cudaStreamSynchronize vs CudaDeviceSynchronize vs cudaThreadSynchronize

Question

What is the difference between these three functions, especially the last two? The library manual says

Note that this function is deprecated because its name does not reflect its behavior. Its functionality is similar to the non-deprecated function cudaDeviceSynchronize(), which should be used instead.

But not very sure what does it mean.

Answer 1

These are all barriers. Barriers prevent code execution beyond the barrier until some condition is met.

1. cudaDeviceSynchronize() halts execution in the CPU/host thread (that the cudaDeviceSynchronize was issued in) until the GPU has finished processing all previously requested cuda tasks (kernels, data copies, etc.)
2. cudaThreadSynchronize() as you've discovered, is just a deprecated version of cudaDeviceSynchronize. Deprecated just means that it still works for now, but it's recommended not to use it (use cudaDeviceSynchronize instead) and in the future, it may become unsupported. But cudaThreadSynchronize() and cudaDeviceSynchronize() are similar.
3. cudaStreamSynchronize() is similar to the above two functions, but it prevents further execution in the CPU host thread until the GPU has finished processing all previously requested cuda tasks that were issued in the referenced stream. So cudaStreamSynchronize() takes a stream id as its only parameter. cuda tasks issued in other streams may or may not be complete when CPU code execution continues beyond this barrier.