CUDA流未实现异步问题：基于NVIDIA 3060 Laptop的解密程序排查

我使用C语言在搭载5个异步引擎的NVIDIA Geforce 3060 Laptop（CUDA）平台上实现异或加密文件的解密功能，程序逻辑正常，但性能分析结果显示，不同CUDA流并未实现预期的异步并行效果（见下图）：

第二次尝试（锁页内存场景） #

第一次尝试（非锁页内存场景） #

核函数调用代码 #

__global__ void decryptKernel(unsigned char* buffer, unsigned char key, int size) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    if (idx < size) {
        buffer[idx] ^= 0x43;  
    }
}

void decrypt_large_shellcode(unsigned char* shellcode, unsigned char key) {
    unsigned char* d_buffers[NUM_STREAMS];
    cudaStream_t streams[NUM_STREAMS];

    unsigned char* pinned_shellcode;
    cudaHostAlloc((void**)&pinned_shellcode, TOTAL_SIZE, cudaHostAllocDefault);
    memcpy(pinned_shellcode, shellcode, TOTAL_SIZE); 
    for (int i = 0; i < NUM_STREAMS; i++) {
        cudaMalloc((void**)&d_buffers[i], CHUNK_SIZE); 
        cudaStreamCreate(&streams[i]);
    }

    for (int offset = 0; offset < TOTAL_SIZE; offset += CHUNK_SIZE * NUM_STREAMS) {
        for (int i = 0; i < NUM_STREAMS; i++) {
            int chunk_offset = offset + i * CHUNK_SIZE;
            if (chunk_offset >= TOTAL_SIZE) break;

            int chunk_size = (TOTAL_SIZE - chunk_offset) < CHUNK_SIZE ? (TOTAL_SIZE - chunk_offset) : CHUNK_SIZE;

            cudaMemcpyAsync(d_buffers[i], pinned_shellcode + chunk_offset, chunk_size, cudaMemcpyHostToDevice, streams[i]);

            int threadsPerBlock = 256;
            int blocksPerGrid = (chunk_size + threadsPerBlock - 1) / threadsPerBlock;

            decryptKernel <<<blocksPerGrid, threadsPerBlock, 0, streams[i]>>> (d_buffers[i], key, chunk_size);

            cudaMemcpyAsync(pinned_shellcode + chunk_offset, d_buffers[i], chunk_size, cudaMemcpyDeviceToHost, streams[i]);
        }
    }

    for (int i = 0; i < NUM_STREAMS; i++) {
        //cudaStreamSynchronize(streams[i]);
        cudaFree(d_buffers[i]);
        cudaStreamDestroy(streams[i]);
    }

    memcpy(shellcode, pinned_shellcode, TOTAL_SIZE);
    cudaFreeHost(pinned_shellcode); 
}

编辑补充：已开启Windows硬件加速，当前性能结果如下，但仍未达到异步并行的预期效果：

问题分析与解决建议 #

1. 流同步时机错误：当前代码在提交完所有流任务后，直接执行cudaFree和cudaStreamDestroy，但此时流内的异步任务可能尚未完成。必须在释放资源前等待所有流执行完毕，取消注释cudaStreamSynchronize(streams[i])，或调用cudaDeviceSynchronize()等待所有设备任务结束。

2. 任务粒度不合理：异或属于轻量计算，若CHUNK_SIZE过小，内存拷贝开销会远大于计算开销，掩盖并行效果；若过大，单流任务耗时过长，多流并行的重叠窗口缩小。建议将CHUNK_SIZE调整至几十MB级别（如64MB），平衡内存拷贝与计算的时间占比。

3. 任务提交模式限制并行重叠：当前外层循环每次批量提交所有流的H2D、Kernel、D2H任务后才进入下一轮，这种模式下多流任务是串行批量执行的，无法形成持续的流水线重叠。建议改为流水线式提交：当一个流完成当前批次的D2H操作后，立即提交下一批次的H2D任务，让不同流的内存拷贝与计算任务持续重叠。

4. Windows WDDM驱动的额外限制：即使开启硬件加速，Windows的WDDM驱动模型对CUDA异步操作存在一定约束，可尝试以下优化：

   • 使用cudaHostAllocWriteCombined分配锁页内存，提升主机到设备的拷贝性能；
   • 临时禁用Windows桌面 compositor（若场景允许），减少GPU资源抢占；
   • 确保CUDA版本与显卡驱动版本匹配，优先使用最新稳定版驱动。

5. 核函数逻辑与性能优化：当前核函数硬编码了异或值0x43，未使用传入的key参数，属于逻辑错误，需修正为buffer[idx] ^= key;；同时可使用向量类型（如unsigned int4）批量处理数据，提升计算吞吐量。

内容的提问来源于stack exchange，提问作者R3dy