CUDA Runtime API :: CUDA Toolkit Documentation

6.8. 执行控制

本节介绍CUDA运行时应用程序编程接口的执行控制函数。

部分函数在C++ API Routines模块中提供了重载的C++ API模板版本，相关文档单独列出。

Functions

__host__ __device__ cudaError_t cudaFuncGetAttributes ( cudaFuncAttributes* attr, const void* func ): Find out attributes for a given function.
__host__ cudaError_t cudaFuncGetName ( const char** name, const void* func ): Returns the function name for a device entry function pointer.
__host__ cudaError_t cudaFuncGetParamInfo ( const void* func, size_t paramIndex, size_t* paramOffset, size_t* paramSize ): Returns the offset and size of a kernel parameter in the device-side parameter layout.
__host__ cudaError_t cudaFuncSetAttribute ( const void* func, cudaFuncAttribute attr, int value ): Set attributes for a given function.
__host__ cudaError_t cudaFuncSetCacheConfig ( const void* func, cudaFuncCache cacheConfig ): Sets the preferred cache configuration for a device function.
__device__ void* cudaGetParameterBuffer ( size_t alignment, size_t size ): Obtains a parameter buffer.
__device__ void cudaGridDependencySynchronize ( void ): Programmatic grid dependency synchronization.
__host__ cudaError_t cudaLaunchCooperativeKernel ( const void* func, dim3 gridDim, dim3 blockDim, void** args, size_t sharedMem, cudaStream_t stream ): Launches a device function where thread blocks can cooperate and synchronize as they execute.
__host__ cudaError_t cudaLaunchCooperativeKernelMultiDevice ( cudaLaunchParams* launchParamsList, unsigned int numDevices, unsigned int flags = 0 ): Launches device functions on multiple devices where thread blocks can cooperate and synchronize as they execute.
__device__ cudaError_t cudaLaunchDevice ( void* func, void* parameterBuffer, dim3 gridDimension, dim3 blockDimension, unsigned int sharedMemSize, cudaStream_t stream ): Launches a specified kernel.
__host__ cudaError_t cudaLaunchHostFunc ( cudaStream_t stream, cudaHostFn_t fn, void* userData ): Enqueues a host function call in a stream.
__host__ cudaError_t cudaLaunchKernel ( const void* func, dim3 gridDim, dim3 blockDim, void** args, size_t sharedMem, cudaStream_t stream ): Launches a device function.
__host__ cudaError_t cudaLaunchKernelExC ( const cudaLaunchConfig_t* config, const void* func, void** args ): Launches a CUDA function with launch-time configuration.
__host__ cudaError_t cudaSetDoubleForDevice ( double* d ): Converts a double argument to be executed on a device.
__host__ cudaError_t cudaSetDoubleForHost ( double* d ): Converts a double argument after execution on a device.
__device__ void cudaTriggerProgrammaticLaunchCompletion ( void ): Programmatic dependency trigger.

Functions

__host__ __device__ cudaError_t cudaFuncGetAttributes ( cudaFuncAttributes* attr, const void* func )

查找给定函数的属性。

参数

attr: - Return pointer to function's attributes
func: - Device function symbol

cudaSuccess, cudaErrorInvalidDeviceFunction

描述

此函数获取通过func指定的函数属性。func是一个设备函数符号，必须声明为__global__函数。获取的属性将存入attr中。如果指定的函数不存在，则假定其为cudaKernel_t并直接使用。对于模板函数，请按以下格式传递函数符号：func_name

请注意，某些函数属性（如maxThreadsPerBlock）可能会根据当前使用的设备而有所不同。

Note:

请注意，此函数也可能返回之前异步启动的错误代码。
在CUDA 4.1中已弃用将函数名称字符串作为func参数的做法，并在CUDA 5.0中移除了该功能。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。
该API也可与内核cudaKernel_t一起使用，通过cudaLibraryGetKernel()或cudaGetKernel查询句柄，然后将其转换为void*传递给API。传递给cudaGetKernel的符号entryFuncAddr应是在同一CUDA Runtime实例中注册的符号。
传递属于不同运行时实例的符号将导致未定义行为。唯一可以可靠传递到不同运行时实例的类型是 cudaKernel_t

另请参阅：

