Debugging and printf in kernel

Is there printf capabilities in CUDA ? More generally, what is the best way of debugging a kernel ?

Yes, there is in-kernel printf available.

I generally recommend to start with proper CUDA error checking and use of compute-sanitizer as I outlined here. More generally, this online training series covers debug in some detail, in session 12.

1 Like

Excellent, thanks very much!

I tried adding printf, but I don’t see any output. Is there anything else I need to do besides adding the call ?

You may need to #include <cstdio> or <stdio.h> (but you would be getting a compile error if you had not included it when needed). You also need to make sure that you have some kind of synchronization like cudaDeviceSynchronize() or cudaMemcpy() after the kernel call that has the printf in it. In other situations, sometime people don’t see printf output because the kernel is failing catastrophically, or didn’t get actually launched at all. In those situations, compute-sanitizer is your friend. If a code reports errors when run under compute-sanitizer, for me personally, I wouldn’t use any other debug methods on it. I would fix those problems first. And running your code under compute-sanitizer is trivially simple to do.

You can find hundreds of questions about using in-kernel printf on various forums, and there is a section in the programming guide on it.

1 Like

Thanks, compute-sanitizer is extremely useful - hopefully I won’t need to resort to printf :)

This topic was automatically closed 14 days after the last reply. New replies are no longer allowed.