поведение atomic_add в opencl

Question

Я играю с примером на opencl:

__kernel void atomic(__global int* x) {
   __local int a, b;
   a = 0; b = 0;
   a++;
   atomic_inc(&b);

   x[0] = a;
   x[1] = b;
   x[2]++;
   atomic_inc(x+3);
  }

Запуск этого кода с global_size = 1024 и workgroup_size = 8, это следующий вывод:

[1 8 1 1024]

Я могу понять, что происходит во всех случаях, кроме значения, указанного для x[1]. Почему значение x[1] не 1024, а 8?

Answer 1

Под x[1] хранится значение b, которое является переменной, находящейся в адресном пространстве __local, что означает, что переменная является общей для всех рабочих элементов в рабочей группе. Каждая рабочая группа имеет b, инициализированную до 0 и атомарно увеличенную до 8, потому что размер рабочей группы равен 8 (каждый рабочий элемент увеличивается на 1).