while loop seems to not working in cuda "whille" infinite loop breaks .

Accelerated Computing CUDA CUDA Programming and Performance
1

I wrote a cuda kernel which use some conditions in some places . I used while loops and if conditions . The results are wrong .
When I tried to debug it by commenting certain lines , i found that the while loop condition has no effect on the code .
To check it out , I tried a loop which should be infinite in principle . But to my surprise the program can pass over the loop somehow
and executing although incorrect .

Any help please ?? I am frustrated trying different conditions .

The infinite loop I tried is

while ( flag != 10 ){
if(flag==10);}
The flag is not going to be 10 at any point in the program and I thought the loop should keep on checking if flag equals 10 . Sorry ,Is this stupid ??

What I want is that the thread should keep on checking whether flag = 10 , until it becomes 10 by the neighboring thread .

Here I commented out the step where flag should get value 10 , to make it infinite . But it is running !! Please help.

2

I wrote a cuda kernel which use some conditions in some places . I used while loops and if conditions . The results are wrong .
When I tried to debug it by commenting certain lines , i found that the while loop condition has no effect on the code .
To check it out , I tried a loop which should be infinite in principle . But to my surprise the program can pass over the loop somehow
and executing although incorrect .

Any help please ?? I am frustrated trying different conditions .

The infinite loop I tried is

while ( flag != 10 ){
if(flag==10);}
The flag is not going to be 10 at any point in the program and I thought the loop should keep on checking if flag equals 10 . Sorry ,Is this stupid ??

What I want is that the thread should keep on checking whether flag = 10 , until it becomes 10 by the neighboring thread .

Here I commented out the step where flag should get value 10 , to make it infinite . But it is running !! Please help.

1 Like
3

Have you declared flag as volatile?

4

Have you declared flag as volatile?

5

Could you post a complete small example that reproduces the problem? This is a common problem that a lot of people end up running into.

Are you doing something like this?

shared done = false;

if(threadIdx.x != 0)
while(!done)
{
…
}

if(threadIdx.x == 0)
{
done = true;
}

1 Like
6

Could you post a complete small example that reproduces the problem? This is a common problem that a lot of people end up running into.

Are you doing something like this?

shared done = false;

if(threadIdx.x != 0)
while(!done)
{
…
}

if(threadIdx.x == 0)
{
done = true;
}

7

My flag is not volatile , it is declared as a global variable .

device int flag ;

My program is somewhat similar to the form you mentioned .

//** Inside the kernel *********

flag=1;

for(t=1;t<=100;t++){

ind = 1;

    while(ind==1){

    {if(flag==t){  ind =2;}

    else

    continue;

            }

//***************code *************//

if (threadIdx.x==0)

flag++;

}

If I comment out his flag++; step , from the 2nd iteration onwards , the loop should hang , shouldn’t it ??

The difference from this code snippet you mentioned is that “done” is a register variable , and the while loop is evaluated by all threads including threadIndex==0 , which updates “done” .

8

My flag is not volatile , it is declared as a global variable .

device int flag ;

My program is somewhat similar to the form you mentioned .

//** Inside the kernel *********

flag=1;

for(t=1;t<=100;t++){

ind = 1;

    while(ind==1){

    {if(flag==t){  ind =2;}

    else

    continue;

            }

//***************code *************//

if (threadIdx.x==0)

flag++;

}

If I comment out his flag++; step , from the 2nd iteration onwards , the loop should hang , shouldn’t it ??

The difference from this code snippet you mentioned is that “done” is a register variable , and the while loop is evaluated by all threads including threadIndex==0 , which updates “done” .