Ядро CUDA, по-видимому, игнорирует «if» заявление

Далее следует часть моего ядра, которая не работает должным образом, затем объяснение того, что я нашел во время отладки.

__global__ void Mangler(float *matrix, int *map)
{
__shared__ signed int localMap[N];

if(0 == threadIdx.x)
{
for(int i=0; i<N; i++)
localMap[i] = -1;
}

__syncthreads();

int fn = ...; // a lot of code goes into this number, skipped for clarity
int rnumber = threadIdx.x;

int X = atomicCAS(&localMap[fn], -1, rnumber); // Spot of bother 1

if(X == -1) // Spot of bother 2
{
// some code
}
else
{
// other code
}
}

Я нашел в документации, что atomicCAS(*address, compare, value) в основном возвращает (и сохраняет по указанному адресу) результат (old == compare ? value : old)где old — это значение по адресу перед выполнением функции.

Идя с этим, я считаю, что выполнение int X = atomicCAS(&localMap[fn], -1, rnumber); должен иметь два возможных результата (согласно Руководству по программированию NVidia Cuda C):

• если localMap[fn] == -1 затем X должен иметь значение rnumber а также localMap[fn] должен иметь значение rnumber, Такого не бывает.
• если localMap[fn] != -1 затем X должно быть установлено значение localMap[fn] и указанное значение следует оставить без изменений.

Вместо этого, как показала мне отладка в NSight, происходит следующее: X присваивается -1, а localMap[fn] присваивается значение rnumber, Я не понимаю этого, но, как вы можете видеть в моем коде, я изменил if чтобы поймать эту ситуацию.

Что подводит меня к проблеме номер 2: хотя NSight показывает ценность X как -1, if {} полностью пропускается (нет точек останова в пределах попадания), и выполнение переходит прямо к else,

Мои вопросы:

• Я неправильно понимаю atomicCAS полностью? Да, я сделал
• Что может вызвать и if который должен быть оценен как истинный, чтобы прыгнуть прямо в else в коде устройства?

Я использую NVidia CUDA 5.5, Visual Studio 2012 x64 для Windows 8, NVidia Nsight Monitor Visual Studio Edition 3.1. Графическим процессором для машины является NVidia GeForce GTX 550 Ti.

Я попытался изменить синтаксис на if(X!=-1); истинная ветвь if все еще не выполняется.