cudaFuncSetCacheConfig ( C API), cudaFuncGetAttributes ( C++ API), cudaLaunchKernel ( C API), cuFuncGetAttribute

__host__ cudaError_t cudaFuncGetName ( const char** name, const void* func )

返回设备入口函数指针对应的函数名称。

参数

name: - The returned name of the function
func: - The function pointer to retrieve name for

cudaSuccess, cudaErrorInvalidValue, cudaErrorInvalidDeviceFunction

描述

返回与符号func关联的函数名称到**name中。函数名称以空终止字符串形式返回。如果函数未声明为具有C链接，此API可能返回一个混淆名称。如果**name为NULL，则返回cudaErrorInvalidValue。如果func不是设备入口函数，则假定其为cudaKernel_t并直接使用。

Note:

请注意，此函数也可能返回之前异步启动的错误代码。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。
该API也可与内核cudaKernel_t一起使用，通过cudaLibraryGetKernel()或cudaGetKernel查询句柄，然后将其转换为void*传递给API。传递给cudaGetKernel的符号entryFuncAddr应是在同一CUDA Runtime实例中注册的符号。
传递属于不同运行时实例的符号将导致未定义行为。唯一可以可靠传递到不同运行时实例的类型是 cudaKernel_t

cudaFuncGetName (C++ API)

__host__ cudaError_t cudaFuncGetParamInfo ( const void* func, size_t paramIndex, size_t* paramOffset, size_t* paramSize )

返回设备端参数布局中内核参数的偏移量和大小。

参数

func: - The function to query
paramIndex: - The parameter index to query
paramOffset: - The offset into the device-side parameter layout at which the parameter resides
paramSize: - The size of the parameter in the device-side parameter layout

CUDA_SUCCESS, CUDA_ERROR_INVALID_VALUE,

描述

查询func参数列表中位于paramIndex处的内核参数，并通过paramOffset和paramSize返回参数信息。paramOffset返回该参数在设备端参数布局中的偏移量。paramSize返回该参数的字节大小。这些信息可用于通过cudaGraphKernelNodeSetParam()和cudaGraphKernelNodeUpdatesApply()从设备端更新内核节点参数。paramIndex必须小于func所接受的参数数量。

Note:

请注意，此函数也可能返回之前异步启动的错误代码。
该API也可与内核cudaKernel_t一起使用，通过cudaLibraryGetKernel()或cudaGetKernel查询句柄，然后将其转换为void*传递给API。传递给cudaGetKernel的符号entryFuncAddr应是在同一CUDA Runtime实例中注册的符号。
传递属于不同运行时实例的符号将导致未定义行为。唯一可以可靠传递到不同运行时实例的类型是 cudaKernel_t

__host__ cudaError_t cudaFuncSetAttribute ( const void* func, cudaFuncAttribute attr, int value )

为指定函数设置属性。

参数

func: - Function to get attributes of
attr: - Attribute to set
value: - Value to set

cudaSuccess, cudaErrorInvalidDeviceFunction, cudaErrorInvalidValue

描述

此函数用于设置通过func指定的函数属性。参数func必须是指向设备上执行函数的指针。由func指定的参数必须声明为__global__函数。将attr定义的枚举设置为value定义的值。如果指定的函数不存在，则假定为cudaKernel_t并按原样使用。如果无法写入指定属性，或值不正确，则返回cudaErrorInvalidValue。

attr的有效取值包括：

