Sycl内核-调用非常慢

Question

我对堆栈溢出、sycl和gpu编程都很陌生.我有一个具有工作的基本sycl内核的项目。逻辑起作用了，所以我跳过了这个问题。在编译和执行过程中也没有错误。

现在最大的问题是sycl代码的调用非常慢。首先，我认为它是一些内存复制或类似的东西，所以我遗漏了任何东西，除了下面可以看到的内容(最简单的注释是代码在非最小内核时所处的位置)。

我的测量时间：(发布x64)

• 使用Visual调试器显示，空内核调用的函数总时间：~100 ms
• 使用Cuda Nsight，OpenCl的时间-内核执行：~5 us

内核gpu时间为5 us非常快，与预期的一个空内核。但是，c++函数在我的代码中的总时间是100 ms慢。

这里有什么问题吗？还是预期系统开销会如此缓慢？(我对此表示怀疑)

我的努力：

• 我将compute++.exe标志从-O2更改为-O3，使总时间提高了5到10 ms。
• 我把内核减到最低

dll函数中的代码：

 { //scope

    sycl::gpu_selector gpuSel;
    sycl::queue myQueue(gpuSel);

    //....buffers

    auto ra = range<1>(size);

    myQueue.submit([&](sycl::handler& hd)
    {
            //....get_access<access::mode::read>

            auto kernel = ([=](cl::sycl::id<1> id)
            {
                    //...some vector math
            });

            hd.parallel_for<someName>(ra, kernel);             
    });

    myQueue.wait();
}

我正在使用：

• 2019
• ComputeCpp社区2.0.0
• 最新的库达司机
• NVIDIA Gtx 980 ptx64 (实验性ComputeCpp支持)

compute++呼叫：

"..\compute++.exe" -sycl -D_ALLOW_COMPILER_AND_STL_VERSION_MISMATCH -O3 -mllvm -inline-threshold=1000 -intelspirmetadata -sycl-target ptx64 -std=c++14 -I"../Codeplay/ComputeCpp/include" -I"../NVIDIA GPU Computing Toolkit/CUDA/v10.2/include"  -sycl-ih something.cpp.sycl -c something.cpp

摘要：

sycl内核的总执行时间很慢。我能在这里做些什么来改进它吗?还是因为在Nvidia gpus上实现了sycl/computecpp，并且预期会很慢？

Answer 1

首先，我要指出，这是一组非常简单的SYCL代码，所以如果您想要度量性能，这可能不是一个非常相关的例子。这里有一篇研究论文，展示了ComputeCpp与CUDA的相似性能，并给出了一种简化算法基准，图见幻灯片40。。您还将在演示文稿中看到，根据正在处理的数据集的大小，性能提高呈指数级增长。对于HPC编程来说，这通常是相同的，因为GPU的好处通常只在处理较大的数据集时才能看到。

您所看到的不同之处在于，ComputeCpp使用OpenCL回调，而NVIDIA OpenCL驱动程序在使用这些回调时似乎确实引入了开销。这是一段时间前关于这件事的相关员额

如果要编写一个使用回调的简单OpenCL内核，它将表现出相同的行为。

我还要补充的是，我们有直接使用CUDA的实现了对DPC++编译器的NVIDIA支持，并且没有看到相同级别的开销。你可以在我们的博客文章中找到更多关于这一点的信息，如果你想在NVIDIA硬件上运行SYCL代码，那是值得一试的。

Answer 2

当您想要添加或乘3或4个数字时，GPU是可怕的。为此，您最好使用CPU，并为此进行优化，并且您可能有一个AVX扩展，该扩展是为进行向量数学而优化的。因此，您应该将cl::sycl::gpu_selector替换为cl::sycl::cpu_selector。我不确定sycl是否使用AVX当你有一个，说它一定会使用多线程。

但是当你试图增加50万个数字时，GPU会比CPU快得多吗？

这段视频解释得很好。