nvcc ICEs on div by 0 kernel

Accelerated Computing CUDA CUDA Programming and Performance
1

@NVIDIA: If I compile the following code with an intentional integer div by 0 using ‘nvcc -ptx test-nvcc-divby0-ice.cu’:

__global__ void kfunc(float* data) {
  unsigned o = 0;
  float v = data[o];
  int i = 1 / o;
  data[o] = v + i;
}

then nvcc exits with status 1 and prints:

Signal: Floating point exception in Code_Expansion phase.
<input>(0): Error: Signal Floating point exception in phase Code_Expansion -- processing aborted
*** Internal stack backtrace:
    /usr/local/cuda-5.5/open64/lib//be() [0x6deb9f]
    /usr/local/cuda-5.5/open64/lib//be() [0x6df7e9]
    /usr/local/cuda-5.5/open64/lib//be() [0x6def3d]
    /usr/local/cuda-5.5/open64/lib//be() [0x6e0186]
    /lib/x86_64-linux-gnu/libc.so.6(+0x364a0) [0x2b169bf7b4a0]
    /usr/local/cuda-5.5/open64/lib//be() [0x59b8f1]
    /usr/local/cuda-5.5/open64/lib//be() [0x59bf24]
    /usr/local/cuda-5.5/open64/lib//be() [0x59c29f]
    /usr/local/cuda-5.5/open64/lib//be() [0x57020f]
    /usr/local/cuda-5.5/open64/lib//be() [0x578e51]
    /usr/local/cuda-5.5/open64/lib//be() [0x579233]
    /usr/local/cuda-5.5/open64/lib//be() [0x578be2]
    /usr/local/cuda-5.5/open64/lib//be() [0x57b3a4]
    /usr/local/cuda-5.5/open64/lib//be() [0x578fb8]
    /usr/local/cuda-5.5/open64/lib//be() [0x57de7b]
    /usr/local/cuda-5.5/open64/lib//be() [0x57e9a9]
    /usr/local/cuda-5.5/open64/lib//be() [0x57f08c]
    /usr/local/cuda-5.5/open64/lib//be() [0x55c7d2]
    /usr/local/cuda-5.5/open64/lib//be() [0x405443]
    /usr/local/cuda-5.5/open64/lib//be() [0x4061f1]
    /usr/local/cuda-5.5/open64/lib//be() [0x40752d]
    /lib/x86_64-linux-gnu/libc.so.6(__libc_start_main+0xed) [0x2b169bf6676d]
    /usr/local/cuda-5.5/open64/lib//be() [0x4038da]
nvopencc INTERNAL ERROR: /usr/local/cuda-5.5/open64/lib//be died due to signal 4
nvopencc ERROR: core dumped

I’m using nvcc 5.5 on Ubuntu Linux 12.10:
Built on Wed_Jul_17_18:36:13_PDT_2013
Cuda compilation tools, release 5.5, V5.5.0

(With 6rc I get the same problem.)

2

This looks like an internal compiler error, and those should not happen. Since you already have a self-contained repro code, it would be helpful if you could file a bug report using the form that is linked from the registered developer website. Thank you for your help.

3

done (#455687)