OpenCL OpenGL interop and Element buffer

Question

I created an OpenGL element buffer to draw a set of triangles. When I try to use this buffer in an OpenCL kernel, the values in the array are shifted.

Example: The first two triangles in the original array [1, 2, 3] [1, 3, 4] and when reading from the OpenCL kernel [1, 2, 3] [3, 4, 0]

Seems OpenCL reads the buffer with different element size / stride than OpenGL. But why?

The OpenGL definition of the buffer:

glm::uvec3 triangles;

...

GLuint gl_buffer_id = 0;
glGenBuffers(1, &gl_buffer_id);
glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, gl_buffer_id);
glBufferData(GL_ELEMENT_ARRAY_BUFFER, sizeof(glm::uvec3) * triangles.size(), triangles.data(), GL_STATIC_DRAW);
glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, 0);

Passing the GL buffer to the kernel

cl_int err;
cl_mem cl_buffer = clCreateFromGLBuffer(cl_context, CL_MEM_READ_ONLY, gl_buffer_id, &err);

...

err = clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *)&cl_buffer);

The OpenCL kernel

__kernel void my_kernel(
    __global uchar4 * points_color,
    __global uint3 * triangles
    )
{
int pid = get_global_id(0);
points_color[pid] = (uchar4)(triangles[1].x, triangles[1].y, triangles[1].z, 255); // use vertex id as color component
}

Jan-Gerd Jan-Gerd · Accepted Answer · 2019-08-06T09:00:14

In OpenCL, uint3 is just an alias for uint4, i.e. its size is 4 * sizeof(char). See section 6.1.5. Alignment of Types of the OpenCL specification for more information. You will probably need to use the vload3 instruction.

For 3-component vector data types, the size of the data type is 4 * sizeof(component). This means that a 3-component vector data type will be aligned to a 4 * sizeof(component) boundary. The vload3 and vstore3 built-in functions can be used to read and write, respectively, 3-component vector data types from an array of packed scalar data type.

OpenCL OpenGL interop and Element buffer

1 Answers