NVIDIA OptiX 光线追踪引擎是一个应用框架，旨在让GPU上的光线追踪性能达到最优。

但用OptiX写的应用，有时候会出一些让人头疼的问题：API参数传错、画面黑屏、或者GPU端埋了bug，躲在成千上万个并发线程里，根本找不到源头。

好消息是，NVIDIA OptiX Toolkit（OTK）里的调试工具能帮上大忙。

OTK是一个GitHub仓库，里面装了一堆实用工具，专门针对GPU光线追踪应用的常见工作流。

它采用BSD 3-Clause许可证，你可以随便复制、修改里面的代码，自由度很高。

这篇文章会聊两个OTK的调试利器：

• 对OptiX和CUDA API错误码的一致性检查
• 设备端的定向调试打印

OTK里还带了一个示例程序叫DemandPbrtScene，里面实际演示了设备端调试打印怎么用。

How do OptiX logging and validation work

在聊OTK具体怎么帮之前，先了解一下OptiX自己的日志机制会比较清晰。

创建OptiX设备上下文时，你需要提供一个选项结构，里面可以配置日志回调函数和验证模式。

如果验证模式设成OPTIX_DEVICE_CONTEXT_VALIDATION_MODE_ALL，OptiX就会帮你检查API函数输入的合法性。

OptiX会把验证错误的提示信息写到日志里，这些信息是人类可读的。

所以要是遇到OPTIX_ERROR_INVALID_VALUE这类API错误，第一个要去翻的就是日志输出。

当然，开启验证会带来一些额外的API开销。

建议在调试和测试版本中启用全部验证，发布版本里关掉。

更多细节可以参考OptiX SDK的示例代码。

How to consistently check return codes for errors

最好尽早发现错误——在后续的失败掩盖掉原始问题之前，就把错误揪出来。

下面就来细说。

API mechanisms

OptiX里大部分函数会返回一个OptixResult错误码，非零就表示出错。

CUDA运行时API和CUDA驱动API也类似。

这三个API都支持下面这些机制：

• 一个专门的枚举类型来表示错误码，比如OptixResult
• 一个函数，能把错误码转成符号名称字符串，比如OPTIX_ERROR_INVALID_VALUE
• 一个函数，能返回错误码对应的人类可读错误信息，比如Invalid value

这三个API的函数签名略有不同，但机制是一样的。

Error checking policy

每个API调用点都手动检查错误码，既繁琐又容易遗漏。

更好的做法是用宏或函数调用，强制统一错误处理策略。

OTK提供了两个宏，分别对应两种策略：

• OTK_ERROR_CHECK( expr )：抛出异常
• OTK_ERROR_CHECK_NOTHROW( expr )：打印消息到std::cerr，然后继续执行

如果你想实现其他策略，也很简单——复用已有的基础设施，再写个新宏就行了。

Minimal use of macro machinery

这个错误检查机制只用了少量宏，实际工作都委托给了内联函数。

你可以在内联函数定义处设断点，这样调试器检测到错误时就会停下来。

宏存在的意义，是为了提供出错代码的上下文信息：

• expr：宏参数的字符串形式，也就是计算出错误码的那个表达式
• __FILE__：宏被调用的源文件名
• __LINE__：宏被调用时在源文件中的行号

这些从宏调用点获取的信息，会传给真正执行错误检查的内联函数。

Unified error checking across APIs

内联模板函数checkError负责检查状态码，如果检测到失败，就生成一条诊断消息。

对于这三个API来说，把错误码直接强转为bool就能判断是否出错——它们都用0表示成功，非零表示失败。

错误消息的格式是这样的：

file(line): expr failed with error nnn (name): message

其中expr是被求值的表达式，nnn是状态码强转为int的结果，name是状态码的符号名称，message是人类可读的错误信息。

如果名称或消息为空，函数会省略对应的部分。

内联模板函数makeErrorString负责构建这条消息，它内部会调用getErrorName和getErrorMessage来拼出完整的消息。

每个API都有自己的状态码类型，所以你可以针对不同的API特化模板函数，以调用正确的API来获取扩展错误信息。

Usage

OTK为每个API提供了一个头文件，里面包含了上述模板函数所需的特化（表1）。

头文件	API
	CUDA驱动API
	CUDA运行时API
	OptiX API

使用时，只需包含你所用API对应的头文件，然后在所有调用点统一使用OTK_ERROR_CHECK宏就行了。

下面这个例子同时用了三个API：

OTK_ERROR_CHECK( cudaSetDevice( m_deviceIndex ) );
OTK_ERROR_CHECK( cuCtxGetCurrent( &m_cudaContext ) );
OTK_ERROR_CHECK( cuStreamCreate( &m_stream, CU_STREAM_DEFAULT ) );
OTK_ERROR_CHECK( optixInit() );

