3D texture problem


I’m new to cuda and I’m trying to use a 3D texture to manipulate data but I can’t seem to get it to work.

I can compile the code but all the outputs just turn out to be 0.

I tried this on a 2D case and it worked.

I believe it’s something wrong with copying the data.

#include <cuda.h>

#include <cutil.h>

#include <iostream>

#define max_H 3

#define max_W 3

#define max_D 3

using namespace std;

texture<float, 3, cudaReadModeElementType> tex;

__global__ void texTest(float* matrix_p)


	int index_x = blockIdx.x*blockDim.x+threadIdx.x;

	int index_y = blockIdx.y*blockDim.y+threadIdx.y;

	int index_z = blockIdx.z*blockDim.z+threadIdx.z;



int main (void)


	float matrix [max_H][max_W][max_D];

	float *matrix_ptr;

	float matrix2 [max_H][max_W][max_D];

	for (int i=0; i < max_H ; i++)


		for (int j=0; j<max_W ; j++)


			for(int k=0; k<max_D;k++)


				matrix[i][j][k]=i*max_W*max_H+j*max_H + k;

				cout<<matrix[i][j][k]<<" ";




		cout << endl;


	cout <<"\n\n\n"<< endl;

	//array init

	cudaArray* cu_array=NULL;

	cudaChannelFormatDesc description = cudaCreateChannelDesc<float>(); 

	const cudaExtent extent = make_cudaExtent(max_W, max_H, max_D);

	cudaMalloc3DArray(&cu_array, &description,extent);

	//copy data

	cudaPitchedPtr pitch_ptr;

	cudaMalloc3D(&pitch_ptr, extent);

	cudaMemset3D(pitch_ptr, 0, extent);

	cudaMemcpy3DParms copyParams = {0};

	copyParams.srcPtr   = pitch_ptr;

	copyParams.dstArray = cu_array;

	copyParams.extent   = extent;

	copyParams.kind      = cudaMemcpyDeviceToDevice;


	//texture parameters


    tex.addressMode[0] = cudaAddressModeClamp;

    tex.addressMode[1] = cudaAddressModeClamp;

	tex.addressMode[2] = cudaAddressModeClamp;

    tex.filterMode = cudaFilterModePoint;

    tex.normalized = false; // do not normalize coordinates


	//Binding texture to data

	cudaBindTextureToArray(tex, cu_array,description);


	//Kernal call to manipulate data


	dim3 block(max_W,max_H,max_D);


	//Copying data from device back to host


	CUDA_SAFE_CALL( cudaUnbindTexture(tex) );

	CUDA_SAFE_CALL( cudaFreeArray(cu_array) );

	for (int i=0; i < max_H ; i++)


		for (int j=0; j<max_W ; j++)


			for (int k=0; k<max_D ; k++)


				cout<<matrix2[i][j][k]<<" ";


			cout << endl;


		cout << endl;


	return 0;