CUDA 如何处理 OpenCL 中的线程偏移量

Question

例如，我可以在 OpenCL 中启动线程 ID 上偏移的内核，因此当使用多个 GPU 时，第二个 GPU 线程可以直接从任意整数值开始：

GPU1: thread idx 0 to k
GPU2: thread idx k+1 to N

内核对于所有 GPU 来说看起来都完全相同：

__kernel(..)
{
   int threadId=get_global_id(0); // this starts from k+1 for GPU2
   ..
}

在驱动程序 API 中调用内核时，CUDA 中相当于什么？

OpenCL 有

cl_int clEnqueueNDRangeKernel ( cl_command_queue command_queue,
    cl_kernel kernel,
    cl_uint work_dim,
    const size_t *global_work_offset, // this can be given k+1
    const size_t *global_work_size,
    const size_t *local_work_size,
    cl_uint num_events_in_wait_list,
    const cl_event *event_wait_list,
    cl_event *event)

我是否必须自己在内核中嵌入偏移量，例如以下示例才能在多个 GPU 上使用外观最相似的内核？

__global__ void vecAdd(.., int * gpuParams)
{
    int offset=gpuParams[0];
    int threadId = offset + blockIdx.x * blockDim.x + threadIdx.x;
}

(这适用于扁平化的缓冲区访问和内核(

这个偏移量将在每次内核启动之前动态更改，所以我猜不能使用它 #define。

Cuda 驱动程序 api 有这个用于启动内核：

CUresult cuLaunchKernel ( CUfunction f, unsigned int  gridDimX, 
                unsigned int  gridDimY, unsigned int  gridDimZ, 
                unsigned int  blockDimX, unsigned int  blockDimY, 
                unsigned int  blockDimZ, unsigned int  sharedMemBytes, 
                CUstream hStream, void** kernelParams, void** extra )

Answer 1

CUDA 中没有等效的功能，但你也不需要它。

在标准 CUDA 模型中，您不能使用单个 API 调用在多个设备上启动内核(从技术上讲，您也不能使用单个 API 调用在多个设备上分配内存(。每个 GPU 实际上都是独立的。

NVIDIA 引入了多设备启动和同步方法，作为 CUDA 9 中新合作组功能的一部分，但这些与您所要求的完全不同。