Writing Device Functions

OpenCL and SYCL do not directly have a notion for device-only functions, i.e. functions that can be only invoked from a kernel and not from a host function. However, the special decorator numba_dpex.func is provided specifically to implement device functions.

@dppy.func
def a_device_function(a):
    """
    A ``func`` is a device callable function that can be invoked from
    ``kernel`` and other ``func`` functions.
    """
    return a + 1

To use a device function from an another device function:

@dppy.func
def another_device_function(a):
    return a_device_function(a)

To use a device function from a kernel function numba_dpex.kernel:

@dppy.kernel
def a_kernel_function(a, b):
    i = dppy.get_global_id(0)
    b[i] = another_device_function(a[i])

Unlike a kernel function, a device function can return a value like normal functions.

Todo

Specific capabilities and limitations for device functions need to be added.