通过表面写入 CUDA 中的浮点 OpenGL 纹理

Question

我正在编写 OpenGL/CUDA (6.5) 互操作应用程序。我在尝试通过 CUDA 内核中的表面引用将浮点值写入 OpenGL 纹理时遇到编译时错误。

这里我给出了我如何设置互操作的高级描述，但我在我的 CUDA 内核中成功地读取了我的纹理，所以我相信这是正确完成的。我有一个用

声明的 OpenGL 纹理

glTexImage2D(GL_TEXTURE_RECTANGLE_ARB, 0, GL_RGB32F_ARB, 512, 512, 0, GL_RGB, GL_FLOAT, NULL);

创建纹理后，我调用 cudaGraphicsGLRegisterImage 并设置 cudaGraphicsRegisterFlagsSurfaceLoadStore。在运行我的 CUDA 内核之前，我解除绑定纹理并在从 cudaGraphicsGLRegisterImage 获得的 cudaGraphicsResource 指针上调用 cudaGraphicsMapResources。然后我从 cudaGraphicsSubResourceGetMappedArray 得到一个 cudaArray，为数组创建一个适当的资源描述符，并调用 cudaCreateSurfaceObject 来得到一个指向 cudaSurfaceObject_t 的指针。然后我用 cudaMemcpyHostToDevice 调用 cudaMemcpy 将 cudaSurfaceObject_t 复制到 cudaMalloc 分配的设备上的缓冲区。

在我的 CUDA 内核中，我可以使用类似这样的内容从表面参考中读取，并且我已经验证它按预期工作。

__global__ void cudaKernel(cudaSurfaceObject_t tex) {
    int x = blockIdx.x*blockDim.x + threadIdx.x;
    int y = blockIdx.y*blockDim.y + threadIdx.y;
    float4 sample = surf2Dread<float4>(tex, (int)sizeof(float4)*x, y, cudaBoundaryModeClamp);

在内核中我想修改样本并将其写回纹理。 GPU 具有 5.0 的计算能力，所以这应该是可能的。我正在尝试这个

surf2Dwrite<float4>(sample, tex, (int)sizeof(float4)*x, y, cudaBoundaryModeClamp);

但是我得到错误：

error: no instance of overloaded function "surf2Dwrite" matches the argument list
argument types are: (float4, cudaSurfaceObject_t, int, int, cudaSurfaceBoundaryMode)

我可以在

中看到

cuda-6.5/include/surface_functions.h

只有 surf2Dwrite 的整数版本的原型接受 void * 作为第二个参数。我确实看到 surf2Dwrite 的原型接受带有模板化 surface 对象的 float4，但是，我不确定如何使用 OpenGL 互操作声明模板化 surface 对象.我还没有找到关于如何执行此操作的任何其他内容。任何帮助表示赞赏。谢谢。

Answer 1

原来答案很简单，虽然我不知道为什么会这样。而不是调用

surf2Dwrite<float4>(sample, tex, (int)sizeof(float4)*x, y, cudaBoundaryModeClamp);

我需要打电话

surf2Dwrite(sample, tex, (int)sizeof(float4)*x, y, cudaBoundaryModeClamp);

老实说，我不确定我是否完全理解 CUDA 在 C++ 中使用模板。有人有解释吗？

Answer 2

有关 CUDA 写入链接到 OpenGL 纹理的表面的完整示例，请参阅此项目：

https://github.com/nvpro-samples/gl_cuda_interop_pingpong_st

Answer 3

来自CUDA Documentation，这里是表面模板函数的定义：

template<class T>
T surf2Dread(cudaSurfaceObject_t surfObj,
              int x, int y,
              boundaryMode = cudaBoundaryModeTrap);

template<class T>
void surf2Dread(T* data,
                 cudaSurfaceObject_t surfObj,
                 int x, int y,
                 boundaryMode = cudaBoundaryModeTrap);

通过表面写入 CUDA 中的浮点 OpenGL 纹理

Writing to a floating point OpenGL texture in CUDA via a surface

c++

opengl

interop

textures

cuda