Cuda error CUDA_ERROR_NO_BINARY_FOR_GPU

Question

I have some PTX code that fails to load. I'm running this on a 650M, with OSX. Other CUDA examples run fine on the system, but when loading the module I always get error 209: CUDA_ERROR_NO_BINARY_FOR_GPU

What am I missing?

 .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;
}

GT650M is a sm_30 gpu. What happens if you change .target sm_20 to .target sm_30 ? Or perhaps I should ask, how did you generate this ptx code? — Robert Crovella
– Robert Crovella, Commented Mar 2, 2013 at 0:16
The code is generated via llvm. I've cut down the code to the level you see above. I've tried target models of sm_10, sm_13, sm_30, and sm_35. All the same. Switching from .entry to .func allows the module to load, but then (of course) I can't find the function. — Timothy Baldridge
– Timothy Baldridge, Commented Mar 2, 2013 at 0:29
You might want to create something similar by compiling some code with nvcc -arch=sm_30 -ptx mymodule.cu and analyze the differences. What happens if you put .visible before .entry i.e. .visible .entry examples_2E_mandelbrot... And you will want .target sm_30, I think. — Robert Crovella
– Robert Crovella, Commented Mar 2, 2013 at 0:37

talonmies · Accepted Answer · 2013-03-02 01:25:16Z

6

You are getting the error because your PTX contains a syntax error and it is never compiling as a result. The line

mov.b64 func_retval0, %rl0;

references a label func_retval0, but that isn't defined in the PTX file anywhere. You can check this by trying to compile the PTX with the toolchain yourself:

$ 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

answered Mar 2, 2013 at 1:25

talonmies

72.8k35 gold badges204 silver badges296 bronze badges

Sign up to request clarification or add additional context in comments.

1 Comment

Timothy Baldridge Over a year ago

Yep! My function should be compiled as returning void. Changing that fixed the issue. Thanks!

talonmies · Accepted Answer · 2014-01-17 09:41:46Z

1

Great advice on running ptxas. I was getting error 209: Problem turned out to be __shared__ memory was oversubscribed. Used to be this would be a warning on compile. I have Cuda 5.5 and no warnings on compile now -- even with verbose turned on. thanks

edited Jan 17, 2014 at 9:41

talonmies

72.8k35 gold badges204 silver badges296 bronze badges

answered Nov 13, 2013 at 3:27

JPM

4651 gold badge5 silver badges16 bronze badges

Collectives™ on Stack Overflow

Cuda error CUDA_ERROR_NO_BINARY_FOR_GPU

2 Answers 2

1 Comment

Comments

Your Answer

Hot Network Questions

Collectives™ on Stack Overflow

2 Answers 2

1 Comment

Comments

Your Answer

Sign up or log in

Post as a guest

Related