CUDA Python 中支持的 Python 功能

CUDA 内置目标弃用通知

Numba 内置的 CUDA 目标已弃用，后续开发已移至 NVIDIA numba-cuda 包。请参阅 内置 CUDA 目标弃用和维护状态。

此页面列出了 CUDA Python 中支持的 Python 功能。这包括所有使用 @cuda.jit 以及其他针对 CUDA GPU 的高级 Numba 装饰器编译的内核和设备函数。

语言

执行模型

CUDA Python 直接映射到 CUDA 的单指令多线程执行 (SIMT) 模型。每条指令都由多个线程隐式并行执行。在此执行模型下，数组表达式用处不大，因为我们不希望多个线程执行相同的任务。相反，我们希望线程以协作方式执行任务。

有关详细信息，请参阅 CUDA 编程指南。

浮点错误模型

默认情况下，CUDA Python 内核使用 NumPy 错误模型执行。在此模型中，除以零不会引发异常，而是产生 inf、-inf 或 nan 的结果。这与正常的 Python 错误模型不同，后者在除以零时会引发 ZeroDivisionError。

启用调试时（通过向 @cuda.jit 装饰器传递 debug=True），将使用 Python 错误模型。这允许识别内核执行期间的除以零错误。

结构

不支持以下 Python 结构

• 异常处理 (try .. except, try .. finally)

• 上下文管理（with 语句）

• 推导式（列表、字典、集合或生成器推导式）

• 生成器（任何 yield 语句）

支持 raise 和 assert 语句，但有以下限制

• 它们只能在内核中使用，不能在设备函数中使用。

• 它们仅在将 debug=True 传递给 @cuda.jit 装饰器时才有效。这与 CUDA C/C++ 中 assert 关键字的行为类似，除非开启设备调试编译，否则该关键字会被忽略。

支持字符串、整数和浮点数的打印，但打印是异步操作——为了确保在内核启动后所有输出都已打印，必须调用 numba.cuda.synchronize()。省略对 synchronize 的调用是可接受的，但内核的输出可能会在其他后续驱动程序操作（例如后续内核启动、内存传输等）期间出现，或者在程序执行完成之前未能出现。最多可以将 32 个参数传递给 print 函数——如果传递更多参数，则会发出格式字符串并产生警告。这是由于 CUDA 打印中的一般限制，如 CUDA C++ 编程指南中关于打印限制的部分所述。

递归

支持自递归设备函数，但递归调用必须与对函数的初始调用具有相同的参数类型。例如，支持以下形式的递归

@cuda.jit("int64(int64)", device=True)
def fib(n):
    if n < 2:
        return n
    return fib(n - 1) + fib(n - 2)

（fib 函数始终具有 int64 参数），而以下形式不支持

# Called with x := int64, y := float64
@cuda.jit
def type_change_self(x, y):
    if x > 1 and y > 0:
        return x + type_change_self(x - y, y)
    else:
        return y

对 type_change_self 的外部调用提供 (int64, float64) 参数，但内部调用使用 (float64, float64) 参数（因为 x - y / int64 - float64 导致 float64 类型）。因此，此函数不受支持。

函数之间的相互递归（例如，函数 func1() 调用 func2()，而 func2() 又调用 func1()）不受支持。

注意

CUDA 中的调用栈大小通常非常有限，因此在 CUDA 设备上通过递归调用使其溢出比在 CPU 上更容易。

栈溢出将导致内核执行期间出现未指定启动失败 (ULF)。为了确定 ULF 是否由栈溢出引起，程序可以在 Compute Sanitizer 下运行，该工具会在栈溢出发生时明确指出。

内置类型

以下内置类型支持从 CPU nopython 模式继承。

• int

• float

• complex

• bool

• None

• tuple

• Enum, IntEnum

参阅 nopython 内置类型。

对 NumPy 数组中使用的字符序列（字节和 Unicode 字符串）也有非常有限的支持。请注意，此支持仅适用于 CUDA 11.2 及更高版本。

内置函数

支持以下内置函数

• abs()

• bool

• complex

• enumerate()

• float

• int: 仅支持单参数形式

• len()

• min(): 仅支持多参数形式

• max(): 仅支持多参数形式

• pow()

• range

• round()

• zip()

标准库模块

cmath

支持 cmath 模块中的以下函数

