Cannot set OpenCL kernel argument with buffer memory object

Question

I have the following simple OpenCL kernel, that simply copies all entries pointed at a to b

__kernel void mmcopy(__global float* a, __global float* b) {
    unsigned pos = get_global_id(0);
    b[pos] = a[pos];
}

The following code snippet shows the opencl function calls for creating a buffer memory object out of four floats, and setting the first argument on the kernel with the buffer object.

let mut v = [1f32, 1f32, 1f32, 1f32];

let size = mem::size_of_val(&v) as size_t;
let mut error_buffer = 0 as i32;
let buffer = unsafe {
    clCreateBuffer(
        context.id.unwrap(),
        (CL_MEM_COPY_HOST_PTR | CL_MEM_READ_WRITE) as u64,
        size,
        v.as_mut_ptr() as *mut c_void,
        &mut error_buffer,
    )
};

let real_size = mem::size_of::<cl_mem>() as size_t;

let error = unsafe {
    clSetKernelArg(
        self.id.unwrap(), // here `self` is a wrapper. `id` is of type `cl_kernel`
        0 as cl_uint,
        real_size,
        buffer as *const c_void,
    )
};

However, executing the code results in an error CL_INVALID_MEM_OBJECT. it looks like creating the buffer didn't succeed, but returned without an error.

The spec is also not very precise when it comes to describe the error in more detail:

for an argument declared to be a memory object when the specified arg_value is not a valid memory object.

note: the OpenCL functions, and types have been generated by rust-bindgen.

update 1

To clarify how the opaque types are represented in rust, here is the representation of cl_mem,

pub struct _cl_mem {
    _unused: [u8; 0],
}
pub type cl_mem = *mut _cl_mem;

the ffi to clSetKernelArg

extern "C" {
    pub fn clSetKernelArg(
        kernel: cl_kernel, 
        arg_index: cl_uint,
        arg_size: size_t,
        arg_value: *const ::std::os::raw::c_void,
    ) -> cl_int;
}

and clCreateBuffer

extern "C" {
    pub fn clCreateBuffer(
        context: cl_context,
        flags: cl_mem_flags,
        size: size_t,
        host_ptr: *mut ::std::os::raw::c_void,
        errcode_ret: *mut cl_int,
    ) -> cl_mem;
}

In my understanding rust(-bindgen) uses zero sized types (ZST) to represent external opaque types. So basically cl_mem is already a pointer.

update 2

According to pmdj's answer the correct way is to pass a pointer to the cl_mem buffer

let error = unsafe {
    clSetKernelArg(
        self.id.unwrap(), // here `self` is a wrapper. `id` is of type `cl_kernel`
        0 as cl_uint,
        real_size,
        &buffer as *const _ as *const c_void,
    )
};

That actually fixes the problem, and set the return value to CL_SUCCESS. The spec for clSetKernelArg also mentions a pointer to data

A pointer to data that should be used as the argument value for argument specified by arg_index. The argument data pointed to by arg_value is copied and the arg_value pointer can therefore be reused by the application after clSetKernelArg returns. The argument value specified is the value used by all API calls that enqueue kernel (clEnqueueNDRangeKernel) until the argument value is changed by a call to clSetKernelArg for kernel [...]

pmdj pmdj · Accepted Answer · 2020-04-20T07:40:15

Before I dig in, I'll point out that I'm a relative beginner with Rust and I'm not particularly familiar with what bindgen produces, but I know OpenCL quite well. So please bear with me if my Rust syntax is off.

The most obvious thing that sticks out for me is that passing the buffer to clSetKernelArg using buffer as *const c_void looks suspicious. My understanding is that your code is roughly equivalent to this C:

cl_int error_buffer = 0;
cl_mem buffer = clCreateBuffer(
        context.id,
        (CL_MEM_COPY_HOST_PTR | CL_MEM_READ_WRITE),
        size,
        v,
        &error_buffer
    );

size_t real_size = siezof(buffer);
cl_int error = clSetKernelArg(self.id, 0, real_size, buffer);

However, the last line is incorrect, it should be:

cl_int error = clSetKernelArg(self.id, 0, real_size, &buffer);
// yes, we want a POINTER to the buffer handle-------^

Although cl_mem is defined as a pointer to an opaque struct type, you need to pass the pointer to that pointer as the argument, just as with any other type of kernel argument: conceptually, I find it useful to think of it as clSetKernelArg performing a memcpy(internal_buffer, arg_value, arg_size); internally - so arg_size must always be the size of the object pointed to by arg_value. I find this helps me work out the correct level of indirection.

So in Rust this is probably along the lines of:

let error = unsafe {
    clSetKernelArg(
        self.id.unwrap(),
        0 as cl_uint,
        real_size,
        &buffer as *const c_void,
    )
};

but I haven't run it past rustc so it's probably wrong. You get the drift though.

Cannot set OpenCL kernel argument with buffer memory object

update 1

update 2

1 Answers