Совместная память OpenCL среди задач

Я работал над созданием программы жизни Конвея на основе GPU. Если вы не знакомы с этим, вот Страница Википедии. Я создал одну версию, которая работает, сохраняя массив значений, где 0 представляет мертвую ячейку, а 1 — живую. Затем ядро ​​просто записывает данные в массив данных буфера изображения, чтобы нарисовать изображение на основе данных ячейки, а затем проверяет соседей каждой ячейки, чтобы обновить массив ячеек для следующего выполнения рендеринга.

Однако более быстрый метод вместо этого представляет значение ячейки как отрицательное число, если оно мертвое, и положительное число, если оно живое. Номер этой ячейки представляет количество соседей, которые у нее есть, плюс один (делая ноль невозможным значением, поскольку мы не можем отличить 0 от -0). Однако это означает, что при создании или уничтожении ячейки мы должны соответственно обновлять значения ее восьми соседей. Таким образом, в отличие от рабочей процедуры, которая должна только читать из соседних слотов памяти, эта процедура должна записывать в эти слоты. Это противоречиво, и выведенный массив недопустим. Например, ячейки содержат числа, такие как 14, что указывает на 13 соседей, невозможное значение. Код правильный, так как я написал ту же процедуру на процессоре, и он работает как положено. После тестирования я считаю, что когда задачи пытаются одновременно выполнить запись в память, возникает задержка, которая приводит к некоторой ошибке записи. Например, возможно, существует задержка между чтением данных массива и настройкой времени изменения данных, что делает процедуру другой задачи некорректной. Я пытался использовать семафоры и барьеры, но только что изучил OpenCL и параллельную обработку и пока не совсем понял их полностью. Ядро выглядит следующим образом.

int wrap(int val, int limit){
int response = val;
if(response<0){response+=limit;}
if(response>=limit){response-=limit;}
return response;
}

__kernel void optimizedModel(
__global uint *output,
int sizeX, int sizeY,
__global uint *colorMap,
__global uint *newCellMap,
__global uint *historyBuffer
)
{
// the x and y coordinates that currently being computed
unsigned int x = get_global_id(0);
unsigned int y = get_global_id(1);

int cellValue = historyBuffer[sizeX*y+x];
int neighborCount = abs(cellValue)-1;
output[y*sizeX+x] = colorMap[cellValue > 0 ? 1 : 0];

if(cellValue > 0){// if alive
if(neighborCount < 2 || neighborCount > 3){
// kill

for(int i=-1; i<2; i++){
for(int j=-1; j<2; j++){
if(i!=0 || j!=0){
int wxc = wrap(x+i, sizeX);
int wyc = wrap(y+j, sizeY);
newCellMap[sizeX*wyc+wxc] -= newCellMap[sizeX*wyc+wxc] > 0 ? 1 : -1;
}
}
}
newCellMap[sizeX*y+x] *= -1;

// end kill
}
}else{
if(neighborCount==3){
// spawn

for(int i=-1; i<2; i++){
for(int j=-1; j<2; j++){
if(i!=0 || j!=0){
int wxc = wrap(x+i, sizeX);
int wyc = wrap(y+j, sizeY);
newCellMap[sizeX*wyc+wxc] += newCellMap[sizeX*wyc+wxc] > 0 ? 1 : -1;
}
}
}
newCellMap[sizeX*y+x] *= -1;

// end spawn
}
}
}

1. Массив выход это данные буфера изображения, используемые для визуализации
   вычисление ядра.
2. SizeX а также SizeY константы — это ширина и высота буфера изображения соответственно.
3. Colormap массив содержит целочисленные значения rgb для черного и белого соответственно, которые используются для правильного изменения значений буфера изображения для отображения цветов.
4. newCellMap массив — это обновленная карта ячеек, вычисляемая после определения рендеринга.
5. historyBuffer старое состояние ячеек в начале вызова ядра Каждый раз, когда выполняется ядро, этот массив обновляется до массива newCellMap.

Дополнительно заворачивать Функция делает пространство тороидальным. Как я мог исправить этот код так, чтобы он работал как положено. И почему глобальная память не обновляется при каждом изменении задачи? Разве это не должна быть общая память?