Як скорегувати суму результатів від локальної до глобальної пам'яті в OpenCL - opencl, зменшити

У мене є ядро ​​OpenCL, в якому кожна робоча група виробляє вектор результатів у локальній пам'яті. Тоді мені потрібно підсумувати всі ці результати у глобальній пам'яті для подальшого пошуку до хоста.
Щоб перевірити це, я створив такий код ядра:

//1st thread in each workgroup initializes local buffer
if(get_local_id(0) == 0){
for(i=0; i<HYD_DIM; i++){
pressure_Local[i] = (float2){1.0f, 0.0f};
}
}

//wait for all workgroups to finish accessing any memory
barrier(CLK_GLOBAL_MEM_FENCE | CLK_LOCAL_MEM_FENCE);

/// sum all the results into global storage
for(i=0; i<get_num_groups(0); i++){

//1st thread in each workgroup writes the group"s local buffer to global memory
if(i == get_group_id(0) && get_local_id(0) == 0){
for(j=0; j<HYD_DIM; j++){
pressure_Global[j] += pressure_Local[j];
// barrier(CLK_GLOBAL_MEM_FENCE);
}
}

//flush global memory buffers:
barrier(CLK_GLOBAL_MEM_FENCE);
}

По суті, я очікував усіх елементіввектор у глобальній пам'яті дорівнює кількості робочих груп (у моєму випадку 128). Насправді вони зазвичай варіюються між 60 і 70, а результати змінюються від запуску до запуску.
Чи може хтось мені сказати, що це таке, що я пропускаю, або як це зробити правильно?

Відповіді:

Ви не можете синхронізувати різні роботигрупи з opencl. CLK_GLOBAL_MEM_FENCE не працює таким чином. Це лише гарантує, що порядок операцій з пам'яттю (доступ до яких працює робоча група) буде збережений. Див. Розділ "6.12.8 Функції синхронізації" в розділі OCL 1,2 спец.

Я вирішував би вашу проблему, використовуючи іншийблок глобальної пам'яті для кожної робочої групи. Ви записуєте дані в глобальний і ваше ядро ​​закінчено. Потім, якщо ви хочете зменшити дані до одного блоку, ви можете зробити інше ядро, щоб прочитати дані з глобального, і об'єднати їх з іншими блоками результатів. Ви можете виконати стільки шарів злиття, скільки хочете, але остаточне злиття повинно бути виконано однією робочою групою.

Шукайте алгоритми скорочення gpu / opencl. Ось з цього пристойно почати. Приклад: Прості скорочення