How to perform targeted device-side debug printing

图形应用的问题在于——导致黑屏的原因太多了，简直防不胜防。

要调试OptiX设备代码中的问题，有几种常见思路：

• 对设备代码进行调试构建，用CUDA调试器
• 在发布构建的设备代码里用printf获取信息

很多应用在调试模式下编译后跑得特别慢，交互式调试器基本没法用。

而printf式调试的主要麻烦是：GPU上同时跑的线程太多了，输出会像开闸放水一样涌出来，根本看不过来。

而且，问题可能只在应用与用户交互一段时间后才出现——问题还没显性化之前的调试输出，全是噪音，只会干扰你找到真正需要的信息。

DebugLocation

头文件提供了一个可复用的调试输出机制。

结构体DebugLocation控制着整个行为：

struct DebugLocation
{
    bool enabled;
    bool dumpSuppressed;
    bool debugIndexSet;
    uint3 debugIndex;
};

enabled成员用于开关整个机制。

dumpSuppressed成员用来在机制启用时临时关闭调试输出。

debugIndexSet表示一个有效的发射索引已经存到了debugIndex里。

当以下条件全部满足时，该机制才会输出调试信息：

• enabled为真
• dumpSuppressed为假
• debugIndexSet为真，且
• 当前发射索引与debugIndex匹配

在OptiX管线的发射参数中包含一个DebugLocation结构体实例，就可以在运行时交互控制调试输出。

The debugInfoDump function

模板函数debugInfoDump提供了发射调试信息的接口：

template 
static __forceinline__ __device__
bool debugInfoDump( const DebugLocation& debug,
                    const Callback &callback )

Callback模板参数应该是一个结构体或类，满足以下接口：

struct Callback
{
    void setColor( float red, float green, float blue );
    void dump( const uint3& index );
};

setColor方法用于在调试位置周围画一个可视化的方框，方便在屏幕上快速定位信息被输出的点。

典型用法是，把当前发射索引对应的输出像素设为指定颜色。

如果不需要可视化指示，让这个方法空着就行。

dump方法用来打印应用认为相关的任何信息，它会收到当前发射索引作为参数。

Displaying the debug location

启用调试位置显示时，回调结构体的setColor方法会在屏幕上画一个方框来指示当前调试位置，即使dumpSuppressed为真也会画。

当禁用时，方框会被隐藏。

具体的视觉方案是：调试位置的那个像素本身是红色的，外面包一圈1像素宽的黑色边框，再外面包一圈1像素宽的白色边框。

这样就能提供一个高对比度的指示器，告诉你dump消息是从哪个位置产生的。

如果输出缓冲区不是传统的颜色缓冲区，你可以自由地把红、绿、蓝值映射成某种可视化时能区分出来的值。

One-shot mode

为了避免被调试输出淹没，一种很有用的做法是采用“单次模式”——只在用户控制下输出一次。

你可以按照以下步骤来安排：

1. 启用DebugLocation机制
2. 正常发射
3. 当用户交互式地选定了调试位置后，将dumpSuppressed设为true，debugIndexSet设为true，并把debugIndex设为所选位置
4. 后续的发射会显示调试位置，但机制不会输出dump
5. 用户与应用交互，把应用操纵到合适的状态（过程中可以移动调试位置）
6. 当用户想获取当前位置的调试信息时，将dumpSuppressed设为false
7. 发射，获得调试输出
8. 发射后将dumpSuppressed重新设为true

DemandPbrtScene example

OTK中的DemandPbrtScene示例演示了pbrt版本3场景的按需加载几何体。

它使用了DebugLocation机制，包括单次行为、交互式切换调试输出和交互式选择调试位置。

UI框架用的是ImGui。

要运行OTK中的这个示例，你需要一个pbrt-v3的场景文件。

Get started debugging with NVIDIA OptiX Toolkit

OptiX Toolkit为常见的OptiX开发问题提供了可复用的调试和测试工具：一致性API错误检查，以及定向的设备端调试输出。

代码托管在NVIDIA/optix-toolkit GitHub仓库中。

准备好了吗？从GitHub下载OptiX Toolkit，然后开始：

• 在调试构建中启用OptiX验证
• 用OTK_ERROR_CHECK包裹你的CUDA和OptiX调用
• 再使用DebugLocation在开发早期就隔离GPU端的问题

OTK采用宽松的BSD 3-Clause许可证，所以你可以直接复制、改编并将这些工具集成到自己的OptiX应用中。