5

I am trying to use cuda with the GNU multiple precision library (gmp). When I add gmp instructions like mpf_init() to my device code I get this compiler error: tlgmp.cu(37): error: calling a host function("__ gmpf_init") from a __ device__ /__ global__ function("histo") is not allowed.

Is it possible to redefine gmp instructions so that they can can be used in device code?

Anton Neururer
  • 53
  • 4
  • not at the library level, no. The library is compiled for x86 and doesn't include functions compiled for running on the device. Since it's open source, it should be possible to go through the library of functions at the source level, and create a new library (or whatever functions you need) that are populated with the necessary __ host__ __ device__ decorations and perhaps other changes, then pass that code through nvcc to generate device-callable functions. – Robert Crovella Nov 17 '12 at 19:17
  • Just googling around, [this CUMP library](http://www.hpcs.cs.tsukuba.ac.jp/~nakayama/cump/) popped up when I googled cuda gmp. May be worth taking a look. – Robert Crovella Nov 17 '12 at 20:58
  • I have already found this library myself and taken a look at the example file axpy.cu. This example includes gmp.h and also uses standard gmp instructions in device code (mpf_mul(), mpf_add()). Therefore it should produce the same error message unless the author used a modified version of the gmp library. – Anton Neururer Nov 17 '12 at 21:25
  • 2
    In that file (axpy.cu) note the presence of the __using namespace cump;__ in the cump_axpy_kernel. This cump [namespace](http://www.cprogramming.com/tutorial/namespaces.html) overrides the gmp library and directs the compiler to use the equivalent functions from the cump library instead. gmp::mpf_mul is different than cump::mpf_mul, for example. – Robert Crovella Nov 17 '12 at 21:42
  • @RobertCrovella : This might be off-topic, but do you know a CUMP equivalent for non-nVidia cards? – user2284570 Apr 11 '14 at 22:18

1 Answers1

5

The GMP library is compiled for the host, and so it can't be used directly in device code. That is the direct reason for the error you are seeing.

Since it's an open-source library, it might be possible with some effort to go through the code and create your own version, that has appropriate __device__ decorators (and possibly other changes) to the various functions you need. This would probably require a substantial amount of work, however.

Another alternative might be to investigate the CUMP library.

Another alternative might be to investigate the xmp library

Another alternative might be to investigate the campary library

Robert Crovella
  • 143,785
  • 11
  • 213
  • 257