cancel
Showing results for 
Search instead for 
Did you mean: 

Archives Discussions

adm271828
Journeyman III

Passing a NULL buffer to a kernel

According to OpenCL 1.1 (§ 5.7.2) it is possible to pass a null buffer to a kernel: "If the argument is a buffer object, the arg_value pointer can be NULL or point to a NULL value in which case a NULL value will be used as the value for the argument declared as a pointer to __global or __constant memory in the kernel."

However, I can't see it working. Here is what I get for various attempts (SDK 2.4).

Regards,

   Antoine

 

err = clSetKernelArg (k, 0, 0, 0); // -> err = CL_INVALID_ARG_VALUE cl_mem buf = 0; err = clSetKernelArg (k, 0, sizeof(cl_mem), &buf); // -> err = CL_INVALID_MEM_OBJECT err = clSetKernelArg (k, 0, 0, &buf); // -> err = CL_INVALID_ARG_SIZE

0 Likes
12 Replies
himanshu_gautam
Grandmaster

Can you please post a test case showing the issue. How do you create the buffer object?

0 Likes

Hi,

the test case is the code already provided. It shall be clear I don't create a buffer, since I want to pass a NULL, and my understanding from the OpenCL spec (quoted in initial message) is that it is sufficient to pass a NULL arg_value, in which case the kernel will receive a null pointer.

Do you suggest I shall create a buffer with zero size and the kernel would receive a null pointer? I didn't try, but if it is the case, it doesn't seem to conform to OpenCL, and it would waste a buffer ressource (in case clCreateBuffer succeeds with null size).

Maybe I missunderstood the spec?

Best regards,

    Antoine

0 Likes

IMHO clSetKernelArg(k, 0, sizeof(cl_mem), 0); is correct way to specify NULL pointer. (but return CL_INVALID_ARG_VALUE) OpenCL specification is little contradiction in this area.

CL_INVALID_ARG_VALUE if arg_value specified is NULL for an argument that is not declared with the __local qualifier or vice-versa.

so you can specify NULL pointer but you get error if do so

0 Likes

I do

size_t localMemBufferSize = xxx;

clSetKernelArg(k,0,localMemBufferSize,NULL);

 

this works on my NVidia cards - I haven't tried it on my ATI yet.

 

I have had problems with local memory like this

 

clSetKernelArg(k,0,sizeof(cl_mem),p1);

clSetKernelArg(k,1,localMemBufferSize,NULL);

when p1 structure wasn't an acceptable multiple of 16 bytes. I haven't resolved this issue except by trial and error with filler fields

0 Likes

clSetKernelArg (k, 0, size, NULL), with size > 0 is for local memory, as explained int the spec. It works on my ATI card.

Once again, I want to pass a null pointer to global or constant memory (as described in the spec...). Seems to be broken (but not critical).

0 Likes

hi adm,

This appears to be a snack in the opencl spec. So it is better to report at khronos forums.

Meanwhile, i was interested if you were having some specific purpose for allocating a NULL buffer or were just experimenting.

Thanks

0 Likes

Hi Himanshu,

I'm not just experimenting. I'd like to pass an optional output buffer to collect extra information from a kernel.

On the kernel side, I'd have written: kernel k (__global int4* out) { ... if (out != 0) { ... write extra info here } ... }.

Of course I can do it by having this optional buffer+code enabled/disabled conditionnaly at compile time. But since I will select each version at runtime in a non predictible pattern, it means I will have two kernels loaded at the same time, wasting resources (the extra code executed when out != 0 is about 10 % of the whole kernel).

This is not critical however (having intrinsics that will help me not to generate 2 instructions when only a single one is necessary is much more critical... for instance)

As for OpenCL spec, as an end-user, I can't say the spec is incorrect (if this is what you mean). Somebody wrote this sentence, that seems to make sense. If during approval process, it was decided to remove it and this was not done, then it's up to people that make products compliant to the spec to say it to the OpenCL committee. Especialy if they are part of the committee.

Best regards,

Antoine

0 Likes

I agree, that part of the specification caught me aswell.

Apart from creating a dummy 'NULL' buffer to bind when that kernel argument is not used, I also passed an int used as a boolean, to indicate whether the kernel should output data into the buffer or not. This way you don't have to create a second kernel.

0 Likes
ajk
Journeyman III

Hello. This issue is still unresolved. clSetKernelArg fails when passing NULL as arg_value.

There was the same bug in Intel OpenCL: https://software.intel.com/en-us/forums/topic/281206 and it was resolved.

0 Likes

Hi,


It would be a great help if you can share a sample code capturing what you want to do and the error you're facing.We will try to reproduce the issue. Please also let us know your system details like SDK, Driver, OS (Window/Linux) (32/64), GPU etc.?


Thanks,

0 Likes
ajk
Journeyman III

I've further studied this case and found that:

suppose, we have the following kernel:

__kernel void test(__global float *a, __global const float *b, __global const float *c) {

  int idx = get_global_id(0);

  a[idx] += b[idx] + (c ? c[idx] : 0);

}

then:

clSetKernelArg(krn, 2, 0, NULL) fails on AMD, succeeds with correct results of the following kernel execution on Intel, succeeds on NVIDIA, but gives incorrect results (NVIDIA just ignores buffer assignment),

while

clSetKernelArg(krn, 2, sizeof(cl_mem), NULL) succeeds on all platforms with correct results, (cl_mem p_null = NULL, clSetKernelArg(krn, 2, sizeof(cl_mem), &p_null) also succeeds).

So, I suppose behaviour of clSetKernelArg with arg_size=0 and arg_value=NULL is vendor specific (the documentation of clSetKernelArg doesn't specify should arg_size be checked or not when arg_value is NULL sadly).

0 Likes

Hi,

Yes, your assumption is correct. To specify a NULL buffer you should use: clSetKernelArg(krn, 2, sizeof(cl_mem), NULL).

As per the clSetKernelArg API spec (clSetKernelArg) :

".....

arg_size: 

     Specifies the size of the argument value. If the argument is a memory object, the size is the size of the buffer or image object type. ...

    Also, it generates error:

    CL_INVALID_ARG_SIZE if arg_size does not match the size of the data type for an argument that is not a memory object or if the argument is a memory object and arg_size != sizeof(cl_mem) or if arg_size is zero and the argument is declared with the __local qualifier or if the argument is a sampler and arg_size != sizeof(cl_sampler).

...."

    So according to API, you can assume that you have to specify the size of the argument type. There is no explicit mention about the arg_size in case of NULL. Some vendor may support 0 as arg_size value in case of NULL. But, for portability, better to use common one.

Regards,

0 Likes