4.1. CUDA Host API

4.1.1. Device Management

4.1.1.1. Device detection and enquiry

The following functions are available for querying the available hardware:

numba.cuda.is_available()

Returns a boolean to indicate the availability of a CUDA GPU.

This will initialize the driver if it hasn’t been initialized.

numba.cuda.detect()

Detect supported CUDA hardware and print a summary of the detected hardware.

Returns a boolean indicating whether any supported devices were detected.

4.1.1.2. Context management

CUDA Python functions execute within a CUDA context. Each CUDA device in a system has an associated CUDA context, and Numba presently allows only one context per thread. For further details on CUDA Contexts, refer to the CUDA Driver API Documentation on Context Management and the CUDA C Programming Guide Context Documentation. CUDA Contexts are instances of the Context class:

class numba.cuda.cudadrv.driver.Context(device, handle)

This object wraps a CUDA Context resource.

Contexts should not be constructed directly by user code.

get_memory_info()

Returns (free, total) memory in bytes in the context.

pop()

Pops this context off the current CPU thread. Note that this context must be at the top of the context stack, otherwise an error will occur.

push()

Pushes this context on the current CPU Thread.

reset()

Clean up all owned resources in this context.

The following functions can be used to get or select the context:

numba.cuda.current_context(devnum=0)

Get the current device or use a device by device number, and return the CUDA context.

numba.cuda.require_context(fn)

A decorator that ensures a CUDA context is available when fn is executed.

Decorating fn is equivalent to writing:

get_context()
fn()

at each call site.

The following functions affect the current context:

numba.cuda.synchronize()

Synchronize the current context.

numba.cuda.close()

Explicitly clears all contexts in the current thread, and destroys all contexts if the current thread is the main thread.

4.1.1.3. Device management

Numba maintains a list of supported CUDA-capable devices:

numba.cuda.gpus

An indexable list of supported CUDA devices. This list is indexed by integer device ID.

Alternatively, the current device can be obtained:

numba.cuda.gpus.current()

Return the currently-selected device.

Getting a device through numba.cuda.gpus always provides an instance of numba.cuda.cudadrv.devices._DeviceContextManager, which acts as a context manager for the selected device:

class numba.cuda.cudadrv.devices._DeviceContextManager(device)

Provides a context manager for executing in the context of the chosen device. The normal use of instances of this type is from numba.cuda.gpus. For example, to execute on device 2:

with numba.cuda.gpus[2]:
    d_a = numba.cuda.to_device(a)

to copy the array a onto device 2, referred to by d_a.

One may also select a context and device or get the current device using the following three functions:

numba.cuda.select_device(device_id)

Make the context associated with device device_id the current context.

Returns a Device instance.

Raises exception on error.

numba.cuda.get_current_device()

Get current device associated with the current thread

numba.cuda.list_devices()

Return a list of all detected devices

The numba.cuda.cudadrv.driver.Device class can be used to enquire about the functionality of the selected device:

class numba.cuda.cudadrv.driver.Device

The device associated with a particular context.

compute_capability

A tuple, (major, minor) indicating the supported compute capability.

id

The integer ID of the device.

name

The name of the device (e.g. “GeForce GTX 970”)

reset()

Delete the context for the device. This will destroy all memory allocations, events, and streams created within the context.

4.1.2. Measurement

4.1.2.1. Profiling

The NVidia Visual Profiler can be used directly on executing CUDA Python code - it is not a requirement to insert calls to these functions into user code. However, these functions can be used to allow profiling to be performed selectively on specific portions of the code. For further information on profiling, see the NVidia Profiler User’s Guide.

numba.cuda.profile_start(*args, **kws)

Enable profile collection in the current context.

numba.cuda.profile_stop(*args, **kws)

Disable profile collection in the current context.

numba.cuda.profiling(*args, **kws)

Context manager that enables profiling on entry and disables profiling on exit.

4.1.2.2. Events

Events can be used to monitor the progress of execution and to record the timestamps of specific points being reached. Event creation returns immediately, and the created event can be queried to determine if it has been reached. For further information, see the CUDA C Programming Guide Events section.

The following functions are used for creating and measuring the time between events:

numba.cuda.event(timing=True)

Create a CUDA event. Timing data is only recorded by the event if it is created with timing=True.

numba.cuda.event_elapsed_time(evtstart, evtend)

Compute the elapsed time between two events in milliseconds.

Events are instances of the numba.cuda.cudadrv.driver.Event class:

class numba.cuda.cudadrv.driver.Event(context, handle, finalizer=None)

query()

Returns True if all work before the most recent record has completed; otherwise, returns False.

record(stream=0)

Set the record point of the event to the current point in the given stream.

The event will be considered to have occurred when all work that was queued in the stream at the time of the call to record() has been completed.

synchronize()

Synchronize the host thread for the completion of the event.

wait(stream=0)

All future works submitted to stream will wait util the event completes.

4.1.3. Stream Management

Streams allow concurrency of execution on a single device within a given context. Queued work items in the same stream execute sequentially, but work items in different streams may execute concurrently. Most operations involving a CUDA device can be performed asynchronously using streams, including data transfers and kernel execution. For further details on streams, see the CUDA C Programming Guide Streams section.

To create a stream:

numba.cuda.stream()

Create a CUDA stream that represents a command queue for the device.

Streams are instances of numba.cuda.cudadrv.driver.Stream:

class numba.cuda.cudadrv.driver.Stream(context, handle, finalizer)

auto_synchronize(*args, **kwds)

A context manager that waits for all commands in this stream to execute and commits any pending memory transfers upon exiting the context.

synchronize()

Wait for all commands in this stream to execute. This will commit any pending memory transfers.