Adress of 3d arrays

Hey!
I was doing some memory allocation and have some doubts. Using cudaMallocPitch i know i can calculate the adress of an element of a 16X8 matrix in the array by defining an index:

TILE_I=16;
TILE_J=8;

i= blockIdx.xTILE_I+threadIdx.x
j=blockIdx.y
TILE_J+threadIdx.y
I2d=i+j*pitch/sizeof(float)

Now i want to try to do the same thing using cudaMalloc3D, but i don’t know how to come up with an index like I2d for the adress of the 3d matrix. Can anyone Help??
PLEASE!!!

Documented in the NVIDIA programming guide but not under “cudaMalloc3D”… Just search for 3D and find it.

if it can help (because i think that Programming Guide is a bit vague on 3D pitch usage) :

[codebox]

real ***tab; //tab[Nx][Ny][Nz]

[…]

//allocation 3D gpu

cudaPitchedPtr pitchptr_gpu;

cudaExtent extent = make_cudaExtent(Nz*sizeof(real),Ny,Nx);

state=cudaMalloc3D(&pitchptr_gpu, extent);

//copie 3D

cudaMemcpy3DParms cp = { 0 };

cp.srcPtr = make_cudaPitchedPtr((void )&tab[0][0][0], Nzsizeof(real),Nz,Ny);

cp.dstPtr = pitchptr_gpu;

cp.extent = extent;

cp.kind = cudaMemcpyHostToDevice;

state=cudaMemcpy3D(&cpA);

[…]

//lecture

tab[y][z] ~ tab_gpu[ z + y*pitch/sizeof(real) + x Nypitch/sizeof(real) ];

Avec pitch = pitchptr_gpu.pitch et tab_gpu= (real *) pitchptr_gpu.ptr

[/codebox]