2D Texture Difficulties


I was wondering if anyone has any ideas about how to fix my current little problem. I am trying to save a buffer containing floats between 0 and 1 (it’s a z-buffer, if you must know) and access it on the device using normalized coordinates x and y both as floats between 0 and 1.

Here’s what I’ve got for declaring the CUDA array and a texture (both are global):

//texture element information for the ZBuffer

texture<float, 2, cudaReadModeElementType> zbuffer_texture;

cudaArray* ZBufferArray = 0;

Here’s what I’ve got for populating the z-buffer and binding the texture:

//load the zBuffer from the host to the array

cudaMallocArray(&ZBufferArray, &channelDesc, sizeof(float)*zBufferSizeX, zBufferSizeY);

cudaMemcpyToArray(ZBufferArray, 0, 0, host_zBuffer, sizeof(float)*zBufferSizeX*zBufferSizeY, cudaMemcpyHostToDevice);

//define the texture parameters and bind the texture to the array

zbuffer_texture.normalized = true;

zbuffer_texture.filterMode = cudaFilterModePoint;

zbuffer_texture.addressMode[0] = cudaAddressModeClamp;

zbuffer_texture.addressMode[1] = cudaAddressModeClamp;

cudaBindTextureToArray(zbuffer_texture, ZBufferArray, channelDesc);

Lastly, here is how I access it in the kernel:

float viewX = ( (float) index.x / (float) resolution.x );

float viewY = ( (float) index.y / (float) resolution.y );

float endDepth = tex2D(zbuffer_texture, (1.0f-viewX), viewY );

(Note that the resolution used in the kernel is not the same as the zBufferSizeX and zBufferSizeY used in populating the array, but index.x and index.y are guaranteed to be in [0, resolution.x) and [0, resolution.y) respectively, so viewX and viewY should always be between 0 and 1.)

So, to test this, I populated the host z-buffer all with 1.0f’s, expecting to get 1.0f from endDepth regardless of the value of viewX and viewY. Alas, I was wrong.

Any ideas?


should be

cudaMallocArray(&ZBufferArray, &channelDesc, zBufferSizeX, zBufferSizeY);