opencl - Serialize work group items in kernel -

opencl - Serialize work group items in kernel -

i have kernel performs 2 tasks, followed b.

a parallelized on work items, , b sequential, first work item performs it.

each work item perform part of b if create them execute in sequence. 2d kernel, , create work items execute in raster pattern.

is possible? 1 thought local variable work items can read, , unique work item matches variable execute , modify variable trigger next work item execute.

any ideas/patterns on how this?

thanks!

edit:

here pseudo code how kernel works:

void mykernel(void) { // perform task (all work items active) barrier(clk_local_memory_fence); if (get_local_id(0) == 0 && get_local_id(1) == 0) { //perform task b } }

and here how work:

// 10 x 10 2d kernel void mykernel(void) { // perform task (all work items active) barrier(clk_local_memory_fence); local activeindex =0; while (activeindex < 100) { if ( get_local_id(0) + 10* get_local_id(1) == activeindex) { // perform part of task b activeindex++; } barrier(clk_local_memory_fence); } } }

edit 2: tried code on hd 7700, slow. guess stick local memory instead.

you can describing on per-workgroup basis. barriers give functionality need. if want on global level, out of luck. using variable check global state before resuming kernel known 'spin-lock', , have caused few wsod (white screen of death) situations trying out myself.

there nil wrong illustration code posted, except maybe lack of params...

void mykernel(void) { // perform task (all work items active) barrier(clk_local_memory_fence); if (get_local_id(0) == 0 && get_local_id(1) == 0) { //perform task b } }

opencl