Really strange gl texture write bug

I’m pretty new to OpenCL and I encountered a really strange error. I have a bunch of OpenGL base code so I implemented gl/cl interop. I have a really simple program, which acquires a texture as a cl image object, and then outputs a color. I then simply render the texture to a screen space quad. I got this working. However, the program I am implementing requires that I use floating point image data. So, all I did was change the type of the opengl texture in my program init function from GL_RGBA8 to GL_RGBA16F_ARB. This caused the kernel to produce the correct output, but only covering the left half of the image!

My really simple kernel is the following:

__kernel void hello( __write_only image2d_t writebuffer

    float4 value = (float4)( (float)get_global_id(0) / (float)get_global_size(0), (float)get_global_id(1) / (float)get_global_size(1), 0.0f, 1.0f);

    write_imagef(writebuffer, (int2)(get_global_id(0), get_global_id(1)), value);


The color is scaled based on the thread id divided by the number of threads, and halfway across the image the red channel is fully bright, as if the program were only queuing up half the necessary threads. Yet, if I change the glTexImage2D call back to use GL_RGB8 it is fine. If I multiply my thread count variable for dimension 0 before calling clEnqueueNDRangeKernel by 2 I get the correct output when using a floating point texture!

I would greatly appreciate any help anyone could provide in this matter. I can provide further details if necessary.


I suggest filing a bug on the vendor who provided the OpenCL drivers you are using. Who is it? Apple, AMD, Intel, NVidia…?