[PyCUDA] Error in NVCC compilation when options are given

classic Classic list List threaded Threaded
9 messages Options
Reply | Threaded
Open this post in threaded view
|

[PyCUDA] Error in NVCC compilation when options are given

Scott Biersdorff
Hi,

I would like to report a possible bug with pycuda. To show the issue
I've modified demo.py adding by passing a option to the nvcc compiler:

  mod = SourceModule("""
      __global__ void doublify(float *a)
      {
        int idx = threadIdx.x + threadIdx.y*4;
        a[idx] *= 2;
      }
      """, options=['-ccbin /usr/bin'])

However I get the following error when running demo.py:

python demo.py
*** compiler options is ['-ccbin /usr/bin', '-arch', 'sm_12',
'-I/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/../include/pycuda']
Traceback (most recent call last):
  File "demo.py", line 22, in <module>
    """, options=['-ccbin /usr/bin'])
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/compiler.py", line 239, in __init__
    arch, code, cache_dir, include_dirs)
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/compiler.py", line 229, in compile
    return compile_plain(source, options, keep, nvcc, cache_dir)
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/compiler.py", line 108, in compile_plain
    cmdline, stdout=stdout, stderr=stderr)
pycuda.driver.CompileError: nvcc compilation of /tmp/tmp4keOYP/kernel.cu
failed
[command: nvcc --cubin -ccbin /usr/bin -arch sm_12
-I/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/../include/pycuda kernel.cu]
[stderr:
nvcc fatal   : Unknown option 'ccbin /usr/bin'
]

Notice that the nvcc command listed in the output is valid and will
compile the kernel correctly, yet I receiving an error from pycuda.

I would welcome any suggestions or workarounds, I'm using pycuda 0.94.2
from source with python 2.7.1+ on Ubuntu 11.04.

Thanks,

- Scott






_______________________________________________
PyCUDA mailing list
[hidden email]
http://lists.tiker.net/listinfo/pycuda
Reply | Threaded
Open this post in threaded view
|

Re: Error in NVCC compilation when options are given

Andreas Kloeckner
On Mon, 20 Jun 2011 14:13:21 -0700, Scott Biersdorff <[hidden email]> wrote:

> Hi,
>
> I would like to report a possible bug with pycuda. To show the issue
> I've modified demo.py adding by passing a option to the nvcc compiler:
>
>   mod = SourceModule("""
>       __global__ void doublify(float *a)
>       {
>         int idx = threadIdx.x + threadIdx.y*4;
>         a[idx] *= 2;
>       }
>       """, options=['-ccbin /usr/bin'])
Pass this as options=['-ccbin', '/usr/bin'].

Andreas

_______________________________________________
PyCUDA mailing list
[hidden email]
http://lists.tiker.net/listinfo/pycuda

attachment0 (195 bytes) Download Attachment
Reply | Threaded
Open this post in threaded view
|

Re: Error in NVCC compilation when options are given

Scott Biersdorff
Hi,

Thanks that fixed the first compile step however the program then
compiles the kernel a second time without the options I specified (I've
added a print statement to show the options being passed onto nvcc):


%> python demo.py
*** compiler options is ['-ccbin', '/usr/bin/nvidia-gnu', '-arch',
'sm_12',
'-I/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/../include/pycuda']
original array:
[[-0.67095202  0.06165788 -0.12977113 -2.18319201]
 [ 0.08371061 -1.0201813  -0.10623035 -1.69607186]
 [ 0.34388322  1.23563218  0.19672823 -0.60138172]
 [ 1.73257256  0.36532807  0.8768267  -0.65081763]]
doubled with kernel:
[[-1.34190404  0.12331576 -0.25954226 -4.36638403]
 [ 0.16742122 -2.0403626  -0.2124607  -3.39214373]
 [ 0.68776643  2.47126436  0.39345646 -1.20276344]
 [ 3.46514511  0.73065615  1.75365341 -1.30163527]]
doubled with InOut:
[[-1.34190404  0.12331576 -0.25954226 -4.36638403]
 [ 0.16742122 -2.0403626  -0.2124607  -3.39214373]
 [ 0.68776643  2.47126436  0.39345646 -1.20276344]
 [ 3.46514511  0.73065615  1.75365341 -1.30163527]]
*** compiler options is ['-arch', 'sm_12',
'-I/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/../include/pycuda']
Traceback (most recent call last):
  File "demo.py", line 44, in <module>
    a_doubled = (2*a_gpu).get()
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/gpuarray.py", line 285, in __rmul__
    return self._axpbz(scalar, 0, result)
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/gpuarray.py", line 160, in _axpbz
    func = elementwise.get_axpbz_kernel(self.dtype)
  File "<string>", line 2, in get_axpbz_kernel
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/tools.py", line 485, in context_dependent_memoize
    result = func(*args)
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/elementwise.py", line 323, in get_axpbz_kernel
    "axpb")
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/elementwise.py", line 99, in get_elwise_kernel
    arguments, operation, name, keep, options, **kwargs)
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/elementwise.py", line 85, in get_elwise_kernel_and_types
    keep, options, **kwargs)
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/elementwise.py", line 74, in get_elwise_module
    options=options, keep=keep)
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/compiler.py", line 239, in __init__
    arch, code, cache_dir, include_dirs)
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/compiler.py", line 229, in compile
    return compile_plain(source, options, keep, nvcc, cache_dir)
  File
"/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/compiler.py", line 108, in compile_plain
    cmdline, stdout=stdout, stderr=stderr)
pycuda.driver.CompileError: nvcc compilation of /tmp/tmpoQTLvI/kernel.cu
failed
[command: nvcc --cubin -arch sm_12
-I/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/../include/pycuda kernel.cu]
[stderr:
In file included
from /usr/local/cuda/bin/../include/cuda_runtime.h:59:0,
                 from <command-line>:0:
/usr/local/cuda/bin/../include/host_config.h:82:2: error: #error --
unsupported GNU version! gcc 4.5 and up are not supported!


I'm I reading this right? Is the kernel being compiled twice?

Thanks,
- Scott


On Mon, 2011-06-20 at 17:36 -0400, Andreas Kloeckner wrote:

> On Mon, 20 Jun 2011 14:13:21 -0700, Scott Biersdorff <[hidden email]> wrote:
> > Hi,
> >
> > I would like to report a possible bug with pycuda. To show the issue
> > I've modified demo.py adding by passing a option to the nvcc compiler:
> >
> >   mod = SourceModule("""
> >       __global__ void doublify(float *a)
> >       {
> >         int idx = threadIdx.x + threadIdx.y*4;
> >         a[idx] *= 2;
> >       }
> >       """, options=['-ccbin /usr/bin'])
>
> Pass this as options=['-ccbin', '/usr/bin'].
>
> Andreas



_______________________________________________
PyCUDA mailing list
[hidden email]
http://lists.tiker.net/listinfo/pycuda
Reply | Threaded
Open this post in threaded view
|

Re: Error in NVCC compilation when options are given

Andreas Kloeckner
On Mon, 20 Jun 2011 14:55:46 -0700, Scott Biersdorff <[hidden email]> wrote:

> Traceback (most recent call last):
>   File "demo.py", line 44, in <module>
>     a_doubled = (2*a_gpu).get()
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/gpuarray.py", line 285, in __rmul__
>     return self._axpbz(scalar, 0, result)
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/gpuarray.py", line 160, in _axpbz
>     func = elementwise.get_axpbz_kernel(self.dtype)
>   File "<string>", line 2, in get_axpbz_kernel
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/tools.py", line 485, in context_dependent_memoize
>     result = func(*args)
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/elementwise.py", line 323, in get_axpbz_kernel
>     "axpb")
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/elementwise.py", line 99, in get_elwise_kernel
>     arguments, operation, name, keep, options, **kwargs)
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/elementwise.py", line 85, in get_elwise_kernel_and_types
>     keep, options, **kwargs)
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/elementwise.py", line 74, in get_elwise_module
>     options=options, keep=keep)
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/compiler.py", line 239, in __init__
>     arch, code, cache_dir, include_dirs)
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/compiler.py", line 229, in compile
>     return compile_plain(source, options, keep, nvcc, cache_dir)
>   File
> "/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/compiler.py", line 108, in compile_plain
>     cmdline, stdout=stdout, stderr=stderr)
> pycuda.driver.CompileError: nvcc compilation of /tmp/tmpoQTLvI/kernel.cu
> failed
> [command: nvcc --cubin -arch sm_12
> -I/usr/local/lib/python2.7/dist-packages/pycuda-0.94.2-py2.7-linux-x86_64.egg/pycuda/../include/pycuda kernel.cu]
> [stderr:
> In file included
> from /usr/local/cuda/bin/../include/cuda_runtime.h:59:0,
>                  from <command-line>:0:
> /usr/local/cuda/bin/../include/host_config.h:82:2: error: #error --
> unsupported GNU version! gcc 4.5 and up are not supported!
>
> I'm I reading this right? Is the kernel being compiled twice?
No. The evidence is in your stack trace, actually. The failing
compilation is trying to compile a kernel for an array operation (the
2*a in your case). I suppose we should add a way to pass options to nvcc
for behind-your-back compilations? A global variable maybe (shudder)?

Andreas

_______________________________________________
PyCUDA mailing list
[hidden email]
http://lists.tiker.net/listinfo/pycuda

attachment0 (195 bytes) Download Attachment
Reply | Threaded
Open this post in threaded view
|

Re: Error in NVCC compilation when options are given

Scott Biersdorff
On Mon, 2011-06-20 at 18:42 -0400, Andreas Kloeckner wrote:
>
> No. The evidence is in your stack trace, actually. The failing
> compilation is trying to compile a kernel for an array operation (the
> 2*a in your case). I suppose we should add a way to pass options to nvcc
> for behind-your-back compilations? A global variable maybe (shudder)?

That's what I ended up doing myself (Like you have done for
CUDA_DEBUGGING). However you might want a more flexible solution so you
don't end up dealing with all possible nvcc options as a separate global
variables. Also be warned that nvcc will not allow some options to set
more than once on the same command.

Thanks for your help,

- Scott



_______________________________________________
PyCUDA mailing list
[hidden email]
http://lists.tiker.net/listinfo/pycuda
Reply | Threaded
Open this post in threaded view
|

Re: Error in NVCC compilation when options are given

Andreas Kloeckner
On Mon, 20 Jun 2011 16:03:39 -0700, Scott Biersdorff <[hidden email]> wrote:

> On Mon, 2011-06-20 at 18:42 -0400, Andreas Kloeckner wrote:
> >
> > No. The evidence is in your stack trace, actually. The failing
> > compilation is trying to compile a kernel for an array operation (the
> > 2*a in your case). I suppose we should add a way to pass options to nvcc
> > for behind-your-back compilations? A global variable maybe (shudder)?
>
> That's what I ended up doing myself (Like you have done for
> CUDA_DEBUGGING). However you might want a more flexible solution so you
> don't end up dealing with all possible nvcc options as a separate global
> variables. Also be warned that nvcc will not allow some options to set
> more than once on the same command.
Added.

http://documen.tician.de/pycuda/driver.html#pycuda.compiler.DEFAULT_NVCC_FLAGS

Andreas

_______________________________________________
PyCUDA mailing list
[hidden email]
http://lists.tiker.net/listinfo/pycuda

attachment0 (195 bytes) Download Attachment
Reply | Threaded
Open this post in threaded view
|

Re: Error in NVCC compilation when options are given

gleb.budylin
This post has NOT been accepted by the mailing list yet.
I had somewhat related problem with the same example demo.py and probably found another bug.

Every function in elementwise.py with "options" argument uses [] as default value of options. This [] is passed to SourceModule constructor. SourceModule.__init__ calls compiler.compile(...) which sets options to DEFAULT_NVCC_FLAGS if options is None. But functions of elementwise.py pass an empty list instead of None, so DEFAULT_NVCC_FLAGS are ignored if the default value of options in elementwise.py were used.

I had DEFAULT_NVCC_FLAGS initialized by compiler.py using environment variable PYCUDA_DEFAULT_NVCC_FLAGS and found that my "--compiler-bindir=/usr/x86_64-pc-linux-gnu/gcc-bin/4.4.6" is ignored. After changing of all occurrences  "options=[]" to "options=None" in elementwise.py I have demo.py working (:
Reply | Threaded
Open this post in threaded view
|

Re: Error in NVCC compilation when options are given

gleb.budylin
In reply to this post by Andreas Kloeckner
I had somewhat related problem with the same example demo.py and probably found another bug.

Every function in elementwise.py with "options" argument uses [] as default value of options. This [] is passed to SourceModule constructor. SourceModule.__init__ calls compiler.compile(...) which sets options to DEFAULT_NVCC_FLAGS if options is None. But functions of elementwise.py pass an empty list instead of None, so DEFAULT_NVCC_FLAGS are ignored if the default value of options in elementwise.py were used.

I had DEFAULT_NVCC_FLAGS initialized by compiler.py using environment variable PYCUDA_DEFAULT_NVCC_FLAGS and found that my "--compiler-bindir=/usr/x86_64-pc-linux-gnu/gcc-bin/4.4.6" was ignored. After changing of all occurrences  "options=[]" to "options=None" in elementwise.py I have demo.py working (:
Reply | Threaded
Open this post in threaded view
|

Re: Error in NVCC compilation when options are given

Andreas Kloeckner
On Tue, 11 Oct 2011 02:55:05 -0700 (PDT), "gleb.budylin" <[hidden email]> wrote:

> I had somewhat related problem with the same example demo.py and probably
> found another bug.
>
> Every function in elementwise.py with "options" argument uses [] as default
> value of options. This [] is passed to SourceModule constructor.
> SourceModule.__init__ calls compiler.compile(...) which sets options to
> DEFAULT_NVCC_FLAGS if options is None. But functions of elementwise.py pass
> an empty list instead of None, so DEFAULT_NVCC_FLAGS are ignored if the
> default value of options in elementwise.py were used.
>
> I had DEFAULT_NVCC_FLAGS initialized by compiler.py using environment
> variable PYCUDA_DEFAULT_NVCC_FLAGS and found that my
> "--compiler-bindir=/usr/x86_64-pc-linux-gnu/gcc-bin/4.4.6" was ignored.
> After changing of all occurrences  "options=[]" to "options=None" in
> elementwise.py I have demo.py working (:
Thanks for the report. Done in git.

Andreas


_______________________________________________
PyCUDA mailing list
[hidden email]
http://lists.tiker.net/listinfo/pycuda

attachment0 (195 bytes) Download Attachment