clSerKernelArg and __global qualified

If one wishes to use __local qualified arguments they should be set using the size in bytes and a NULL pointer.

What if one wishes to do something similar for __global qualified arguments that are not read from or written to host memory, i.e., a buffer analogous to __local but instead __global. Would this be possible and called using the size in bytes and a NULL pointer, or must one necessarily create memory objects?

Obviously this doesn‘t make sense for __constant qualified arguments.

I think I found the answer to my question in the specification, section 5.7.2; however it says that it‘s also possible for __constant qualified arguments. Can anyone provide an example for such a use of __constant?

Would this be possible and called using the size in bytes and a NULL pointer, or must one necessarily create memory objects?

No, it’s not possible. You must create a buffer.

Section 5.7.2 is not saying that with __constant pointers you can do the same as with __local pointers. __constant pointers behave as __global pointers.

Hi David,

Would you mind explaining the meaning of the last paragraph on the first page of Section 5.7.2: “If the argument is a memory object …“

In that paragraph it describes __global and __constant arguments (using “can“) in a similar manner as __local arguments (using “must“) with respect to NULL or pointer to NULL arg_value. This is the place where my understanding seems to get cloudy.

Thanks for your help!

Ah, I see what you mean. All that paragraph says is that if you have a kernel like this:


__kernel void foo(__global float *bar)
{
...
}

Then you can do this:

cl_int errcode = clSetKernelArg(myKernel, 0, sizeof(cl_mem), &myBuffer);

And the semantics are the same as passing a NULL pointer to function foo(). In other words, “bar” will be NULL. There’s no automatic memory allocation like with __local pointers.

Thanks, that clears things up. So if you have a kernel like this:


__kernel void foo(__global float *bar)
{
...
}

Then you try to do this:

cl_int errcode = clSetKernelArg(myKernel, 0, n * sizeof(cl_float), NULL);

It should return the error code CL_INVALID_ARG_SIZE or CL_INVALID_ARG_VALUE?

It should return the error code CL_INVALID_ARG_SIZE or CL_INVALID_ARG_VALUE?

Short answer: yes.

Long pedantic answer: Assuming n is large it should return CL_INVALID_ARG_SIZE. If “n” was equal to one or two, then “n * sizeof(cl_float)” may be equal to sizeof(cl_mem) and in that case no error would be returned. Instead, it would have the effect of making “bar” equal to NULL.