CODEWITHSUNDEEP
opencl
Tigran84
Tigran84

Reputation: 193

Can I have boolean buffer in OpenCL and change its value during kernel execution, example to break while loop

I want to do some experiments in OpenCL and I want to know possibility to change states during kernel execution from host code using buffer.

I attempted to alter the state of a while loop in the kernel code by modifying the buffer value from within the host code, however the execution is hung.

void my_kernel(
                __global bool *in,
                __global int *out)
{
        int i = get_global_id(0);
        while(1) {
                if(1 == *in) {
                        printf("while loop is finished");
                        break;
                }
        }
        printf("out[0] = %d\n", out[0]);

}

I call second time the function clEnqueueWriteBuffer() to change state of input value.

                input[0] = 1; 
                err = clEnqueueWriteBuffer(commands, input_buffer,
                                CL_TRUE, 0, sizeof(int), (void*)input,
                                0, NULL,NULL);

Upvotes: 1

Views: 51

Answers (1)

pmdj
pmdj

Reputation: 23438

At least for OpenCL 1.x, this is not permitted, and any behaviour you may observe in one implementation cannot be relied upon.

See the NOTE in the OpenCL 1.2 specification, section 5.2.2, Reading, Writing and Copying Buffer Objects:

Calling clEnqueueWriteBuffer to update the latest bits in a region of the buffer object with the ptr argument value set to host_ptr + offset, where host_ptr is a pointer to the memory region specified when the buffer object being written is created with CL_MEM_USE_HOST_PTR, must meet the following requirements in order to avoid undefined behavior:

  • The host memory region given by (host_ptr + offset, cb) contains the latest bits when the enqueued write command begins execution.
  • The buffer object or memory objects created from this buffer object are not mapped.
  • The buffer object or memory objects created from this buffer object are not used by any command-queue until the write command has finished execution.

The final condition is not met by your code, therefore its behaviour is undefined.

I am not certain if the situation is different with OpenCL 2.x's Shared Virtual Memory (SVM) feature, as I have no practical experience using it, perhaps someone else can contribute an answer for that.

Upvotes: 1

Related Questions