• cmath.acos()

• cmath.acosh()

• cmath.asin()

• cmath.asinh()

• cmath.atan()

• cmath.atanh()

• cmath.cos()

• cmath.cosh()

• cmath.exp()

• cmath.isfinite()

• cmath.isinf()

• cmath.isnan()

• cmath.log()

• cmath.log10()

• cmath.phase()

• cmath.polar()

• cmath.rect()

• cmath.sin()

• cmath.sinh()

• cmath.sqrt()

• cmath.tan()

• cmath.tanh()

math

支持 math 模块中的以下函数

• math.acos()

• math.asin()

• math.atan()

• math.acosh()

• math.asinh()

• math.atanh()

• math.cos()

• math.sin()

• math.tan()

• math.hypot()

• math.cosh()

• math.sinh()

• math.tanh()

• math.atan2()

• math.erf()

• math.erfc()

• math.exp()

• math.expm1()

• math.fabs()

• math.frexp()

• math.ldexp()

• math.gamma()

• math.lgamma()

• math.log()

• math.log2()

• math.log10()

• math.log1p()

• math.sqrt()

• math.remainder()

• math.pow()

• math.ceil()

• math.floor()

• math.copysign()

• math.fmod()

• math.modf()

• math.isnan()

• math.isinf()

• math.isfinite()

operator

支持 operator 模块中的以下函数

• operator.add()

• operator.and_()

• operator.eq()

• operator.floordiv()

• operator.ge()

• operator.gt()

• operator.iadd()

• operator.iand()

• operator.ifloordiv()

• operator.ilshift()

• operator.imod()

• operator.imul()

• operator.invert()

• operator.ior()

• operator.ipow()

• operator.irshift()

• operator.isub()

• operator.itruediv()

• operator.ixor()

• operator.le()

• operator.lshift()

• operator.lt()

• operator.mod()

• operator.mul()

• operator.ne()

• operator.neg()

• operator.not_()

• operator.or_()

• operator.pos()

• operator.pow()

• operator.rshift()

• operator.sub()

• operator.truediv()

• operator.xor()

NumPy 支持

由于 CUDA 编程模型，内核内部的动态内存分配效率低下且通常不需要。Numba 禁止任何内存分配功能。这会禁用大量 NumPy API。为了获得最佳性能，用户应编写代码，使每个线程一次处理一个元素。

支持的 NumPy 特性

• 访问 ndarray 属性 .shape、.strides、.ndim、.size 等。

• 索引和切片有效。

• 支持 ufuncs 的子集，但输出数组必须作为位置参数传入（请参阅 调用 NumPy UFunc）。请注意，ufuncs 在每个线程中顺序执行——ufuncs 在输入数组的元素之间没有自动并行化。

  支持以下 ufuncs

  • numpy.sin()

  • numpy.cos()

  • numpy.tan()

  • numpy.arcsin()

  • numpy.arccos()

  • numpy.arctan()

  • numpy.arctan2()

  • numpy.hypot()

  • numpy.sinh()

  • numpy.cosh()

  • numpy.tanh()

  • numpy.arcsinh()

  • numpy.arccosh()

  • numpy.arctanh()

  • numpy.deg2rad()

  • numpy.radians()

  • numpy.rad2deg()

  • numpy.degrees()

  • numpy.greater()

  • numpy.greater_equal()

  • numpy.less()

  • numpy.less_equal()

  • numpy.not_equal()

  • numpy.equal()

  • numpy.log()

  • numpy.log2()

  • numpy.log10()

  • numpy.logical_and()

  • numpy.logical_or()

  • numpy.logical_xor()

  • numpy.logical_not()

  • numpy.maximum()

  • numpy.minimum()

  • numpy.fmax()

  • numpy.fmin()

  • numpy.bitwise_and()

  • numpy.bitwise_or()

  • numpy.bitwise_xor()

  • numpy.invert()

  • numpy.bitwise_not()

  • numpy.left_shift()

  • numpy.right_shift()

不支持的 NumPy 功能

• 数组创建 API。

• 数组方法。

• 返回新数组的函数。

CFFI 支持

支持 cffi.FFI 对象的 from_buffer() 方法。这对于获取可传递给外部 C/C++/PTX 函数的指针很有用（请参阅 CUDA FFI 文档）。