waiting on variable from host fails

Accelerated Computing CUDA CUDA Programming and Performance
1

When I run this code:

[codebox]global void kernel(int *arr) {

    int i;

    for(i = 0;i < 10; i++) {

            while(arr[i] == 0) ;

    }

}

int main() {

    int *d_arr;

    int arr[10] = {0,0,0,0,0,0,0,0,0,0};

    cudaMalloc((void **)&d_arr, sizeof(int) * 10);

    cudaMemcpy(d_arr, arr, sizeof(int) * 10, cudaMemcpyHostToDevice);

    kernel<<<1,256>>>(d_arr);

    cudaThreadSynchronize();

    return 0;

}

[/codebox]

the program seems to exit instantly, when it seems to me this should be an infinite loop on the device, with the host waiting for it. Even adding printf(“%s\n”,cudaGetErrorString(cudaGetLastError())); at the end of the host code detects no error. Does anyone know of an explanation for this?

Thanks for any help or suggestions.

2

I haven’t tried compiling this myself to be sure, but have you looked at the PTX for this kernel? The compiler is very good at removing dead code – by which I essentially mean code that does not contribute to a write to global memory. In that sense, this entire kernel is “dead”, so I wouldn’t be too surprised if the compiler was just making the whole thing a noop…

And of course it’s also possible that the answer is even more simple: there could be a CUDA error being returned from one of these functions that you’re not catching. Add checks for errors (at least in debug mode) to be sure.

3

Thanks for the reply, that was exactly the issue. Added a write to the loop and the device code doesn’t return. Thanks very much!