Synchronous and Asnychronous kernel calls (Ref: NVidia manual)

Types of kernel calls

There are 2 types (kinds) of kernel calls (launches):

Synchronous kernel call:

The CPU will launch (= starts execution) a kernel function and then
Wait until the kernel function completes before executing the next CPU program statement

Asynchronous kernel call:

Kernel function calls (launches) are asynchronous

Important fact in CUDA programming:

Launching a (user-defined) kernel function is always asynchronous !!!

Example:

int main() { /* ------------------------------------ Call the hello( ) kernel function ------------------------------------ */ hello<<< 1, 4 >>>( ); // Asynchronous !!! printf("I am the CPU: Hello World ! \n"); return 0; }

Effect:

Because hello<<<1, 4>>>( ) is asynchronous, the call will return immediately

Then:

printf("I am the CPU: Hello World ! \n");

is executed

Finally, the CPU program will return (= exit !):
And the GPU function does not have the chance to run !!!

Forcing the CPU to wait on the execution on the GPU (Ref: click here)

To force the CPU execution to wait on the termination of all activities on the GPU, use:

cudaDeviceSynchronize(); This call will forces the CPU to wait until all preceeding command/function calls on the GPU have completed

The Hello World CUDA program revisited

The correct way to write the hello world CUDA program is:

#include <stdio.h> #include <unistd.h> __global__ void hello( ) { printf("Hello World !\n"); } int main() { hello<<< 1, 4 >>>( ); printf("I am the CPU: Hello World ! \n"); // Try moving the "printf" statement // after cudaDeviceSynchronize() cudaDeviceSynchronize(); return 0; }

Example Program: (Demo above code)

/home/cs355001/demo/CUDA/1-intro/hello-sync Output: I am the CPU: Hello World ! Hello World ! Hello World ! Hello World ! Hello World !