cpu 代码延迟了 cuda 内核执行

Question

我无法理解以下公认的非常简单的代码，这是一个更复杂的项目的简化版本，我现在花了很多时间在这个项目上。这段代码将在大约 2000 毫秒内在我的系统上运行。但是，当我启用该线路使 cpu 进入睡眠状态 500 毫秒时，该程序总共将运行更长的时间，大约为 2500 毫秒。

我无法理解这如何符合cuda 内核相对于主机异步执行的声明？

在 Visual Studio 2019 上运行 cuda 11.1

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <chrono>
#include <iostream>
#include <numeric>
#include <thread>

__global__ void kernel(double* val, int siz) {
    for (int i = 0; i < siz; i++) val[i] = sqrt(val[i]); //calculate square root for every value in array
}

int main() {
    auto t1 = std::chrono::high_resolution_clock::now();

    const int siz = 1'000'000; //array length
    double* val = new double[siz];
    std::iota(val, val + siz, 0.0); //fill array with 0, 1, 2,...
    double* d_val;

    cudaMalloc(&d_val, sizeof(double) * siz);
    cudaMemcpy(d_val, val, sizeof(double) * siz, cudaMemcpyDefault);
    kernel <<<1, 1 >>> (d_val, siz); //start kernel
    //std::this_thread::sleep_for(std::chrono::milliseconds(500)); //---- putting cpu to sleep also delays kernel execution?
    cudaError_t err = cudaDeviceSynchronize();
    auto t2 = std::chrono::high_resolution_clock::now();

    std::cout << "status: " << cudaGetErrorString(err) << std::endl;
    std::chrono::duration<double, std::milli> ms = t2 - t1;
    std::cout << "duration: " << ms.count() << std::endl;

    delete[] val;
}

Answer 1

我无法理解这如何符合 cuda 内核相对于主机异步执行的声明？

您正在体验此处所述的 WDDM 命令批处理。

简而言之，在 Windows 上，当处于 WDDM 驱动程序模型时，GPU 命令（例如来自 cuda 运行时 API 的任何内容，加上内核启动）将被发送到命令队列。每隔一段时间，根据未发布的启发式方法，并且没有提供明确的用户控制，命令队列将被“刷新”，即发送到 GPU，此时（如果当前不忙）GPU 将开始处理这些命令。

因此，在 WDDM 设置中，内核向命令队列的分派是非阻塞的（控制立即返回给 CPU 线程）。从命令队列到 GPU 的工作分派遵循一些其他启发式方法。（无论如何，内核执行与主机线程异步）

如果这是一个问题，您至少有几个选择：

1. 在 Windows 上，切换到 TCC 驱动程序模型中的 GPU。
2. 在 Windows 上，尝试使用链接答案中描述的“黑客”之一。
3. 切换到Linux