为什么 cuda 内核函数成本 cpu？

Question

我发现在某些特定情况下，cuda 内核函数可以 花费 cpu。

我正在使用 win7(32bit) + vs2013(sp4) + cuda 6.5 + GTX 650

我的代码如下所示：

cudaMalloc(args...); // malloc buffer outside the busy loop
while(1) {
    Sleep(1);
    kernel<<<B, T>>>(args...);
}

// and the kernel function(uses 30 registers) 
__global__ void kernel(args...) {
    char l_plain[256] = {0}; // use local variable
    // copy memory from argument buffer to l_plain
    // to avoid using global memory in the loop below

    // it costs CPU when the ROUND is very huge, like 1024
    for(int i=0; i<ROUND; i++) {
        uint u1, u2, u3, u4;
        md5_vfy(l_plain, 16, &u1, &u2, &u3, &u4); // a __device__ function that calculates md5 hash
        // prepare for next plain text
    }
}

我可以验证 <<<B, T>>> 优化得很好，实际上我使用 <<<32,128>>>,在Visual Profiler中占用率几乎是100%。

函数 md5_vfy 可以在这里找到：http://pastebin.com/KU3zUxpb

在我的机器上，当循环 ROUND 小于 720 时，cpu 总是免费的，它需要花费任务管理器中为 0%。并且将循环 ROUND 更改为 750/800/900/1000... cpu 成本也线性增长。

我想知道是什么造成了差异，当 ROUND 非常大时，cpu 的成本是多少？我可以提供 Visual Profiler 截图。

更新原因

我想知道这个，因为我想减少所有 cpu 成本。我的程序应该支持两种模式：正常模式和游戏模式。正常模式占用 100% cpu 和 gpu，可能每秒运行 800m 哈希。游戏模式占用 0% cpu 和 5-xx% gpu，每秒运行 50m 哈希。

谢谢。

Answer 1

据推测，当您增加 ROUND 时，内核执行时间会变长。

考虑您的主循环：

while(1) {
    Sleep(1);
    kernel<<<B, T>>>(args...);
}

看来 windows Sleep() 函数 suspends thread execution for the specified period in milliseconds。因此，上述循环中 kernel 的执行时间是否多于或少于 1 毫秒将是影响程序动态行为的一个非常重要的因素。

假设执行时间小于1毫秒。在这种情况下，每次来自 Sleep 的 CPU 线程 returns 时，先前的内核调用都已完成，并且新的内核调用或多或少会立即开始执行。由于内核调用是异步的，因此控制权返回给 CPU 线程，然后该线程返回休眠状态，平均而言，该线程的总体 CPU 使用率将非常低。

现在考虑内核执行时间超过1ms的情况（可能是因为ROUND比较大）。在这种情况下，您的主循环发出内核调用，进入休眠状态，1 毫秒后唤醒，但之前发出的内核仍在执行。没问题，CPU 线程仍然可以发出另一个内核调用，该调用将进入等待队列。控制返回给 CPU 线程，然后 CPU 线程返回休眠状态。然而，随着时间的推移（并且非常快，对于 1 毫秒的休眠期）这个 "excess" 将累加起来，最终队列中将有两个等待启动，然后是 3 个等待启动，等等。最终等待启动队列将满。

在这一点，行为发生了显着变化。具有完整启动等待队列的新内核启动 不再异步 ，并且 CPU 线程将在此时阻塞，等待队列槽打开。此时，CPU 利用率将达到 100%，同时线程正忙于等待队列槽打开。一旦队列槽打开，它将发出下一个内核调用（它将在新打开的队列槽中发生），然后进入休眠状态。

在这一点，随着您进一步增加内核执行的长度，CPU 忙等待（100% 利用率）与休眠的时间（0% 利用率）将随着内核执行时间的增加而按比例变化。

我不完全确定你的目标是什么，但为了避免这种情况，你可以尝试增加 Sleep() 周期以大致匹配（即略大于）你的内核执行时间，随着你增加ROUND 值。

为什么 cuda 内核函数成本 cpu？

why cuda kernel function costs cpu?

cuda