有没有办法在nvcc中为设备代码（内核代码）启用`-Wconversion`？

Question

我最近正在学习用 CUDA 编写前缀和算法。我有一个愚蠢的错误，我将一个浮点变量分配给一个失去精度的整数变量：

// Phase 3: populate last element o previous subsection.
__shared__ float XY[SECTION_SIZE];
// ...
const int prev_sec_sum = ((subsection_idx > 0) ? XY[subsection_idx - 1] : 0.0f);
//    ^ Should be float here.
for (size_t i = 0; i < COARSEN_FACTOR - 1; i++) {
    XY[subsection_idx + i] += prev_sec_sum;
}

我应该对此更加小心，但令我惊讶的是

nvcc

没有警告转换。我进行了搜索，然后我意识到以下编译器标志仅适用于主机代码：

--compiler-options -Wall,-Wextra,-Wconversion

搜索 nvcc 文档后，我能找到的唯一编译器标志是

--Werror all-warnings

，它不会对此转换生成任何警告。你知道

nvcc

是否支持这种转换检查吗？例如，我们的设备代码是否有类似

-Wconversion

的标志？

我累了：

nvcc -o bin/prefix_sum prefix_sum.cu --compiler-options -Wall,-Wextra,-Wconversion --Werror all-warnings --compiler-bindir /usr/bin/gcc-12 -arch=native

但它只检查主机代码的转换错误。

感谢您的帮助。

Answer 1

回答我自己的问题，为刚开始学习CUDA编程的人提供一些参考。

感谢@paleonix的帮助，我使用clang来构建我的CUDA程序。

使用构建选项

-Weverything -Werror

，我现在可以从中获取

-Wimplicit-int-float-conversion

错误/警告。

对于此代码片段：

// Phase 3: populate last element of previous subsection.
const int prev_sec_sum = ((subsection_idx > 0) ? XY[subsection_idx - 1] : 0.0f);
//    ^ Should be float here.
for (size_t i = 0; i < COARSEN_FACTOR - 1; i++) {
    XY[subsection_idx + i] += prev_sec_sum;
}

我使用了以下命令：

clang++-18 -Weverything -Werror -Wno-reserved-identifier -Wno-unsafe-buffer-usage prefix_sum.cu -o bin/prefix_sum --cuda-gpu-arch=sm_89 -L/usr/local/cuda/lib64 -lcudart_static

这是结果：

prefix_sum.cu:467:51: error: implicit conversion turns floating-point number into integer: 'float' to 'const int' [-Werror,-Wfloat-conversion]
  467 |         const int prev_sec_sum = ((subsection_idx > 0) ? XY[subsection_idx - 1] : 0.0f);
      |                   ~~~~~~~~~~~~                           ^~~~~~~~~~~~~~~~~~~~~~
prefix_sum.cu:469:29: error: implicit conversion from 'const int' to 'float' may lose precision [-Werror,-Wimplicit-int-float-conversion]
  469 |                 XY[subsection_idx + i] += prev_sec_sum;
      |                                        ~~ ^~~~~~~~~~~~
2 errors generated when compiling for sm_89.

这正是我想要的。