binding texture with linear memory

So it says the size is the memory in Bytes for the texture,

but if I want to use a 2D texture, how can I specify the dimention of the array ?

so here is my code :

float *data;

uint32_t data_size;

CUDA_SAFE_CALL(cudaMalloc((void**) &(data), data_size));

CUDA_SAFE_CALL(cudaMemcpy(data, host_data,data_size,

       cudaMemcpyHostToDevice) );

texture<float4, 2, cudaReadModeElementType> texRef;

 // set texture parameters

  texRef.addressMode[0] = cudaAddressModeClamp;

  texRef.addressMode[1] = cudaAddressModeClamp;

  texRef.filterMode = cudaFilterModePoint;

  texRef.normalized = false;    // access with normalized texture coordinates or not


  cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc(32, 32,32, 32, cudaChannelFormatKindFloat);

  CUDA_SAFE_CALL(cudaBindTexture(texRef,(void *)data,channelDesc,data_size,0));

__global__ void 

myKer(float* in,float* out)


  float4 fetch = texfetch(texRef, x,y); <- this doesn't work !

  out[(threadIdx.y*blockDim.x+threadIdx.x)*4] = fetch.x;

  out[(threadIdx.y*blockDim.x+threadIdx.x)*4+1] = fetch.y;

  out[(threadIdx.y*blockDim.x+threadIdx.x)*4+2] = fetch.z;

  out[(threadIdx.y*blockDim.x+threadIdx.x)*4+3] = fetch.w;


when I do a fetch in the kernel I can only get the first element no matter what x & y values are. I’m guessing it’s because it doesn’t know what is the size of each dimension.

So If I want to use a 2D texture do I have to use a cudaArray with necessarily ? or could I tell to the texture what is my array shape ?

Yes. See Section 4.3.4 of the programming guide: Textures allocated in linear memory can only be of dimensionality equal to 1.