Global does not update correctly
Hi All, I am looking for some help with this code
__kernel void ReduceBoundsFromGBuffer(
__global int *p_Semaphors,
__global PartitionBoundsC* p_PartionBounds
)
{
uint2 groupId; // SV_GroupID.xy
uint2 groupThreadId; // SV_GroupThreadID.xy
uint groupIndex ; // SV_GroupIndex
groupId.x = get_group_id(0);
groupId.y = get_group_id(1);
groupThreadId.x = get_local_id(0);
groupThreadId.y = get_local_id(1);
groupIndex = get_local_size(0)*groupThreadId.y+groupThreadId.x;
// Now write out the result from this pass
if (groupIndex == 0)
{
GetSemaphor(&p_Semaphors[2]);
{
// p_Semaphors[0] += 1.0f;
p_PartionBounds[1].c_MinCoord.y += 1.0;
p_PartionBounds[2].c_MinCoord.y += 1.0f;
p_PartionBounds[3].c_MinCoord.y += 1.0f;
p_PartionBounds[0].c_MinCoord.y += 1.0f;
}
ReleaseSemaphor(&p_Semaphors[2]);
}
}
If I run this on cpu's only it works correctly
p_PartionBounds[1].c_MinCoord.y has number of threads run.
If I run it on a GPU (5870)
p_PartionBounds[1].c_MinCoord.y has 1,
all other y values are correct
if I uncomment the line
// p_Semaphors[0] += 1.0f;
p_PartionBounds[1].c_MinCoord.y has the correct value
Am I doing something obviouly wrong?
PS the semaphore code seems to be ok