How are Values Loaded into the Texture Memory from Global Memory

Accelerated Computing CUDA CUDA Programming and Performance
1

Hi everyone,

Logistics: I’m using CUDA 2.2, CUDA Programming Guide 2.2.1, GPU = Quadro NVS 295 (Compute Capability 1.1)

This is a related question to the topic I posted a few hours ago that avidday helped me solve. This time, my question revolves around how texture values are loaded into a texture cache? My code (which is attached), is a simple program that just loads an array of values. The array is really a 2D array, indexed by row major order. Each “row” in the host array is set to the row number + 1.

In the code, when I’m in the device, right now I’m simply trying to read the values out that are written to the texture cache. When I run my program, I get the following (for a 3x3 test):

Initial Array

1.000000 1.000000 1.000000
2.000000 2.000000 2.000000
3.000000 3.000000 3.000000

Texture Array

1.000000 1.000000 1.000000
1.500000 1.500000 1.500000
2.500000 2.500000 2.500000

which seems a bit odd to me since I would expect the Texture Array to have the same values as the Initial (Host) Array. I’ve tried searching the forum for an issue like this, and searching online, but I haven’t found anything like this. Does anyone know why this might be happening? My guess is that it probably has something to do with the cudaMemcpyToArray() function, but I can’t find anything wrong with my call.

Thanks for your help,
Matt

2

It appears that my code wasn’t attached, sorry.

Matt
textureCache_ex.cu (5.5 KB)

3

Have a look at the interpolation rules in the back of the programming guide. Textures are nominally texel(cell) centred, so you probably need to centre the coordinates you supply to the texture read.

4

I went and looked at them again in D.2, but it doesn’t seem very clear to me.

I was able to figure out how to center it though, I added 0.5 in both dimensions.

Thanks,
Matt