格式说明符：
至于标准printf（），格式说明符的格式为：％[flags] [width] [.precision] [size] type
支持以下字段（有关所有行为的完整说明，请参阅广泛可用的文档）：
- Flags：
#''0'+' `-' - Width：
*'0-9' - Precision：`0-9'
- Size：
h'l' `ll' - Type：`%cdiouxXpeEfgGaAs'
请注意，CUDA的printf（）将接受标志，宽度，精度，大小和类型的任意组合，不管它们是否总体上形成有效的格式说明符。 换句话说，“％hd”将被接受，并且printf将在参数列表中的相应位置期待一个双精度变量。
限制：
printf（）输出的最终格式在主机系统上进行。 这意味着格式字符串必须被主机系统的编译器和C库理解。 我们尽力确保CUDA的printf函数支持的格式说明符构成最常见主机编译器的通用子集，但确切的行为将取决于主机操作系统。
如格式说明符中所述，printf（）将接受有效标志和类型的所有组合。 这是因为它无法确定在最终输出格式化的主机系统上什么将会有效并且无效。 这样做的效果是，如果程序发出包含无效组合的格式字符串，则输出可能不确定。
除格式字符串外，printf（）命令最多可以接受32个参数。 除此之外的其他参数将被忽略，并且格式说明符按原样输出。
由于64位Windows平台上的长类型（64位Windows平台上的四个字节，其他64位平台上的八个字节）的大小不同，这是一个在非Windows 64位机器上编译的内核，但是 那么在win64机器上运行将会看到包含“％ld”的所有格式字符串的损坏输出。 建议编译平台匹配执行平台以确保安全。
在内核启动之前，printf（）的输出缓冲区设置为固定大小（请参阅关联的主机端API）。 它是循环的，如果在内核执行过程中产生的输出比在缓冲区中更多，那么旧的输出将被覆盖。 只有在执行其中一项操作时才会刷新它：
- 通过<<< >>>或cuLaunchKernel（）启动内核（在启动开始时，如果CUDA_LAUNCH_BLOCKING环境变量设置为1，也在启动结束时）
- 通过cudaDeviceSynchronize（），cuCtxSynchronize（），cudaStreamSynchronize（），cuStreamSynchronize（），cudaEventSynchronize（）或cuEventSynchronize（）同步，
- 内存通过任何阻止版本的cudaMemcpy （）或cuMemcpy （）进行复制，
- 模块通过cuModuleLoad（）或cuModuleUnload（）加载/卸载，
- 通过cudaDeviceReset（）或cuCtxDestroy（）来破坏上下文。
- 在执行由cudaStreamAddCallback或cuStreamAddCallback添加的流回调之前。
请注意，程序退出时，缓冲区不会自动刷新。 用户必须显式调用cudaDeviceReset（）或cuCtxDestroy（），如下例所示。
内部printf（）使用共享数据结构，因此调用printf（）可能会更改线程的执行顺序。 特别是，调用printf（）的线程可能比不调用printf（）的线程执行时间更长，并且该路径长度取决于printf（）的参数。 但是，请注意，除明确的__syncthreads（）障碍外，CUDA不保证线程执行顺序，因此无法判断执行顺序是否已由printf（）修改或通过硬件中的其他调度行为。
关联的主机端API：
以下API函数获取并设置用于将printf（）参数和内部元数据传输到主机的缓冲区的大小（默认值为1兆字节）：
- cudaDeviceGetLimit(size_t* size,cudaLimitPrintfFifoSize)
- cudaDeviceSetLimit(cudaLimitPrintfFifoSize, size_t size)
例子：
#include <stdio.h>
__global__ void helloCUDA(float f)
{
printf("Hello thread %d, f=%f\n", threadIdx.x, f);
}
int main()
{
helloCUDA << <1, 5 >> >(1.2345f);
cudaDeviceSynchronize();
getchar();
return 0;
}结果：
Hello thread 2, f=1.2345
Hello thread 1, f=1.2345
Hello thread 4, f=1.2345
Hello thread 0, f=1.2345
Hello thread 3, f=1.2345注意每个线程是如何遇到printf（）命令的，因此输出的行数与网格中启动的线程数一样多。 正如预期的那样，全局值（即float f）在所有线程之间是公共的，并且本地值（即threadIdx.x）不同于每个线程。
#include <stdio.h>
__global__ void helloCUDA(float f)
{
if (threadIdx.x == 0)
printf("Hello thread %d, f=%f\n", threadIdx.x, f);
}
int main()
{
helloCUDA << <1, 5 >> >(1.2345f);
cudaDeviceSynchronize();
return 0;
}结果：
Hello thread 0, f=1.2345不言而喻，if（）语句限制了哪些线程将调用printf，以便只能看到一行输出。
