CUDA - nvcc -G - 如果工作不正常

Posted 2023-03-23

【中文标题】CUDA - nvcc -G - 如果工作不正常

【英文标题】：CUDA - nvcc -G - if not working properly

【发布时间】：2015-07-19 02:38:21

【问题描述】：

我目前正致力于在 CUDA 中移植熔岩流模型（github 上的完整代码：Full source of the CUDA-SCIARA Fv2 lava flow model。

编辑

为了重现该问题，请使用 github README 中的说明获取项目 SCIARA_FV2_CUDA_MULTICELLS 的开发副本。然后编译没有 -G 选项并作为命令行参数传递 -c ../data/2006/PARAMETERS.cfg 。 （引用的代码在 file 的第 260 行）

我遇到了 __device__ 函数中的 if 构造问题。

根据当前的熔岩数量和温度计算新的温度，如果它低于一个常数参数（变量d_PTsol=1143.0），熔岩就会凝固。

下面代码中的问题是，如果我使用 -G 选项（用于生成设备代码调试信息）进行编译，但如果没有，则行为错误。

double new_temp = d_computeNewTemperature(sommah,sommath);        
if(new_temp <= d_PTsol)
            printf("Solidified %.5f,%.5f\n",new_temp,d_PTsol);
            double newQuote = d_sbts_updated[d_getIdx(row,col,ALTITUDE)]+d_sbts_current[d_getIdx(row,col,THICKNESS)];
            //CODE FOR LAVA SOLIDIFICATION HERE
    else
           //there is lava and is not solidified -> activate this cell!
           adjustAdaptiveGrid(row,col);
 

在模拟的某个点输出类似的东西：

Solidified 1344.68654 1143.00000
Solidified 1343.99509 1143.00000
Solidified 1320.50061 1143.00000
Solidified 1325.53942 1143.00000

如果我将 if 条件更改为严格的不等式 if(new_temp < d_PTsol)，问题会完全消失。

使用以下选项并在单独的编译模式下进行编译

-O3 -Xcompiler -fPIC -std=c++11

并使用链接

--cudart static --relocatable-device-code=true -gencode arch=compute_35,code=compute_35 -gencode arch=compute_35,code=sm_35

以前有人遇到过类似的问题吗？我是不是做错了什么？

更新

这个问题似乎与以

if(new_temp <= d_PTsol) 
        //solidification
else
        //something else

到

if(new_temp <= d_PTsol) 
        //solidification

if(!(new_temp <= d_PTsol))
        //something else

使代码完美运行。

【问题讨论】：

这通常表示竞争条件/未初始化的内存/未定义的行为。我建议你通过 cuda-memcheck 运行它，它是子工具（racecheck、synccheck、initcheck）。

@Jez 我已经运行了所有的 cudamemcheck 套件：泄漏、比赛初始化和同步检查。一切都顺利通过。那么为什么只有在没有启用设备调试的情况下才会出现竞争条件呢？此外， new_temp 变量是一个寄存器，并没有明显地初始化（它被正确打印出来了！）。同样适用于变量d_PTsol。

我第一次略过这个问题有点太快了 - 抱歉。这确实是一个奇怪的问题。有没有检查下print语句执行后的代码？可能在优化过程中错误地移动了打印语句。一个重现的例子会很有用。

我今天将尝试重新编译所有不同的系统和不同的目标设备，我会告诉更新问题并提供更多详细信息。打印不是唯一错误执行的指令（因为这种行为会导致数值差异，这让我发现了这个“问题候选人”）。

好吧，我可以在这里问很多问题，比如您是否有任何包含此语句或将其一分为二的条件线程行为，您是否在任何您没有的代码中使用__syncthreads()显示等。但 SO 实际上希望这样的问题有一个 MCVE，这样你可能会更快地得到更令人满意的答案。 （带有字面意思的 no 指令的 github 存储库链接不构成 MCVE。）这很可能是一个编译器错误，但如果没有简短的复制器，该途径将没有任何吸引力。由于缺少 MCVE，我已投票关闭。

【参考方案1】：

我遇到了类似的问题。我的代码使用 -G 选项但并非没有。我只需要将 -fmad=false 和 -prec-div=false 添加到编译器选项（有关更多信息，请参阅：http://docs.nvidia.com/cuda/cuda-compiler-driver-nvcc/index.html#options-for-steering-gpu-code-generation）。

默认情况下，此值为 true 并优化您的单精度操作，但以您的精度为代价。如果您启用调试模式，此优化将被禁用。我需要精确的值，因此我的代码只能与 -G 一起使用。也许你有类似的问题。 （更多关于CUDA浮点精度的信息：http://docs.nvidia.com/cuda/cuda-c-programming-guide/#mathematical-functions-appendix）