Code refactor: avoid some unnecessary device memory copying.

5 files changed, 25 insertions, 34 deletions

diff --git a/intern/cycles/device/device_cpu.cpp b/intern/cycles/device/device_cpu.cpp
index af1bbc0db18..0ba00da16a6 100644
--- a/intern/cycles/device/device_cpu.cpp
+++ b/intern/cycles/device/device_cpu.cpp
@@ -379,7 +379,7 @@ public:
 				texture_info.resize(flat_slot + 128);
 			}
 
-			TextureInfo& info = texture_info.get_data()[flat_slot];
+			TextureInfo& info = texture_info[flat_slot];
 			info.data = (uint64_t)mem.data_pointer;
 			info.cl_buffer = 0;
 			info.interpolation = interpolation;
diff --git a/intern/cycles/device/device_cuda.cpp b/intern/cycles/device/device_cuda.cpp
index c245d7d8408..0f17b67c8c6 100644
--- a/intern/cycles/device/device_cuda.cpp
+++ b/intern/cycles/device/device_cuda.cpp
@@ -840,7 +840,7 @@ public:
 				}
 
 				/* Set Mapping and tag that we need to (re-)upload to device */
-				TextureInfo& info = texture_info.get_data()[flat_slot];
+				TextureInfo& info = texture_info[flat_slot];
 				info.data = (uint64_t)tex;
 				info.cl_buffer = 0;
 				info.interpolation = interpolation;
@@ -1911,9 +1911,10 @@ uint64_t CUDASplitKernel::state_buffer_size(device_memory& /*kg*/, device_memory
 	                           0, 0, (void**)&args, 0));
 
 	device->mem_copy_from(size_buffer, 0, 1, 1, sizeof(uint64_t));
+	size_t size = size_buffer[0];
 	device->mem_free(size_buffer);
 
-	return *size_buffer.get_data();
+	return size;
 }
 
 bool CUDASplitKernel::enqueue_split_kernel_data_init(const KernelDimensions& dim,
diff --git a/intern/cycles/device/device_memory.h b/intern/cycles/device/device_memory.h
index 20707ad04c9..eeeca61496e 100644
--- a/intern/cycles/device/device_memory.h
+++ b/intern/cycles/device/device_memory.h
@@ -270,31 +270,14 @@ public:
 		return &data[0];
 	}
 
-	T *copy(T *ptr, size_t width, size_t height = 0, size_t depth = 0)
+	void steal_data(array<T>& from)
 	{
-		T *mem = resize(width, height, depth);
-		if(mem != NULL) {
-			memcpy(mem, ptr, memory_size());
-		}
-		return mem;
-	}
-
-	void copy_at(T *ptr, size_t offset, size_t size)
-	{
-		if(size > 0) {
-			size_t mem_size = size*data_elements*datatype_size(data_type);
-			memcpy(&data[0] + offset, ptr, mem_size);
-		}
-	}
-
-	void reference(T *ptr, size_t width, size_t height = 0, size_t depth = 0)
-	{
-		data.clear();
-		data_size = width * ((height == 0)? 1: height) * ((depth == 0)? 1: depth);
-		data_pointer = (device_ptr)ptr;
-		data_width = width;
-		data_height = height;
-		data_depth = depth;
+		data.steal_data(from);
+		data_size = data.size();
+		data_pointer = (data_size)? (device_ptr)&data[0]: 0;
+		data_width = data_size;
+		data_height = 0;
+		data_depth = 0;
 	}
 
 	void clear()
@@ -318,6 +301,11 @@ public:
 		return &data[0];
 	}
 
+	T& operator[](size_t i)
+	{
+		return data[i];
+	}
+
 private:
 	array<T> data;
 };
diff --git a/intern/cycles/device/opencl/opencl_base.cpp b/intern/cycles/device/opencl/opencl_base.cpp
index 6747a8a83ac..48c32a9dc5c 100644
--- a/intern/cycles/device/opencl/opencl_base.cpp
+++ b/intern/cycles/device/opencl/opencl_base.cpp
@@ -494,20 +494,21 @@ void OpenCLDeviceBase::mem_free_sub_ptr(device_ptr device_pointer)
 void OpenCLDeviceBase::const_copy_to(const char *name, void *host, size_t size)
 {
 	ConstMemMap::iterator i = const_mem_map.find(name);
+	device_vector<uchar> *data;
 
 	if(i == const_mem_map.end()) {
-		device_vector<uchar> *data = new device_vector<uchar>();
-		data->copy((uchar*)host, size);
+		data = new device_vector<uchar>();
+		data->resize(size);
 
 		mem_alloc(name, *data, MEM_READ_ONLY);
-		i = const_mem_map.insert(ConstMemMap::value_type(name, data)).first;
+		const_mem_map.insert(ConstMemMap::value_type(name, data));
 	}
 	else {
-		device_vector<uchar> *data = i->second;
-		data->copy((uchar*)host, size);
+		data = i->second;
 	}
 
-	mem_copy_to(*i->second);
+	memcpy(data->get_data(), host, size);
+	mem_copy_to(*data);
 }
 
 void OpenCLDeviceBase::tex_alloc(const char *name,
diff --git a/intern/cycles/device/opencl/opencl_split.cpp b/intern/cycles/device/opencl/opencl_split.cpp
index b4e9419ebbd..920106f92d4 100644
--- a/intern/cycles/device/opencl/opencl_split.cpp
+++ b/intern/cycles/device/opencl/opencl_split.cpp
@@ -309,6 +309,7 @@ public:
 		device->opencl_assert_err(device->ciErr, "clEnqueueNDRangeKernel");
 
 		device->mem_copy_from(size_buffer, 0, 1, 1, sizeof(uint64_t));
+		size_t size = size_buffer[0];
 		device->mem_free(size_buffer);
 
 		if(device->ciErr != CL_SUCCESS) {
@@ -318,7 +319,7 @@ public:
 			return 0;
 		}
 
-		return *size_buffer.get_data();
+		return size;
 	}
 
 	virtual bool enqueue_split_kernel_data_init(const KernelDimensions& dim,