tex3D axis order problem

Accelerated Computing CUDA CUDA Programming and Performance
1

Hello !

I hope you will be able to help me as this is driving me crazy for a few days.

Our array is a bit unusual as our main axis is Z, so to get index into array using x,y,z coordinates you calc it

as follows: array[ xnynz + y*nz + z ]

nx - number of elements in x direction, etc…

Here is an example how did I try to overcome this unusual ordering;

cudaArray *dev_vx(NULL), *dev_vy(NULL), *dev_vz(NULL); 

	cudaExtent const ext = {nz, ny, nx}; 

	cudaChannelFormatDesc floatTex_vx2 = cudaCreateChannelDesc<float>(); 

	cudaChannelFormatDesc floatTex_vy2 = cudaCreateChannelDesc<float>(); 

	cudaChannelFormatDesc floatTex_vz2 = cudaCreateChannelDesc<float>(); 

	

	cudaMalloc3DArray(&dev_vx, &floatTex_vx2, ext);

	cudaMalloc3DArray(&dev_vy, &floatTex_vy2, ext);

	cudaMalloc3DArray(&dev_vz, &floatTex_vz2, ext);

	 cudaMemcpy3DParms copyParams = {0}; 

	 copyParams.extent = make_cudaExtent(nz, ny, nx); 

	 copyParams.kind = cudaMemcpyHostToDevice; 

	 copyParams.dstArray = dev_vx; 

	 copyParams.srcPtr =  make_cudaPitchedPtr((void*)&vx1[0], nz*sizeof(float), nz, ny); 

	 cudaMemcpy3D(&copyParams);	

	 	 

	 copyParams.dstArray = dev_vy; 

	 copyParams.srcPtr = make_cudaPitchedPtr((void*)&vy1[0], nz*sizeof(float), nz, ny); 

	 cudaMemcpy3D(&copyParams);	

	 copyParams.dstArray = dev_vz; 

	 copyParams.srcPtr = make_cudaPitchedP

..and for accessing: val = tex3D(tex_vx, _z, _y, _x);

All this works ok when nx=ny=nz, but as soon as nx!=ny!=nz everything falls apart.

Thank you for your help.

Shokre