cudaFuncAttributeMaxDynamicSharedMemorySize - 请求的动态分配共享内存的最大字节数。该值与函数属性sharedSizeBytes的总和不能超过设备属性cudaDevAttrMaxSharedMemoryPerBlockOptin。可请求的动态共享内存最大容量可能因GPU架构而异。
cudaFuncAttributePreferredSharedMemoryCarveout - 在L1缓存和共享内存使用相同硬件资源的设备上，此属性设置共享内存的预留偏好比例（以总共享内存的百分比表示）。详见cudaDevAttrMaxSharedMemoryPerMultiprocessor。这仅是一个提示值，驱动程序可根据函数执行需要选择不同的比例。
cudaFuncAttributeRequiredClusterWidth: 所需的块集群宽度。宽度、高度和深度值必须全为0或全为正数。集群维度的有效性将在启动时检查。如果在编译时设置该值，则无法在运行时设置。在运行时设置将返回cudaErrorNotPermitted。
cudaFuncAttributeRequiredClusterHeight: 所需的块集群高度。宽度、高度和深度值必须全为0或全为正数。集群维度的有效性将在启动时检查。如果在编译时设置了该值，则无法在运行时设置。在运行时设置将返回cudaErrorNotPermitted。
cudaFuncAttributeRequiredClusterDepth: 所需的块集群深度。宽度、高度和深度值必须全为0或全为正数。集群维度的有效性将在启动时检查。如果在编译时设置该值，则无法在运行时设置。在运行时设置将返回cudaErrorNotPermitted错误。
cudaFuncAttributeNonPortableClusterSizeAllowed: 表示该函数是否允许使用非便携式集群大小启动。1表示允许，0表示禁止。
cudaFuncAttributeClusterSchedulingPolicyPreference: 函数的块调度策略。值类型为cudaClusterSchedulingPolicy。

Note:

请注意，此函数也可能返回之前异步启动的错误代码。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。
该API也可与内核cudaKernel_t一起使用，通过cudaLibraryGetKernel()或cudaGetKernel查询句柄，然后将其转换为void*传递给API。传递给cudaGetKernel的符号entryFuncAddr应是在同一CUDA Runtime实例中注册的符号。
传递属于不同运行时实例的符号将导致未定义行为。唯一可以可靠传递到不同运行时实例的类型是 cudaKernel_t

cudaLaunchKernel (C++ API), cudaFuncSetCacheConfig ( C++ API), cudaFuncGetAttributes ( C API),

__host__ cudaError_t cudaFuncSetCacheConfig ( const void* func, cudaFuncCache cacheConfig )

为设备函数设置首选缓存配置。

参数

func: - Device function symbol
cacheConfig: - Requested cache configuration

cudaSuccess, cudaErrorInvalidDeviceFunction

描述

在L1缓存和共享内存使用相同硬件资源的设备上，通过cacheConfig设置由func指定的函数的首选缓存配置。这只是一个偏好设置。如果可能，运行时将使用请求的配置，但若需要执行func，运行时可以自由选择不同的配置。

func 是一个设备函数符号，必须声明为 __global__ 函数。如果指定的函数不存在，则会返回 cudaErrorInvalidDeviceFunction。对于模板化函数，请按以下方式传递函数符号：func_name

在L1缓存和共享内存大小固定的设备上，此设置不起作用。

使用与最近偏好设置不同的偏好启动内核可能会插入一个设备端同步点。

支持的缓存配置包括：

cudaFuncCachePreferNone: 对共享内存或L1缓存无偏好（默认设置）
cudaFuncCachePreferShared: 优先使用更大的共享内存和更小的L1缓存
cudaFuncCachePreferL1: 优先使用更大的L1缓存和更小的共享内存
cudaFuncCachePreferEqual: 偏好L1缓存与共享内存大小相等

Note:

请注意，此函数也可能返回之前异步启动的错误代码。
在CUDA 4.1中已弃用将函数名称字符串作为func参数的做法，并在CUDA 5.0中移除了该功能。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。
此API不接受将cudaKernel_t强制转换为void*。如果需要对cudaKernel_t(或__global__函数)进行缓存配置修改，可以替换为调用cudaFuncSetAttributes函数，并使用属性cudaFuncAttributePreferredSharedMemoryCarveout来指定更细粒度的L1缓存和共享内存分配配置。

另请参阅：

cudaFuncSetCacheConfig ( C++ API), cudaFuncGetAttributes ( C API), cudaLaunchKernel ( C API), cuFuncSetCacheConfig

__device__ void* cudaGetParameterBuffer ( size_t alignment, size_t size )

获取参数缓冲区。

参数

alignment: - Specifies alignment requirement of the parameter buffer
size: - Specifies size requirement in bytes

返回指向已分配参数缓冲区的指针

描述

获取一个参数缓冲区，可用于填充内核启动所需的参数。传递给cudaLaunchDevice的参数必须通过此函数分配。

这是一个底层API，只能通过并行线程执行(PTX)访问。CUDA用户代码应使用<<< >>>来启动内核。

Note:

