cuda.jit compiles a Python function with Numba for CUDA execution. Without options, the function acts as a GPU kernel: each invocation handles one element, writes through its arguments, and is launched across configured thread blocks. section 9.2.2
A basic kernel follows this execution model:
- Use cuda.grid to calculate the current thread's position. section 9.2.2
- Guard against threads whose positions exceed the array dimensions. section 9.2.2
- Launch with kernel[blocks_per_grid, threads_per_block](arguments), rather than calling the function normally. section 9.2.2
- Pass output arrays as arguments because ordinary kernel functions cannot return values. section 9.2.2
Doubling an array
A one-dimensional kernel uses one thread per array element: position = cuda.grid(1); if position < my_array.shape[0]: my_array[position] *= 2. For 1,000 elements, 50 blocks of 20 threads cover the array exactly; when coverage is not exact, bounds checks prevent invalid accesses.
Mandelbrot rendering
Use @cuda.jit(device=True) for a helper that computes one Mandelbrot point and returns its value. A separate two-dimensional kernel maps threads to image coordinates, fills a NumPy array, and launches a 2D grid rounded up with ceil(size / 16). section 9.2.3
