NVIDIA OptiX光线追踪引擎是一款专为在GPU上实现最优光线追踪性能而设计的应用框架。使用OptiX的应用程序可能会以难以诊断的方式出现故障，例如无效的API参数、全黑画面，或隐藏在数千个并发线程中的GPU端错误。
NVIDIA OptiX工具包（OTK）提供的调试工具可以帮助解决上述问题。OTK是一个GitHub代码仓库，包含一系列支持GPU光线追踪应用常见工作流程的实用工具，并采用BSD 3条款风格的许可证，允许用户自由复制和修改其中的代码。
本文将重点介绍OTK的两项调试功能：对OptiX和CUDA API错误码的一致性检查，以及有针对性的设备端调试打印。OTK还提供了一个示例程序DemandPbrtScene，展示了设备端调试打印在实际场景中的应用。
OptiX日志背景
在使用OTK提供的辅助功能之前，有必要先了解一些关于OptiX日志的背景知识。创建OptiX设备上下文时，需要提供一个选项结构体，用于配置日志回调和验证模式。当验证模式设置为OPTIX_DEVICE_CONTEXT_VALIDATION_MODE_ALL时，OptiX能够协助验证API函数的输入参数。
OptiX会将验证错误的可读信息写入日志。遇到类似OPTIX_ERROR_INVALID_VALUE这样的API错误时，日志输出应是首先排查的地方。验证功能会在API层面带来一定的额外开销，因此建议在调试和测试版本中启用全部验证，而在正式发布版本中关闭验证。更多详情请参阅OptiX SDK示例。
尽早发现错误
在后续失败掩盖原始问题之前，尽早发现错误至关重要。
OptiX中的大多数函数会返回一个OptixResult错误码，非零值表示发生了错误。CUDA运行时API和CUDA驱动API遵循类似的模式。三套API均支持以下机制：
- 用于错误码的独立枚举类型，例如OptixResult
- 返回错误码符号名称字符串的函数，例如OPTIX_ERROR_INVALID_VALUE
- 返回错误码可读信息的函数，例如"Invalid value"
这三套API中上述函数的签名略有不同，但机制相同。
OTK错误检查宏
在每个API调用处手动处理这些错误码既繁琐又容易出错，最好使用宏或函数调用来统一执行错误处理策略。OTK提供了两种错误检测策略的宏：
- OTK_ERROR_CHECK( expr )：抛出异常
- OTK_ERROR_CHECK_NOTHROW( expr )：向std::cerr打印信息并继续执行
通过复用已有机制并创建适当的宏，也可以轻松实现其他处理策略。
该错误检查机制将宏的使用降至最低，并将实际工作委托给内联函数完成。可以在内联函数定义处设置断点，使调试器在检测到错误时暂停执行。
宏的作用在于提供导致错误的代码的诊断信息：
- expr：传入宏的参数字符串形式，即求值为错误码的表达式
- __FILE__：调用宏所在的源文件名称
- __LINE__：调用宏所在源文件的行号
这些来自宏调用处的信息将被传递给负责实际错误检查的内联函数。
错误检查的内部实现
内联模板函数checkError用于检查状态码是否存在错误，并在检测到失败时生成诊断信息。对于上述三套API，将错误码简单转换为bool类型即可判断是否出错。三套API均以零表示成功，非零表示失败。
错误信息的格式如下：
file(line): expr failed with error nnn (name): message
其中expr为求值后的表达式，nnn为将状态码转换为int后的结果，name为状态码的符号名称，message为可读的错误信息。若name或message为空，函数将自动省略对应部分。
内联模板函数makeErrorString负责构建该信息，它调用内联模板函数getErrorName和getErrorMessage来合并生成完整信息。由于每套API的状态码类型各不相同，可以对模板函数进行特化，以调用相应API获取扩展错误信息。
OTK为每套API提供一个头文件，包含上述模板函数的必要特化实现（见表1）。只需引入所用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() );
设备端调试打印
图形应用程序的问题在于，导致黑屏的原因实在太多。要调试OptiX设备端代码的问题，可以采取以下几种方式：
- 在设备代码的调试版本上使用CUDA调试器
- 通过printf在发布版本的设备代码中获取信息
许多应用程序在以调试模式编译时运行速度过慢，影响交互式调试器的正常使用。
printf式调试的主要难点在于GPU上有大量线程并发运行，输出内容极易泛滥成灾。此外，问题可能只在与应用程序进行一定交互后才会出现，问题可视化之前产生的调试输出纯属干扰，反而妨碍找到有效信息。
DebugLocation机制
头文件<OptiXToolkit/ShaderUtil/DebugLocation.h>提供了一种可复用的调试输出机制。DebugLocation结构体用于控制其行为：
struct DebugLocation {
bool enabled;
bool dumpSuppressed;
bool debugIndexSet;
uint3 debugIndex;
};
其中enabled成员用于开关整个机制；dumpSuppressed成员在机制启用时关闭调试输出；debugIndexSet成员表示debugIndex中已存储有效的启动索引。
满足以下条件时，该机制将输出调试信息：
- enabled为true
- dumpSuppressed为false
- debugIndexSet为true，且当前启动索引与debugIndex匹配
在OptiX管线的启动参数中加入DebugLocation结构体实例，即可实现对调试输出的交互式控制。
debugInfoDump函数
模板函数debugInfoDump提供了发出调试信息的接口：
template <typename Callback>
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方法用于打印应用程序在指定启动索引处认为相关的任何信息。
调试位置的可视化标记
启用机制后，即使关闭了dump输出，回调结构体的setColor方法也会在屏幕上绘制方框以标示当前调试位置；禁用时方框隐藏。
调试位置的红色像素位于一个一像素宽的黑色方框内，黑色方框外再套一个一像素宽的白色方框，从而形成高对比度的位置指示器。若输出缓冲区不是传统的颜色缓冲区，可以自由地将提供的红、绿、蓝值映射为某种便于可视化的特殊值。
一次性调试输出模式
为避免调试输出泛滥，建议采用一次性输出模式，即根据用户操作仅输出一次调试信息。可按以下顺序使用该模式：
1. 启用DebugLocation机制
2. 按常规发起启动
3. 当用户交互式选择调试位置时，将dumpSuppressed设为true、debugIndexSet设为true，并将debugIndex设为所选位置
4. 后续启动将显示调试位置标记，但不会输出dump信息
5. 用户与应用程序交互，将其调整至适当状态，期间可移动调试位置
6. 当用户希望获取当前位置的调试信息时，将dumpSuppressed设为false
7. 发起启动以获取调试输出
8. 启动完成后将dumpSuppressed重新设为true
DemandPbrtScene示例
OTK中的DemandPbrtScene示例演示了针对pbrt第3版场景的按需加载几何体功能，其中使用了DebugLocation机制，包括一次性输出行为、调试输出的交互式切换以及调试位置的交互式选取，UI框架采用ImGui。如需从OTK运行此示例，需要准备一个pbrt-v3的场景文件。
总结
OptiX工具包为常见的OptiX开发问题提供了可复用的调试与测试工具：一致性API错误检查和有针对性的设备端调试输出。相关代码可通过NVIDIA/optix-toolkit GitHub仓库获取。
如需立即上手，可从GitHub下载OptiX工具包，从在调试版本中启用OptiX验证开始，使用OTK_ERROR_CHECK封装CUDA和OptiX调用，并借助DebugLocation在开发早期隔离GPU端错误。OTK采用宽松的BSD 3条款风格许可证，可直接将这些工具复制、改编并集成到自己的OptiX应用程序中。
Q&A
Q1：NVIDIA OptiX工具包（OTK）是什么？有什么用？
A：OTK是NVIDIA提供的一个开源GitHub代码仓库，包含一系列用于GPU光线追踪应用开发的实用工具。它主要提供两大调试能力：对OptiX和CUDA API错误码的一致性检查，以及有针对性的设备端调试打印功能，帮助开发者更高效地定位和修复光线追踪应用中的问题。OTK采用BSD 3条款许可证，可自由复制和修改。
Q2：OTK_ERROR_CHECK宏怎么用？和直接判断返回值有什么区别？
A：OTK_ERROR_CHECK是一个封装宏，只需将API调用包裹其中，例如OTK_ERROR_CHECK( optixInit() )，它会在检测到非零错误码时自动抛出异常，并附带出错文件名、行号和错误详情。相比手动判断每个调用的返回值，这种方式更简洁、更不容易遗漏，也更便于调试器设置断点。OTK还提供了不抛异常、仅打印信息的OTK_ERROR_CHECK_NOTHROW版本。
Q3：DebugLocation机制如何避免GPU调试时输出内容过多？
A：DebugLocation通过只在特定启动索引（即屏幕上特定像素位置）处输出调试信息来解决这一问题，避免了数千个并发线程同时打印造成的信息泛滥。配合一次性输出模式，可以先让应用程序运行到目标状态，再由用户手动触发单次dump输出，之后自动关闭，从而精准获取所需调试信息，不受无关输出干扰。