请注意，此函数也可能返回之前异步启动的错误代码。

另请参阅：

cudaLaunchDevice

__device__ void cudaGridDependencySynchronize ( void ) [inline]

以编程方式实现网格依赖同步。

描述

该设备函数将阻塞线程，直到所有直接网格依赖项都已完成。此API旨在与编程/启动事件/依赖项结合使用。更多信息请参阅cudaLaunchAttributeID::cudaLaunchAttributeProgrammaticStreamSerialization和cudaLaunchAttributeID::cudaLaunchAttributeProgrammaticEvent。

__host__ cudaError_t cudaLaunchCooperativeKernel ( const void* func, dim3 gridDim, dim3 blockDim, void** args, size_t sharedMem, cudaStream_t stream )

启动一个设备函数，线程块在执行过程中可以协作和同步。

参数

func: - Device function symbol
gridDim: - Grid dimentions
blockDim: - Block dimentions
args: - Arguments
sharedMem: - Shared memory
stream: - Stream identifier

描述

该函数在gridDim（gridDim.xgridDim.ygridDim.z）网格块上调用内核func。每个块包含blockDim（blockDim.xblockDim.yblockDim.z）个线程。

调用此内核的设备必须具有非零值的设备属性cudaDevAttrCooperativeLaunch。

启动的块总数不能超过每个多处理器最大块数（由cudaOccupancyMaxActiveBlocksPerMultiprocessor或cudaOccupancyMaxActiveBlocksPerMultiprocessorWithFlags返回）乘以设备属性cudaDevAttrMultiProcessorCount指定的多处理器数量。

该内核无法利用CUDA动态并行功能。

如果内核有N个参数，args应指向包含N个指针的数组。每个指针从args[0]到args[N - 1]，都指向内存中实际参数将被复制的区域。

对于模板化函数，按以下方式传递函数符号：func_name

sharedMem 设置每个线程块可用的动态共享内存量。

stream 指定调用关联到的流。

Note:

此函数使用标准的默认流语义。
请注意，此函数也可能返回之前异步启动的错误代码。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。
该API也可与内核cudaKernel_t一起使用，通过cudaLibraryGetKernel()或cudaGetKernel查询句柄，然后将其转换为void*传递给API。传递给cudaGetKernel的符号entryFuncAddr应是在同一CUDA Runtime实例中注册的符号。
传递属于不同运行时实例的符号将导致未定义行为。唯一可以可靠传递到不同运行时实例的类型是 cudaKernel_t

另请参阅：

cudaLaunchCooperativeKernel (C++ API), cudaLaunchCooperativeKernelMultiDevice, cuLaunchCooperativeKernel

__host__ cudaError_t cudaLaunchCooperativeKernelMultiDevice ( cudaLaunchParams* launchParamsList, unsigned int numDevices, unsigned int flags = 0 )

在多个设备上启动设备函数，线程块在执行过程中可以协作和同步。

参数

launchParamsList: - List of launch parameters, one per device
numDevices: - Size of the launchParamsList array
flags: - Flags to control launch behavior

已弃用

该函数自CUDA 11.3起已弃用。

描述

调用launchParamsList数组中指定的内核，其中数组的每个元素都包含执行单个内核启动所需的所有参数。这些内核在执行过程中可以协作和同步。数组的大小由numDevices指定。

无法在同一设备上启动两个内核。此多设备启动针对的所有设备必须完全相同。所有设备的设备属性cudaDevAttrCooperativeMultiDeviceLaunch必须具有非零值。

必须在所有设备上启动相同的kernel。请注意，任何__device__或__constant__变量都会在每个设备上独立实例化。应用程序有责任确保这些变量被正确初始化和使用。

所有启动的内核中，网格大小（以块为单位指定）、块本身的大小以及每个线程块使用的共享内存量也必须匹配。

用于启动这些内核的流必须通过cudaStreamCreate或cudaStreamCreateWithPriority或cudaStreamCreateWithPriority创建。不能使用NULL流或cudaStreamLegacy或cudaStreamPerThread。

