cupy.RawModule#

class cupy.RawModule(unicode code=None, *, unicode path=None, tuple options=(), unicode backend=u'nvrtc', bool translate_cucomplex=False, bool enable_cooperative_groups=False, name_expressions=None, bool jitify=False)[源代码]#

用户定义的自定义模块。

此类可用于编译原始CUDA源代码或加载CUDA模块（*.cubin, *.ptx）。当需要获取同一源代码中的多个CUDA内核时，此类非常有用。

对于前一种情况，当调用任何方法时，CUDA源代码会被编译。对于后一种情况，可以通过提供其路径来加载现有的CUDA二进制文件（*.cubin）或PTX文件。

在 RawModule 中的 CUDA 内核可以通过调用 get_function() 来获取，这将返回一个 RawKernel 的实例。（与在 RawKernel 中一样，生成的二进制文件也会被缓存。）

参数:

code (str) – CUDA 源代码。与 path 互斥。
path (str) – cubin/ptx 的路径。与 code 互斥。
options (tuple of str) – 传递给后端（NVRTC 或 NVCC）的编译器选项。详情请参阅 https://docs.nvidia.com/cuda/nvrtc/index.html#group__options 或 https://docs.nvidia.com/cuda/cuda-compiler-driver-nvcc/index.html#command-option-description。
backend (str) – 要么是 nvrtc 要么是 nvcc。默认是 nvrtc。
translate_cucomplex (bool) – CUDA 源代码是否包含头文件 cuComplex.h。如果设置为 True，任何使用 cuComplex.h 中函数的代码将被翻译为其 Thrust 对应部分。默认为 False。
enable_cooperative_groups (bool) – 是否在CUDA源代码中启用协作组。如果设置为``True``，编译选项将正确配置，并且内核将使用``cuLaunchCooperativeKernel``启动，以便可以从CUDA源代码中使用协作组。此功能仅在CUDA 9或更高版本中受支持。
name_expressions (sequence of str) – 一个字符串序列（例如列表），引用C++全局/模板内核的名称。例如，对于模板内核 func1<T> 和非模板内核 func2，可以使用 name_expressions=['func1<int>', 'func1<double>', 'func2']。这个元组中的字符串必须逐个传递给 get_function() 以检索相应的内核。
jitify (bool) – 是否使用 Jitify 来协助 NVRTC 编译 C++ 内核。默认为 False。

备注

从 CuPy v13.0.0 开始，如果未在 options 中指定，RawModule 默认使用 C++11 标准 (-std=c++11) 进行编译。

备注

RawModule 中的每个内核都具有独立的功能属性。

备注

在 CuPy v8.0.0 之前，编译发生在初始化时。现在，它发生在第一次从模块中检索任何对象（内核或指针）时。

方法

compile(self, log_stream=None)#

编译当前模块。

通常情况下，您不需要调用此方法；内核在第一次调用时会隐式编译。

参数:: log_stream (object) – 将 sys.stdout 或一个文件对象传递给编译器输出将被写入的位置。默认为 None。

备注

调用 compile() 将重置 RawKernel 的内部状态。

get_function(self, unicode name)#

从模块中通过名称检索一个CUDA内核。

参数:: name (str) – 内核函数的名称。对于C++全局/模板内核，name 指的是在初始化当前 RawModule 实例时指定的名称表达式之一。
返回:: 一个 RawKernel 实例。
返回类型:: RawKernel

备注

以下示例展示了如何检索其中一个专门的 C++ 模板内核：

code = r'''
template<typename T>
__global__ void func(T* in_arr) { /* do something */ }
'''

kers = ('func<int>', 'func<float>', 'func<double>')
mod = cupy.RawModule(code=code, options=('--std=c++11',),
                     name_expressions=kers)

// retrieve func<int>
ker_int = mod.get_function(kers[0])

参见

来自 NVRTC 文档访问降低的名称的 nvrtcAddNameExpression 和 nvrtcGetLoweredName。

get_global(self, name)#

通过模块中的名称检索全局符号的指针。

参数:: name (str) – 全局符号的名称。
返回:: 全局符号的句柄。
返回类型:: MemoryPointer

备注

此方法可用于访问常量内存，例如：

# to get a pointer to "arr" declared in the source like this:
# __constant__ float arr[10];
memptr = mod.get_global("arr")
# ...wrap it using cupy.ndarray with a known shape
arr_ndarray = cp.ndarray((10,), cp.float32, memptr)
# ...perform data transfer to initialize it
arr_ndarray[...] = cp.random.random((10,), dtype=cp.float32)
# ...and arr is ready to be accessed by RawKernels