Есть ли способ остановить запуск ядра OpenCL?

Question

Есть ли способ остановить запуск ядра OpenCL?Например, я запускаю ядро, выполняю некоторые вычисления и затем останавливаю его, если выполняются некоторые условия, в противном случае я жду, пока оно не завершится:

clEnqueueNDRange(queue, ...); // start kernel function

// do other stuff...
// ...

if (some condition met) {
    stopKernel();
} else { 
    clFinish(queue);
}

Спасибо за помощь

Answer 1

Нет. Как только вы поставили ядро ​​в очередь, оно будет работать до конца.

Один из способов достижения чего-то , как указано выше, заключается в следующем:

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);

Это позволяет вам избежать обработки ВСЕХ данных с очевидным компромиссом в том, что вы сейчас отправляете работу небольшими порциями, что, вероятно, скажется на производительности.

Answer 2

Возможно.

1. Создайте две очереди команд в контексте.
2. Создайте два ядра, одно для выполнения работы, а другое для остановки выполнения.Каждое ядро ​​имеет доступ к общему глобальному буферу.
3. Загрузка первого ядра в очередь 1.
4. Загрузка второго ядра в очередь2, если вы хотите остановить выполнение.

В качестве альтернативы вы можете использовать очередь с ошибками и загрузить второе ядро ​​в ту же очередь команд, чтобы остановить выполнение.Вы должны быть немного более осторожны (при необходимости используйте clFinish / clFlush), однако это более естественный способ сделать это.

Некоторый псевдокод (для нескольких очередей):

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

Используйте буфер с целочисленными значениями или числами с плавающей запятой в качестве переменной остановки и обращайтесь к ним через global_id в ваших ядрах, чтобы снизить стоимость чтения из глобального цикла.Недостатком является то, что ваше состояние будет неопределенным: без дополнительных переменных для подсчета выполнений и т. Д. Вы не будете знать, сколько рабочих элементов и какие были выполнены.

А ядра:

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;
}