3.2.8 错误处理
所有的运行时函数都返回错误码，但对于异步函数(参见3.2.6节)，由于会在任务结束前返回，因此错误码不能报告异步调用的错误;错误码只报告在任务执行之前的错误，典型的错误有关参数有效性;如果异步调用出错，错误将会在后面某个无关的函数调用中出现。
唯一能够检查异步调用出错的方式是通过在异步调用函数后面使用cudaThreadSynchronize()同步(或使用3.2.6节介绍的其它同步机制)，然后检查cudaThreadSynchronize()的返回值。
运行时为每个主机线程维护着一个初始化为cudaSuccess的错误变量，每次错误发生(可以是参数不正确或异步错误)时，该变量会被错误码重写。cudaGetLastError()会返回这个变量，并将它重新设置为cudaSuccess。
内核发射不返回任何错误码，所以应当在内核发射后立刻调用cudaGetLastError()检测发射前错误。为保证cudaGetLastError()返回的错误值不是由于内核发射前错误导致的，必须保证运行时错误变量在内核发射前被设置为cudaSuccess，可以通过在内核发射前调用cudaGetLastError()实现。内核发射是异步的，因此为了检测异步错误，应用必须在内核发射和cudaGetLastError()之间同步。
注意cudaStreamQuery()可能返回cudaErrorNotReady，而由于cudaEventQuery()没有考虑错误，因此不会被cudaGetLastError()报告。
3.2.9 使用设备模拟模式调试
CUDA-GDB可以用于调试计算能力大于1.0的设备(参看 CUDA-GDB用户手册获得支持平台)。编译器和运行时也支持模拟模式调试，模拟模式可以在没有CUDA支持的设备上运行。当使用模拟模式编译应用时(使用-deviceemu选项)，设备代码为主机编译，并在主机上运行，允许程序员使用主机上的本地调试器，就好像是主机程序一样。预处理宏__DEVICE_EMULATION__是为这种模式定义的。使用了任何库的所有应用代码必须一致的用设备模拟模式或设备执行模式编译。链接设备模拟模式编译的代码和设备执行模式编译的代码会返回下面的运行时初始化错误：cudaErrorMixedDeviceExecution。
在设备模拟模式下运行应用，运行时模拟编程模型。对于块内的每个线程，运行时在主机上建立一个线程。程序员要确定：
1、主机能够运行每个块的最大线程数，外加一主线程。
2、足够的可用存储器用于运行所有线程，已知每个线程都要占据256 KB栈空间。
设备模拟模式提供的许多特性让他成为一种有效的调试工具：
1、通过使用主机的本地调试支持，程序员可利用调试器支持的所有特性，例如断点设置和数据检查。
2、由于设备代码编译后在主机上运行，故可调用无法在设备上运行的代码扩展功能，如文件输入和输出操作或输出到屏幕(printf() 等)。
3、由于所有数据都位于主机上，因而可从设备或主机代码可读取任何特定于设备或主机的数据;类似地，设备和主机可调用任何设备或主机函数。
4、万一误用了内置同步函数，运行时将检测到死锁情况。
程序员必须牢记，设备模拟模式的目的在于模拟设备，而非仿真。因此，设备模拟模式在查找算法错误时非常有用，但是很难发现某些错误：
1、竞争条件在设备模拟模式中体现的可能很小，因为同时执行的线程数量要比实际运行在设备上的少得多。
2、在主机上解引用指向全局存储器的指针或在设备上解引用指向主机存储器的指针时，设备执行几乎必然以某种没有定义的方式失败，而设备模拟能得到正确的结果。
3、大多数时候，相同的浮点计算在设备上执行时所得到的结果，与在设备模拟模式中得到的结果并不完全相同。这是可预见的，因为一般只要使用略有差异的编译器选项，相同的浮点计算就会得到不同的结果，更不用说不同的编译器、不同的指令集或不同的架构了。
特别地，某些主机平台会将单精度浮点计算的中间结果存储在一个扩展精度寄存器中，在设备模拟模式中运行时，这潜在地导致精确性的显著差异。当发生这种情况时，程序员可尝试下面的任何一种方法，但没有任何一种方法能确保成功：
Ø 声明一些易变的浮点变量，强制单精度存储;
Ø 使用 gcc 的 -ffloat-store 编译器选项;
Ø 使用 Visual C++ 编译器的 /Op 或 /fp 编译器选项;
Ø 在 Linux 上使用 _FPU_GETCW()、在 Windows 上使用 _controlfp()，使用这些代码环绕原代码的一部分，强制使用单精度浮点计算：
_FPU_GETCW(originalCW);
unsigned int cw = (originalCW & ~0x300) | 0x000;
或
_controlfp(_PC_24, _MCW_PC);
在开头处，为了存储控制字的当前值并更改它以使尾数存储在 24 个位中，可使用：
或在结尾处，使用下面代码，重置原始控制字：
另外，对于单精度浮点数，计算能力1.1的设备不支持非正规数(参见附录 G)，而主机平台通常支持。这可能会导致设备模拟模式和设备执行模式的结果显著不同，因为某些计算可能会在一种模式下产生有穷的结果，而在另一种模式下产生无穷的结果。
4、在设备模拟模式中，束的大小等于 1(参见的4.1节了解束的定义)。因而，束表决(warp vote)函数的结果将与设备执行模式不同。
