传递给OpenCL内核时,image2d_t损坏

4
我正在使用Haskell和OpenCL编写路径追踪器,但是在将image2d_t传递到我的内核以写入输出时遇到了问题。具体来说,在image2d_t上调用OpenCL中的任何get_image_*函数都会返回无意义的值(通常为0或2^24-1),而write_imagef则不起作用。这仅在GPU上运行时发生--CPU可以正常运行。在主机上调用clGetImageInfo返回正确的值。OpenCL的Haskell绑定将错误代码转换为异常,因此不是忘记检查错误的问题。clinfo将我的版本报告为“OpenCL 1.2 AMD-APP(1084.2)”。值得注意的是,我遇到过(并报告过)多个导致OpenCL编译器崩溃或链接失败的错误,因此这可能是由于这些错误而不是我的代码中的错误导致的。
我像这样初始化OpenCL(希望对不了解Haskell的人相对易懂):
(platform:_) <- clGetPlatformIDs
(device:_) <- clGetDeviceIDs platform CL_DEVICE_TYPE_GPU
glContext <- glXGetCurrentContext
glDisplay <- glXGetCurrentDisplay
context <- clCreateContext [CL_GL_CONTEXT_KHR glContext, CL_GLX_DISPLAY_KHR glDisplay] [device] putStrLn
queue <- clCreateCommandQueue context device []
source <- readFile "pt.cl"
program <- clCreateProgramWithSource context source
clBuildProgram program [device] "-cl-strict-aliasing"
        `catch` (λe -> case (e :: CLError) of
                            CL_BUILD_PROGRAM_FAILURE -> putStrLn "Building OpenCL program failed:"
                                                     >> clGetProgramBuildLog program device >>= putStrLn
                                                     >> throw e
                            _                        -> return ())
kernel <- clCreateKernel program "sample"
pCorners <- mallocArray 4
buffer <- clCreateBuffer context [CL_MEM_READ_ONLY, CL_MEM_USE_HOST_PTR] (4*sizeOf (undefined :: V.Vec4F), castPtr pCorners)
clSetKernelArgSto kernel 1 buffer
tex@(TextureObject texid) <- head <$> (genObjectNames 1)
activeTexture $= TextureUnit 0
textureBinding Texture2D $= Just tex
textureFilter Texture2D $= ((Nearest, Nothing), Nearest)
textureWrapMode Texture2D S $= (Repeated, Clamp)
textureWrapMode Texture2D T $= (Repeated, Clamp)
texImage2D Nothing NoProxy 0 RGBA′ (TextureSize2D initialWidth initialHeight) 0 (PixelData RGBA UnsignedByte nullPtr)
image <- clCreateFromGLTexture2D context [CL_MEM_READ_WRITE] gl_TEXTURE_2D 0 texid
clSetKernelArgSto kernel 2 image

我将这个(稍微简化过的)称之为运行内核并呈现结果的过程:
clSetKernelArgSto kernel 0 position
pokeArray pCorners orientedCorners -- update the pCorners array
finish -- This is glFinish()
clEnqueueAcquireGLObjects queue [image] []
clEnqueueNDRangeKernel queue kernel [width, height] [] []
clEnqueueReleaseGLObjects queue [image] []
clFinish queue
drawElements TriangleFan 4 UnsignedInt offset0
swapBuffers

最后,测试内核:
__kernel void sample(float3 position, __constant float3 corner[4], image2d_t output) {
        write_imagef(output, (int2)(get_global_id(0), get_global_id(1)), (float4)(0, 0.5f, 1, 1));
}

这将输出一个全屏的四边形,显示GPU内存中随机未初始化的区域。它本应该是一个全屏的青色四边形。我在其中添加了一些printf来显示get_image_*函数的结果,但它们导致程序挂起。
2个回答

1

OpenCL规范对此有规定 - image2d_t对象需要访问限定符。

有两种这样的限定符:

  • read_only(或__read_only
  • write_only(或__write_only

它们是互斥的,不能同时使用(因此您不能同时读写纹理 - 如果您打算使用像路径跟踪这样的Monte-Carlo应用程序进行累积工作,则这很重要)。省略限定符是有效的,因为它将默认为read_only,但不幸的是,这是输出图像的错误选择。

解决方案就是只需使用write_only限定符来限定您的图像参数,或者如果您也需要从中读取,则使用某种交换系统(或使用全局内存缓冲区,可以同时读取和写入,但这使得CL / GL互操作更加困难,并且您会失去采样功能...)。

我认为CPU上可以这样操作的原因是因为CPU上没有“只读”纹理内存,因此即使写入图像在技术上是非法的,但是仍然有可能,并且运行时允许您这样做。另一方面,GPU具有只读内存部分,在内核正在运行时无论你如何努力都无法对其进行写入(或者也许您的GPU设备的运行时更严格)。
当我提到运行时时,我指的是支持OpenCL的设备,而不是您的程序,当然。

谢谢,我现在已经添加了一个__write_only限定符。不幸的是,这没有产生任何影响。 - klkblake
@klkblake 哎呀,那真是奇怪。可能还有其他问题。你的GPU平台/设备/驱动程序是什么?说实话,我也曾经遇到过GPU内核的瞬态问题,它们会在一个驱动程序版本到另一个版本时突然失败,而没有任何原因。 - Thomas
我使用的是Mobility Radeon 5400系列(我不知道确切的型号)。我在Linux上使用12.11 beta驱动程序,其中包含最新的APP SDK。正如我所提到的,当构建内核时我遇到了一些段错误(我已经向AMD报告了),我开始认为这可能只是一个单一的错误导致了两个问题。 - klkblake
1
@klkblake你尝试过回退到一些旧的驱动程序版本,比如12.8或12.9,并尝试运行代码吗?如果可以运行,那肯定是驱动程序的bug。 - Thomas
我现在已经尝试了从12.6到现在的每个版本(早期版本与Linux 3.6不兼容)。12.10表现出与当前版本相同的行为。所有其他版本都无法运行它--要么是segfault或打印“链接失败”。我现在相当确定这是一个驱动程序错误。 - klkblake

0

我曾经遇到过类似的问题。在重新排列内核参数时,将所有image2d_t参数作为第一个参数,这样可以解决问题。特别是调用get_image_dim返回了正确的结果。不知道这是否是一个bug。我的GPU是ATI Radeon 7950。


我在使用AMD的OpenCL时遇到了一些问题,后来通过按照参数大小降序排列解决了。因此我认为这可能是一个驱动程序的错误(或者曾经是,我最近没有尝试过)。 - klkblake

网页内容由stack overflow 提供, 点击上面的
可以查看英文原文,
原文链接