Problems with local atomics

Discussion created by SiegeLord on Sep 23, 2010
Latest reply on Oct 29, 2010 by SiegeLord

The following kernel when run on a CPU device causes the error_buffer to return non-zero values:

#pragma OPENCL EXTENSION cl_khr_local_int32_base_atomics : enable
void test(__global int* error_buffer) { __local int fire_table_idx; int local_id = get_local_id(0); if(local_id == 0) fire_table_idx = 0; barrier(CLK_LOCAL_MEM_FENCE); if(fire_table_idx != 0) error_buffer[0] = fire_table_idx; else error_buffer[0] = 0; atom_inc(&fire_table_idx); }

If I remove the call to atom_inc at the bottom, the kernel functions fine (i.e. fire_table_idx is set to 0).

Any ideas as to why this might be happening? Am I not using __local variables correctly?