了解 CUDA 中的纹理获取答案

【问题标题】：Understanding texture fetching in CUDA了解 CUDA 中的纹理获取
【发布时间】：2016-07-18 17:07:01
【问题描述】：

我按照 cuda 内核执行，以了解纹理获取机制，但看起来数组中的第 A[i][j] 个条目应该由 tex2D(tex_ref,j,i) 获取这不是违反直觉吗？我这里有什么遗漏吗？

 texture<float, cudaTextureType2D, cudaReadModeElementType> texRef;
 surface<void,cudaSurfaceType2D> surfRef;

__global__ void transformKernel(float device_array[3][3],
                                int width, int height,
                                float theta)
{
    unsigned int x = blockIdx.x * blockDim.x + threadIdx.x;
    unsigned int y = blockIdx.y * blockDim.y + threadIdx.y;
    printf("\n  Array : %d %d = %f",x,y,device_array[x][y]);
    printf("\n  Texture : %d %d = %f",x,y,tex2D(texRef,x,y));
    float temp;
    surf2Dread(&temp, surfRef, x*sizeof(float),y);
    printf("\n  Surface : %d %d = %f",x,y,temp);

}

通过将 host_array[i][j] 设置为 (i-j) 来填充数组，并复制到 device_array 并且相同的数组绑定到 texRef,surfRef。

输出看起来像：

数组：0 0 = 0.000000

数组：1 0 = 1.000000

数组：0 1 = -1.000000

数组：1 1 = 0.000000

纹理：0 0 = 0.000000

纹理：1 0 = -1.000000

纹理：0 1 = 1.000000

纹理：1 1 = 0.000000

从表面读取返回的值与获取纹理后返回的值相似。

【问题讨论】：

标签： cuda

【解决方案1】：

您对此感到困惑：

printf("\n  Array : %d %d = %f",x,y,device_array[x][y]);
                                                 ^^^^

x 纹理维度是 2D 纹理操作中的水平（即宽度）维度。这意味着它有效地选择了数据数组中的列。但是，在您的 printf 语句中，您使用 x 来选择行。

您可能希望阅读the documentation，它实际上给出了 2D 纹理的示例（请注意，“纹理获取”与“纹理”是一个不同的术语，但我们可以在这里忽略这种区别）。在此处的 2D 纹理示例中，您将看到“宽度”维度索引首先传递（即作为 x），然后高度维度索引传递（即作为 y）。

documentation还表示在多维纹理操作中纹理维度的顺序是width、height、depth。 p>

您应该能够在打印输出上通过反转索引来创建可比较的结果：

printf("\n  Array : %d %d = %f",x,y,device_array[y][x]);
                                                 ^^^^

【讨论】：