每个内核启动的块总数不能超过cudaOccupancyMaxActiveBlocksPerMultiprocessor（或cudaOccupancyMaxActiveBlocksPerMultiprocessorWithFlags）返回的每个多处理器最大活动块数乘以设备属性cudaDevAttrMultiProcessorCount指定的多处理器数量。由于每个设备启动的块总数必须在所有设备上保持一致，因此每个设备可启动的最大块数将受限于具有最少多处理器数量的设备。

该内核无法利用CUDA动态并行功能。

cudaLaunchParams 结构体定义为：

‎        struct cudaLaunchParams
              {
                  void *func;
                  dim3 gridDim;
                  dim3 blockDim;
                  void **args;
                  size_t sharedMem;
                  cudaStream_t 
                  stream;
              };

where:

cudaLaunchParams::func 指定要启动的内核函数。必须在所有设备上启动相同的函数。对于模板化函数，请按以下方式传递函数符号：func_name
cudaLaunchParams::gridDim 指定网格的宽度、高度和深度（以块为单位）。所有启动的内核必须保持一致。
cudaLaunchParams::blockDim 表示每个线程块的宽度、高度和深度。所有启动的内核必须保持该参数一致。
cudaLaunchParams::args 指定内核函数的参数。如果内核有N个参数，那么cudaLaunchParams::args应该指向一个包含N个指针的数组。每个指针，从cudaLaunchParams::args[0]到cudaLaunchParams::args[N - 1]，都指向内存中实际参数将被复制的区域。
cudaLaunchParams::sharedMem 是每个线程块的动态共享内存大小（以字节为单位）。所有启动的内核必须保持一致。
cudaLaunchParams::stream 是用于执行启动操作的流句柄。该流不能是NULL流、cudaStreamLegacy或cudaStreamPerThread。

默认情况下，内核在所有指定流中的所有先前工作完成之前不会在任何GPU上开始执行。此行为可以通过指定标志cudaCooperativeLaunchMultiDeviceNoPreSync来覆盖。当指定此标志时，每个内核将仅等待对应GPU流中的先前工作完成，然后才开始执行。

类似地，默认情况下，在所有GPU上的内核完成之前，任何后续推送到指定流中的工作都不会开始执行。可以通过指定标志cudaCooperativeLaunchMultiDeviceNoPostSync来覆盖此行为。当指定此标志时，任何后续推送到指定流中的工作将仅等待对应该流的GPU上启动的内核完成后就开始执行。

Note:

此函数使用标准的默认流语义。
请注意，此函数也可能返回之前异步启动的错误代码。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。

另请参阅：

cudaLaunchCooperativeKernel (C++ API), cudaLaunchCooperativeKernel, cuLaunchCooperativeKernelMultiDevice

__device__ cudaError_t cudaLaunchDevice ( void* func, void* parameterBuffer, dim3 gridDimension, dim3 blockDimension, unsigned int sharedMemSize, cudaStream_t stream )

启动指定的内核。

参数

func: - Pointer to the kernel to be launched
parameterBuffer: - Holds the parameters to the launched kernel. parameterBuffer can be NULL. (Optional)
gridDimension: - Specifies grid dimensions
blockDimension: - Specifies block dimensions
sharedMemSize: - Specifies size of shared memory
stream: - Specifies the stream to be used

cudaSuccess, cudaErrorInvalidDevice, cudaErrorLaunchMaxDepthExceeded, cudaErrorInvalidConfiguration, cudaErrorStartupFailure, cudaErrorLaunchPendingCountExceeded, cudaErrorLaunchOutOfResources

描述

使用指定的参数缓冲区启动指定的内核。参数缓冲区可以通过调用cudaGetParameterBuffer()获取。

这是一个底层API，只能通过并行线程执行(PTX)访问。CUDA用户代码应使用<<< >>>来启动内核。

Note:

请注意，此函数也可能返回之前异步启动的错误代码。

有关启动配置和参数布局的详细描述，请分别参阅CUDA编程指南中的执行配置和参数缓冲区布局部分。

另请参阅：

cudaGetParameterBuffer

__host__ cudaError_t cudaLaunchHostFunc ( cudaStream_t stream, cudaHostFn_t fn, void* userData )

在流中排队一个主机函数调用。

参数

stream
fn: - The function to call once preceding stream operations are complete
userData: - User-specified data to be passed to the function

cudaSuccess, cudaErrorInvalidResourceHandle, cudaErrorInvalidValue, cudaErrorNotSupported

