CUDA编程之CUDA上下文

一、CUDA上下文的基本概念

1. 什么是CUDA上下文？

   CUDA上下文是一个环境，它定义了如何在GPU上执行CUDA代码。每个上下文都与一个特定的设备相关联，并且可以包含多个流（Streams），这些流用于管理并发执行的任务。

2. 上下文的作用

   • 资源管理：上下文管理着GPU上的内存分配、内核执行等资源。

   • 状态管理：它包含了执行CUDA代码所需的状态信息，如当前执行的流、内存分配情况等。

   • 并发控制：通过使用多个流，可以在同一个上下文中实现任务的并发执行。

二、创建与销毁‌

• ‌显式创建（驱动API）‌：
  • 使用cuCtxCreate(&ctx, flags, device)创建，需手动销毁cuCtxDestroy(ctx)。

    #include <cuda.h>
    
    int main() {
        // 初始化驱动API
        cuInit(0);  // 参数为保留标志位，必须为0‌
    
        // 获取设备句柄
        CUdevice device;
        cuDeviceGet(&device, 0);  // 获取索引为0的GPU设备‌
    
        // 创建上下文
        CUcontext ctx;
        cuCtxCreate(&ctx, CU_CTX_SCHED_SPIN, device);  // 设置调度策略为主动轮询‌
    
        // 后续操作（如内存分配、内核启动）...
        
        // 销毁上下文
        cuCtxDestroy(ctx);  // 显式释放资源‌
        return 0;
    }
    

    关键参数说明‌：
    ‌flags‌：控制上下文行为，常用值包括：
    CU_CTX_SCHED_SPIN：主动轮询任务状态，适合低延迟场景‌
    CU_CTX_SCHED_YIELD：任务阻塞时主动释放CPU资源，适合多任务共享GPU
    CU_CTX_MAP_HOST：允许主机内存映射到设备地址空间（需设备支持）
    ‌device‌：通过cuDeviceGet获取的设备句柄，对应物理GPU索引
  • 支持上下文堆栈操作：cuCtxPushCurrent(ctx)和cuCtxPopCurrent(&ctx)。
    1）详细方法说明
    ‌cuCtxPushCurrent(ctx)‌：将指定上下文压入当前线程的上下文堆栈顶部，使其成为当前活动上下文。所有后续CUDA操作均基于此上下文执行‌。
    ‌cuCtxPopCurrent(&ctx)‌：从堆栈顶部弹出当前上下文，恢复前一个上下文为活动状态。弹出的上下文通过参数返回，便于后续管理‌。
    2）上下文堆栈机制‌
    每个线程维护独立的上下文堆栈，栈顶始终为当前上下文‌。
    通过Push/Pop操作实现多上下文切换，适用于需临时切换设备或隔离任务的场景‌。

    CUcontext ctx1, ctx2;
    cuCtxCreate(&ctx1, 0, device0);  // 创建并自动压栈ctx1‌
    cuCtxPushCurrent(ctx2);          // 显式压入ctx2，当前上下文变为ctx2
    // 执行需ctx2的CUDA操作
    cuCtxPopCurrent(&ctx2);          // 弹出ctx2，恢复ctx1为当前上下文
    

  • cuCtxPushCurrent/cuCtxPopCurrent与cuCtxSetCurrent对比
    cuCtxSetCurrent(ctx)：直接设置当前上下文，不维护堆栈结构，适用于单次切换‌。
    Push/Pop：维护堆栈状态，适合嵌套式上下文切换（如函数内临时切换）‌。

• ‌隐式创建（运行时API）‌：
  • 首次调用运行时函数（如cudaMalloc）时自动创建设备的‌主上下文‌。
  • 主上下文通过引用计数管理，所有线程共享，最后一个线程释放后销毁。

三、上下文与设备关系‌

• ‌一对一关联‌：每个上下文绑定到特定GPU设备。
• ‌多上下文支持‌：一个设备可创建多个上下文（如不同线程或任务隔离）。
• ‌多GPU编程‌：通过cudaSetDevice()或cuCtxSetCurrent()切换当前上下文。
  在CUDA编程中，经常需要管理多个GPU设备或上下文（contexts），特别是在一个应用程序中需要同时使用多个GPU时。cudaSetDevice()和cuCtxSetCurrent()是CUDA提供的两个API，用于在不同的GPU设备或上下文之间切换。下面是如何使用这两个函数来切换当前上下文。

  示例：
  首先，你需要通过cuDeviceGet()和cuCtxCreate()创建和管理上下文：

  CUdevice device;
  CUcontext ctx;
  cuInit(0);  // 初始化CUDA驱动
  cuDeviceGet(&device, 0);  // 获取设备0
  cuCtxCreate(&ctx, 0, device);  // 创建上下文

  然后，使用cuCtxSetCurrent()切换到这个上下文：

  cuCtxSetCurrent(ctx);  // 将创建的上下文设置为当前上下文

四、上下文与多个流

在同一个上下文中，你可以创建多个流来并发执行任务，这可以通过cudaStreamCreate实现：

cudaStream_t stream1, stream2;
cudaStreamCreate(&stream1);
cudaStreamCreate(&stream2);

 然后，你可以在核函数调用时指定不同的流：

kernel<<<blocks, threads, 0, stream1>>>(); // 使用stream1执行核函数
kernel<<<blocks, threads, 0, stream2>>>(); // 使用stream2执行核函数

五、多线程管理‌

• ‌主上下文共享‌：运行时API中，同一设备的主上下文被所有线程共享，自动附加。
• ‌线程局部上下文（驱动API）‌：
  • 每个线程维护独立的上下文堆栈，需显式绑定（cuCtxSetCurrent）。
  • 多线程访问同一上下文需同步（CUDA API非线程安全）。

六、资源共享与同步

• ‌同一上下文‌：资源（内存、流等）直接共享。
• ‌跨上下文共享‌：
  • 使用‌IPC内存句柄‌（cuIpcGetMemHandle, cuIpcOpenMemHandle）。
  • 需设备支持统一寻址（UVA）和Peer-to-Peer访问。
• ‌同步机制‌：事件和流可用于跨上下文同步（需相同设备）。