Is there a way to stop the OpenCL kernel from working?

Question

Is it possible to stop the launch of the OpenCL kernel? For example, I start the kernel, do some calculations, and then stop it if some conditions are met, otherwise I wait until it ends:

clEnqueueNDRange(queue, ...); // start kernel function // do other stuff... // ... if (some condition met) { stopKernel(); } else { clFinish(queue); } 

thanks for the help

Answer 1

Not. After you close the kernel, it will be completed.

One way to achieve something similar above:

 while ( data_left_to_process ) { clEnqueueNDRangeKernel( ..., NDRange for a portion of the data, ... ) // other work if (condition) { break; } // adjust NDRange for next execution to processes the next part of your data } clFinish(queue); 

This avoids the processing of ALL data with the obvious compromise that you are sending now, working in small pieces, which is likely to have an impact on performance.

Answer 2

Maybe.

• Create two command queues in context.
• Create two cores, one to do the work, and the other to stop the execution. Each core has access to a global global buffer.
• Load the first kernel in queue1.
• Load the second core in queue2 if you want to stop execution.

Alternatively, you can use the queue out of turn and load the second core into the same command queue to stop execution. You should be a little more careful (using clFinish / clFlush if necessary), however this is a more natural way to do this.

Some pseudo codes (for multiple queues):

 clEnqueueNDRange(queue1, kernel1, ...); //start work kernel // do other stuff // ... if (condition) clEnqueueNDRange(queue2, kernel2, ...); //stop work kernel clFinish(queue1); clFinish(queue2); // entirely unnecessary if using in-order queues 

Use the ints or float buffer as a stop variable and access them through global_id in your kernels to reduce the cost of reading from the global in a loop. The disadvantage is that your state will be undefined: without additional variables for counting performances, etc. You will not know how many work items and which ones were completed.

And the kernels:

 void kernel1( ... ,global int * g_stop) { int index_from_ids = ...; while (g_stop[index_from_ids] == 0) // or 'if' for single pass execution { // do work here } } void kernel2( ... ,global int * g_stop) { int index_from_ids = ...; g_stop[index_from_ids] = 1; }