从多个 MPI 进程调用 cudaMemcpy 的最可靠方法是什么？

Question

我正在开发一个库，该库为使用 CUDA 和 MPI 求解微分方程进行动态工作负载分配。我有许多节点，每个节点都有一个 NVIDIA GPU。当然，每个节点也有多个进程。该方程采用一定数量的输入（本例中为 6 个）并构建一个解决方案，该解决方案在 GPU 的全局内存中表示为一个数组。

我目前的策略是在每个节点上的根进程上分配输入数据缓冲区：

if (node_info.is_node_root_process)
{
    cudaMalloc(&gpu_input_buffer.u_buffer, totalsize);
    cudaMalloc(&gpu_input_buffer.v_buffer, totalsize);
}

然后，我希望每个进程单独调用 cudaMemcpy 以将输入数据复制到 GPU 全局内存中，每个进程都复制到此输入缓冲区中的不同位置。这样输入缓冲区在内存中是连续的，可以实现内存的合并。

我了解从多个进程（或线程）调用cudaMemcpy，调用将在设备上连续执行。这很好。

我想要做的是分享地址，例如gpu_input_buffer.u_buffer 指向每个进程。这样，每个进程都有一个偏移量process_gpu_io_offset，这样与该进程相关的数据就是gpu_input_buffer.u_buffer + process_gpu_io_offset 到gpu_input_buffer.u_buffer + process_gpu_io_offset + number_of_points - 1。

我已经读过，由于使用了虚拟寻址，因此通过 MPI 共享指针值是禁忌，但是由于所有 GPU 数据都驻留在单个内存空间中，并且由于 gpu_input_buffer.u_buffer 是设备指针，我认为这应该没问题.

这是实现我想要的可靠方法吗？

编辑：基于 CUDA 文档：

主机线程创建的任何设备内存指针或事件句柄都可以 被同一进程中的任何其他线程直接引用。它 然而，在这个过程之外是无效的，因此不能 由属于不同进程的线程直接引用。

这意味着我原来的方法是无效的。正如已经指出的那样，CUDA API 具有用于此目的的 IPC 内存句柄，但我找不到有关如何使用 MPI 共享它的任何信息。 cudaIpcMemHandle_t 的文档只是：

CUDA IPC 内存句柄

它没有提供任何信息来支持我需要做的事情。可以创建 MPI 派生类型并进行通信，但这需要我知道 cudaIpcMemHandle_t 的成员，而我不知道。

Answer 1

CUDA 运行时 API 特别支持在同一台机器上的进程之间共享内存区域（和事件）。就用那个吧！

这里是示例 sn-ps（使用我的 modern-C++ wrappers for the CUDA Runtime API）

主要流程：

auto buffer = cuda::memory::device::make_unique<unsigned char[]>(totalsize);
gpu_input_buffer.u_buffer = buffer.get(); // because it's a smart pointer
auto handle_to_share = cuda::memory::ipc::export_(gpu_input_buffer.u_buffer);
do_some_MPI_magic_here_to_share_the_handle(handle_to_share);

其他进程：

auto shared_buffer_handle = do_some_MPI_magic_here_to_get_the_shared_handle();
auto full_raw_buffer = cuda::memory::ipc::import<unsigned char>(shared_buffer_handle);
auto my_part_of_the_raw_buffer = full_raw_buffer + process_gpu_io_offset;

注意：如果您对句柄类型的确切布局非常好奇，这里是 CUDA 的driver_types.h 的摘录：

typedef __device_builtin__ struct __device_builtin__ cudaIpcMemHandle_st 
{
    char reserved[CUDA_IPC_HANDLE_SIZE];
} cudaIpcMemHandle_t;