CUDA错误 CUDA_ERROR_NO_BINARY_FOR_GPU

Question

我有一些PTX代码无法加载。我的电脑是650M型号的，运行OSX系统。其他CUDA示例在该系统上都能正常运行，但当加载模块时，我总是得到错误209：CUDA_ERROR_NO_BINARY_FOR_GPU。

我该怎么办？

 .version 3.1
.target sm_20, texmode_independent
.address_size 64


    // .globl   examples_2E_mandelbrot_2F_calc_2D_mandelbrot_2D_ptx
.entry examples_2E_mandelbrot_2F_calc_2D_mandelbrot_2D_ptx(
    .param .u64 .ptr .global .align 8 examples_2E_mandelbrot_2F_calc_2D_mandelbrot_2D_ptx_param_0,
    .param .f64 examples_2E_mandelbrot_2F_calc_2D_mandelbrot_2D_ptx_param_1,
    .param .f64 examples_2E_mandelbrot_2F_calc_2D_mandelbrot_2D_ptx_param_2,
    .param .f64 examples_2E_mandelbrot_2F_calc_2D_mandelbrot_2D_ptx_param_3
)
{
    .reg .pred %p<396>;
    .reg .s16 %rc<396>;
    .reg .s16 %rs<396>;
    .reg .s32 %r<396>;
    .reg .s64 %rl<396>;
    .reg .f32 %f<396>;
    .reg .f64 %fl<396>;

    ld.param.u64    %rl0, [examples_2E_mandelbrot_2F_calc_2D_mandelbrot_2D_ptx_param_0];
    mov.b64 func_retval0, %rl0;
    ret;
}

- Timothy Baldridge

---

GT650M是一款sm_30的GPU。如果您将“.target sm_20”更改为“.target sm_30”，会发生什么？或者我应该问，您是如何生成这个PTX代码的？ - Robert Crovella

代码是通过llvm生成的。我已经将代码削减到您在上面看到的级别。我尝试了sm_10、sm_13、sm_30和sm_35的目标模型。都一样。从.entry切换到.func允许模块加载，但是（当然）我找不到函数。 - Timothy Baldridge

cuModuleGetFunction返回CUDA_ERROR_NOT_FOUND。 - Timothy Baldridge

你可能想通过使用nvcc -arch=sm_30 -ptx mymodule.cu编译一些代码，并分析它们之间的差异来创建类似的东西。如果在.entry之前添加.visible，例如：.visible .entry examples_2E_mandelbrot...会发生什么呢？我认为你需要添加.target sm_30。 - Robert Crovella

Answer 1

您之所以出现错误是因为您的PTX包含语法错误，从而无法编译。导致错误的行为：

mov.b64 func_retval0, %rl0;

引用了一个标签func_retval0，但在PTX文件中并没有定义。您可以通过尝试使用工具链自己编译PTX来检查此问题：

$ ptxas -arch=sm_20 own.ptx 
ptxas own.ptx, line 24; error   : Arguments mismatch for instruction 'mov'
ptxas own.ptx, line 24; error   : Unknown symbol 'func_retval0'
ptxas own.ptx, line 24; error   : Label expected for forward reference of 'func_retval0'
ptxas fatal   : Ptx assembly aborted due to errors

- talonmies

---

是的！我的函数应该被编译为返回void。更改后问题得到解决。谢谢！ - Timothy Baldridge

Answer 2

关于运行ptxas的好建议。我遇到了209错误：问题是__shared__内存超额使用。以前这只是编译时的警告。我使用的是Cuda 5.5，现在即使打开详细模式也没有编译警告。谢谢。

- JPM