描述

将一个主机函数加入流中排队执行。该函数将在当前已排队的工作完成后被调用，并会阻塞之后添加的工作。

主机函数不得进行任何CUDA API调用。尝试使用CUDA API可能会导致cudaErrorNotPermitted错误，但这不是必然的。主机函数不得执行任何可能依赖于未完成CUDA工作的同步操作，除非这些工作被明确要求提前运行。没有明确执行顺序要求的主机函数（例如在独立流中）将以未定义的顺序执行，并可能被序列化。

就统一内存管理而言，执行过程提供以下保证：

在函数执行期间，该流被视为空闲状态。因此，例如，该函数可以始终使用附加到其入队流的内存。
函数开始执行的效果等同于在同一个流中同步记录在函数之前的事件。因此，它会同步那些在函数执行前已被"连接"的流。
在任何流中添加设备工作不会使该流变为活动状态，直到所有先前的主机函数和流回调执行完毕。因此，例如，即使工作已添加到另一个流中，如果该工作通过事件被排序在函数调用之后，函数仍可能使用全局附加内存。
函数的完成不会导致流变为活动状态，除非如上所述。如果没有后续的设备工作，流将保持空闲状态，并且在连续的宿主函数或流回调之间没有设备工作时也将保持空闲。因此，例如，可以通过在流结束时从宿主函数发出信号来完成流同步。

请注意，与cuStreamAddCallback不同，如果CUDA上下文中发生错误，该函数将不会被调用。

Note:

此函数使用标准的默认流语义。
请注意，此函数也可能返回之前异步启动的错误代码。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。

另请参阅：

cudaStreamCreate, cudaStreamQuery, cudaStreamSynchronize, cudaStreamWaitEvent, cudaStreamDestroy, cudaMallocManaged, cudaStreamAttachMemAsync, cudaStreamAddCallback, cuLaunchHostFunc

__host__ cudaError_t cudaLaunchKernel ( const void* func, dim3 gridDim, dim3 blockDim, void** args, size_t sharedMem, cudaStream_t stream )

启动一个设备函数。

参数

func: - Device function symbol
gridDim: - Grid dimentions
blockDim: - Block dimentions
args: - Arguments
sharedMem: - Shared memory
stream: - Stream identifier

cudaSuccess, cudaErrorInvalidDeviceFunction, cudaErrorInvalidConfiguration, cudaErrorLaunchFailure, cudaErrorLaunchTimeout, cudaErrorLaunchOutOfResources, cudaErrorSharedObjectInitFailed, cudaErrorInvalidPtx, cudaErrorUnsupportedPtxVersion, cudaErrorNoKernelImageForDevice, cudaErrorJitCompilerNotFound, cudaErrorJitCompilationDisabled

描述

该函数在gridDim（gridDim.xgridDim.ygridDim.z）网格块上调用内核func。每个块包含blockDim（blockDim.xblockDim.yblockDim.z）个线程。

如果内核有N个参数，args应指向包含N个指针的数组。每个指针从args[0]到args[N - 1]，都指向内存中实际参数将被复制的区域。

对于模板化函数，按以下方式传递函数符号：func_name

sharedMem 设置每个线程块可用的动态共享内存量。

stream 指定调用关联到的流。

Note:

此函数使用标准的默认流语义。
请注意，此函数也可能返回之前异步启动的错误代码。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。
该API也可与内核cudaKernel_t一起使用，通过cudaLibraryGetKernel()或cudaGetKernel查询句柄，然后将其转换为void*传递给API。传递给cudaGetKernel的符号entryFuncAddr应是在同一CUDA Runtime实例中注册的符号。
传递属于不同运行时实例的符号将导致未定义行为。唯一可以可靠传递到不同运行时实例的类型是 cudaKernel_t

另请参阅：

cudaLaunchKernel (C++ API), cuLaunchKernel

__host__ cudaError_t cudaLaunchKernelExC ( const cudaLaunchConfig_t* config, const void* func, void** args )

使用启动时配置启动一个CUDA函数。

参数

config: - Launch configuration
func: - Kernel to launch
args: - Array of pointers to kernel parameters

描述

请注意，功能上等效的可变参数模板 cudaLaunchKernelEx 适用于 C++11 及更新版本。

