执行 CUDA 内核时黑屏C++输出正确的结果

black screen while executing C++ CUDA kernel, outputs correct result

本文关键字：输出 结果 C++ CUDA 内核 执行      更新时间：2023-10-16

我正在使用 CUDA C++对数组进行一些简单的并行计算。一切正常，内核输出正确的结果(使用串行 CPU 代码检查)，但是当内核正在执行时，我的整个屏幕在整个内核执行期间都会变黑。我是 CUDA 的新手，所以我可能做错了什么，我似乎无法弄清楚是什么。

#define KERNEL_FOR_ITERS 1e6
__global__ void addKernel(float *c, const float *a, const float *b)
{
int i = blockIdx.x * blockDim.x + threadIdx.x;
if (i % 2 == 0)
for (int j = 0; j < KERNEL_FOR_ITERS; j++)
c[i] += sqrt(abs(sin(a[i] * b[i])));
else
for (int j = 0; j < KERNEL_FOR_ITERS; j++)
c[i] += sqrt(abs(cos(a[i] * b[i])));
}

我的问题是，我可以防止我的屏幕在整个内核执行期间变黑，而不会影响太多的速度吗？

如果您描述您的设置，包括正在运行的操作系统和 GPU，GPU 是否驱动显示器，以及在 Windows 操作系统上，GPU 处于 WDDM 或 TCC 模式，这有点有用。

但是，我们可以在没有这些的情况下作一些一般性的发言。

正如评论中指出的那样，目前，运行 CUDA 内核的 GPU 如果也支持显示，则不会为显示请求提供服务。 这意味着在 GPU 内核运行时，显示器将显示为"冻结"或可能变黑。 当然，这种情况将来可能会改变，但这是当前和预期的行为。

在这种情况下，通常的建议是使用第二个GPU来运行CUDA，如果您根本不想干扰显示，如果您在Windows上，最好是该GPU能够并置于TCC模式。

为了在仅使用单个 GPU 时减轻影响，并且确实在单 GPU 显示环境中为生产目的提供 CUDA 支持，重要的是应用程序的 CUDA 端的设计方式是有限的内核持续时间。 为了获得良好的交互性，一个合理的起点是将内核持续时间限制为 0.1 秒或更短，因为交互性的损失水平可能不是特别明显。 如果您或某人不同意人为因素的说法，没关系;我们不需要争论它。 将内核持续时间减少到您决定的任何级别，这将带来良好的显示交互性。

在 Windows 情况下(据我所知，在 linux 情况下不是)，WDDM 命令批处理的情况进一步复杂化。 为了提高性能，可以对命令进行批处理，并且背靠背内核调用的批处理可能会导致感知到更长的交互性丢失时间，而不仅仅是单个内核调用所指示的。 我知道没有正式解决此问题的方法。 您可以通过在每次内核调用后发出虚假(即不需要)CUDA 操作(例如cudaStreamQuery())来"刷新"WDDM 命令队列。 同样，我不知道为此正式记录的方法，在某种程度上它可能取决于您的应用程序设计。

在性能方面，CUDA 内核启动通常涉及大约 100 微秒或更少的启动开销(我们可以称之为浪费时间)。 因此，如果我们将一个长时间运行的内核分解为 100 毫秒的"块"，并且每个块增加 ~100 微秒的开销，那么对性能的净影响可能是 CUDA 计算吞吐量降低 0.1%(假设显示任务微不足道)。

以您提供的代码为例，您可能希望将该内核分解为一系列内核，在您选择的 GPU 上对其进行基准测试/计时，以便内核运行不超过大约 100 毫秒(或您选择的数字)。

#define KERNEL_FOR_ITERS 1e6

__global__ void addKernel(float *c, const float *a, const float *b,const int iters)
{
int i = blockIdx.x * blockDim.x + threadIdx.x;
if (i % 2 == 0)
for (int j = 0; j < iters; j++)
c[i] += sqrt(abs(sin(a[i] * b[i])));
else
for (int j = 0; j < iters; j++)
c[i] += sqrt(abs(cos(a[i] * b[i])));
}
...
const int loop_iters = 1e4; // chosen by tuning or benchmarking
cudaStream_t str;
cudaStreamCreate(&str);
for (int i = 0; i < KERNEL_FOR_ITERS; i+= loop_iters){
addKernel<<<...,0,str>>>(d_c, d_a, d_b, loop_iters);
cudaStreamQuery(str);//probably unnecessary on linux}

我不认为这是您实际使用的内核，但顺便说一句，通过将线程之间实际不同的内容限制为一小部分代码，可以提高其性能特征。 例如：

__global__ void addKernel(float *c, const float *a, const float *b,const int iters)
{
int i = blockIdx.x * blockDim.x + threadIdx.x;
float val = a[i] * b[i];
if (i % 2 == 0)
val = sin(val);
else 
val = cos(val);
for (int j = 0; j < iters; j++)
c[i] += sqrt(abs(val));
}

无论如何，编译器可能会弄清楚这种收缩，但我通常会尝试给它最好的"领先优势"。

相关文章：