cudaMemcpy3D problem

Hi everybody

I’m trying to copy the memory named d_volume to a cudaArray named d_volumeArray but get an “invalid argument” error.

Definitions:
cudaExtent volumeSize = make_cudaExtent(…);
uchar *d_volume; //used for selective data write
cudaArray *d_volumeArray;

Allocation code:
size_t size = volumeSize.widthvolumeSize.heightvolumeSize.depth*sizeof(u
char);
cutilSafeCall( cudaMalloc((void**)&d_volume, size));
channelDesc = cudaCreateChannelDesc();
cutilSafeCall( cudaMalloc3DArray(&d_volumeArray, &channelDesc, volumeSize) );

I have tried two options of copying from d_volume to d_volumeArray:

cutilSafeCall(cudaMemcpyToArray(d_volumeArray, 0, 0, d_volume, size, cudaMemcpyDeviceToDevice));

cudaMemcpy3DParms copyParams = {0};
copyParams.srcPtr = make_cudaPitchedPtr((void*)d_volume, volumeSize.width*sizeof(uchar), volumeSize.width, volumeSize.height);
copyParams.dstArray = d_volumeArray;
copyParams.extent = volumeSize;
copyParams.kind = cudaMemcpyDeviceToDevice;
cutilSafeCall( cudaMemcpy3D(&copyParams) );

In both trials, I got the “invalid argument” runtime message.

Does anybody have any idea how to solve it?

Try allocating device mem using cudaMalloc3D() instead of cudaMalloc(). cudaMalloc3D() returns a pitchedPtr, so may be some differene between internal formats between two allocations!

Did you check to see the error message?

-Oj

Thanks Ojaswa

The only problem I encounter is that using cudaMalloc3D() force me (as you mentioned) to define the memory as cudaPitchedPtr rather than void*.

In such a case, how can I access (write) a specific address within this allocated memory?

Like this (inside a kernel):

__global__ somekernel(cudaPitchedPtr accumPPtr)

  {

	...

	int __x, __y, __z;//Compute global x, y, z coords for the volume.

	...

	float somevalue = 0.0;

	row = (float*)((char*) accumPPtr.ptr + (__z* accumPPtr.ysize + __y)* accumPPtr.pitch);

	row[__x] += somevalue;

  }

Does this answer your question?

-Oj

float* row;