在由config->gridDim (config->gridDim.xconfig->gridDim.yconfig->gridDim.z)定义的网格块上调用内核函数func。每个块包含config->blockDim (config->blockDim.xconfig->blockDim.yconfig->blockDim.z)个线程。

config->dynamicSmemBytes 设置每个线程块可用的动态共享内存量。

config->stream 指定了调用关联的流。

除了网格和块维度、动态共享内存大小以及流之外，还可以通过config:的以下两个字段提供更多配置：

config->attrs 是一个包含 config->numAttrs 个连续 cudaLaunchAttribute 元素的数组。如果 config->numAttrs 为零，则不考虑该指针的值。但在这种情况下，建议将指针设置为 NULL。config->numAttrs 表示填充 config->attrs 数组前 config->numAttrs 个位置的属性数量。

如果内核有N个参数，args应指向包含N个指针的数组。每个指针从args[0]到args[N - 1]，都指向内存中实际参数将被复制的区域。

注意：此函数如此命名是为了避免无意中调用模板化版本cudaLaunchKernelEx，该版本适用于接收单个void**或void*参数的核函数。

Note:

此函数使用标准的默认流语义。
请注意，此函数也可能返回之前异步启动的错误代码。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。
该API也可与内核cudaKernel_t一起使用，通过cudaLibraryGetKernel()或cudaGetKernel查询句柄，然后将其转换为void*传递给API。传递给cudaGetKernel的符号entryFuncAddr应是在同一CUDA Runtime实例中注册的符号。
传递属于不同运行时实例的符号将导致未定义行为。唯一可以可靠传递到不同运行时实例的类型是 cudaKernel_t

另请参阅：

cudaLaunchKernelEx(const cudaLaunchConfig_t *config, void (*kernel)(ExpTypes...), ActTypes &&... args) "cudaLaunchKernelEx (C++ API)", cuLaunchKernelEx

__host__ cudaError_t cudaSetDoubleForDevice ( double* d )

将双精度参数转换为在设备上执行。

参数

d: - Double to convert

cudaSuccess

已弃用

该函数自CUDA 7.5起已弃用

描述

如果设备不支持双精度算术运算，则将d的双精度值转换为内部浮点表示形式。如果设备原生支持双精度，则此函数不执行任何操作。

Note:

请注意，此函数也可能返回之前异步启动的错误代码。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。

另请参阅：

cudaFuncSetCacheConfig ( C API), cudaFuncGetAttributes ( C API), cudaSetDoubleForHost

__host__ cudaError_t cudaSetDoubleForHost ( double* d )

在设备上执行后转换双精度参数。

参数

d: - Double to convert

cudaSuccess

已弃用

该函数自CUDA 7.5起已弃用

描述

如果设备不支持双精度算术运算，则将d的双精度值从可能的内部浮点表示形式转换。如果设备原生支持双精度，则此函数不执行任何操作。

Note:

请注意，此函数也可能返回之前异步启动的错误代码。
请注意，如果此调用尝试初始化CUDA RT内部状态，该函数也可能返回cudaErrorInitializationError、cudaErrorInsufficientDriver或cudaErrorNoDevice。
请注意，根据cudaStreamAddCallback的规定，回调函数中不得调用任何CUDA函数。在这种情况下，可能会（但不保证）返回cudaErrorNotPermitted作为诊断信息。

另请参阅：

cudaFuncSetCacheConfig ( C API), cudaFuncGetAttributes ( C API), cudaSetDoubleForDevice

__device__ void cudaTriggerProgrammaticLaunchCompletion ( void ) [inline]

编程式依赖触发器。

描述

该设备函数确保程序化启动完成的边沿/事件得到满足。有关更多信息，请参阅cudaLaunchAttributeID::cudaLaunchAttributeProgrammaticStreamSerialization和cudaLaunchAttributeID::cudaLaunchAttributeProgrammaticEvent。事件/边沿触发仅当网格中的每个CTA至少退出或调用此函数一次时才会发生，否则在所有warp完成执行后但在网格完成前会自动触发。该触发仅启用次级内核的调度，本身不提供内存可见性保证。用户可以通过插入正确作用域的内存栅栏来强制实现内存可见性。