Replacing the njit decorator with the cuda.jit decorator

Question

I have an Nvidia GPU, downloaded CUDA, and am trying to make use of it.

Say I have this code:

#@cuda.jit (Attempted fix #1)
#@cuda.jit(device = True) (Attempted fix #2)
#@cuda.jit(int32(int32,int32)) (Attempted fix #3)

@njit
def product(rho, theta):
    x = rho * (theta)
    return(x)
a = product(1,2)
print(a)

How do I make it work with the cuda.jit decorator instead of njit?

Things I've tried:

When I switch the decorator from @njit to @cuda.jit, I get: TypingError: No conversion from int64 to none for '$0.5', defined at None.

When I switch the decorator @cuda.jit(device = True), I get: TypeError: 'DeviceFunctionTemplate' object is not callable.

And when I specify the types for my inputs and outputs, and use the decorator @cuda.jit(int32(int32,int32)), I get: TypeError: CUDA kernel must have void return type.

Robert Crovella · Answer 1 · 2019-05-03T21:35:27.393

numba cuda kernels don't return anything. You must return results via parameters/arguments to the function. The starting point for this is usually some kind of numpy array. Here is an example:

$ cat t44.py
from numba import cuda
import numpy as np

@cuda.jit
def product(rho, theta, x):
        x[0] = rho * (theta)

x = np.ones(1,dtype=np.float32)
product(1,2,x)
print(x)
$ python t44.py
[ 2.]
$

There's potentially many other things that could be said; you may wish to avail yourself of the documentation linked above, or e.g. this tutorial. Usually you will want to handle problems much bigger than multiplying two scalars, before GPU computation will be interesting.

Also, numba provides other methods to access GPU computation, that don't depend on use of the @cuda.jit decorator. These methods, such as @vectorize are documented.

I'm also omitting any kernel launch configuration syntax on the call to product. This is legal in numba cuda, but it results in launching a kernel of 1 block that contains 1 thread. This works for this particular example, but this is basically a nonsensical way to use CUDA GPUs.

Replacing the njit decorator with the cuda.jit decorator

1 Answers1