oneAPI GPU 优化指南 - 在 SYCL kernel 中使用标准库函数

翻译已于 2024-01-05 16:37:03 修改 · 356 阅读

文章标签：

#oneapi

于 2024-01-05 16:24:36 首次发布

oneAPI GPU 优化指南专栏收录该内容

46 篇文章

订阅专栏

本文探讨了在SYCLkernel中使用标准C++函数，如rand()，在GPU设备上的限制，并介绍了InteloneAPI的DPC++和oneMKL库作为替代，以在GPU上生成随机数。

本章节翻译by chenchensmail@163.com 原文：Using Standard Library Functions in SYCL Kernels (intel.com)

在 SYCL kernel 中使用标准库函数#

一些标准的 C++ 函数可以在 SYCL kernel 中调用，但并非所有函数都可以。请参阅 Data Parallel C++ 的第18章（库）以了解支持的函数概述。这里提供了一个简单的示例，以说明从 SYCL kernel 中调用不受支持的函数时会发生什么。以下程序使用 rand() 函数生成一系列随机数：

 #include <CL/sycl.hpp>
 #include <iostream>
 #include <random>
 
 constexpr int N = 5;
 
 extern SYCL_EXTERNAL int rand(void);
 
 int main(void) {
 #if defined CPU
   sycl::queue Q(sycl::cpu_selector_v);
 #elif defined GPU
   sycl::queue Q(sycl::gpu_selector_v);
 #else
   sycl::queue Q(sycl::default_selector_v);
 #endif
 
   std::cout << "Running on: "
             << Q.get_device().get_info<sycl::info::device::name>() << std::endl;
 
   // Attempt to use rand() inside a DPC++ kernel
   auto test1 = sycl::malloc_shared<float>(N, Q.get_device(), Q.get_context());
 
   srand((unsigned)time(NULL));
   Q.parallel_for(N, [=](auto idx) {
      test1[idx] = (float)rand() / (float)RAND_MAX;
    }).wait();
 
   // Show the random number sequence
   for (int i = 0; i < N; i++)
     std::cout << test1[i] << std::endl;
 
   // Cleanup
   sycl::free(test1, Q.get_context());
 }

该程序可以编译以在 CPU （即 cpu_selector ）或 GPU （即 gpu_selector ）设备上执行 SYCL kernel。它在这两个设备上编译时没有错误，并且在 CPU上正确运行，但在 GPU 上运行时失败:

$ icpx -fsycl -DCPU -std=c++17 external_rand.cpp -o external_rand
$ ./external_rand
Running on: Intel(R) Xeon(R) E-2176G CPU @ 3.70GHz
0.141417
0.821271
0.898045
0.218854
0.304283

$ icpx -fsycl -DGPU -std=c++17 external_rand.cpp -o external_rand
$ ./external_rand
Running on: Intel(R) Graphics Gen9 [0x3e96]
terminate called after throwing an instance of 'cl::sycl::compile_program_error'
  what():  The program was built for 1 devices
Build program log for 'Intel(R) Graphics Gen9 [0x3e96]':

error: undefined reference to `rand()'

error: backend compiler failed build.
 -11 (CL_BUILD_PROGRAM_FAILURE)
Aborted

在 Just-In-Time (JIT) 编译期间发生故障，因为对 rand() 的引用未定义。即使此函数声明为 SYCL_EXTERNAL, 但在 GPU 设备上没有 SYCL 等效函数。

幸运的是， SYCL 库包含许多标准 C++ 函数的替代方法，包括生成随机数的函数。以下示例显示了使用|intel_r| oneAPI DPC ++库 (oneDPL) 和 Intel® oneAPI Math Kernel Library (oneMKL) 实现等效功能：

 #include <CL/sycl.hpp>
 #include <iostream>
 #include <oneapi/dpl/random>
 #include <oneapi/mkl/rng.hpp>
 
 int main(int argc, char **argv) {
   unsigned int N = (argc == 1) ? 20 : std::stoi(argv[1]);
   if (N < 20)
     N = 20;
 
   // Generate sequences of random numbers between [0.0, 1.0] using oneDPL and
   // oneMKL
   sycl::queue Q(sycl::gpu_selector_v);
   std::cout << "Running on: "
             << Q.get_device().get_info<sycl::info::device::name>() << std::endl;
 
   auto test1 = sycl::malloc_shared<float>(N, Q.get_device(), Q.get_context());
   auto test2 = sycl::malloc_shared<float>(N, Q.get_device(), Q.get_context());
 
   std::uint32_t seed = (unsigned)time(NULL); // Get RNG seed value
 
   // oneDPL random number generator on GPU device
   clock_t start_time = clock(); // Start timer
 
   Q.parallel_for(N, [=](auto idx) {
      oneapi::dpl::minstd_rand rng_engine(seed, idx); // Initialize RNG engine
      oneapi::dpl::uniform_real_distribution<float>
          rng_distribution;                      // Set RNG distribution
      test1[idx] = rng_distribution(rng_engine); // Generate RNG sequence
    }).wait();
 
   clock_t end_time = clock(); // Stop timer
   std::cout << "oneDPL took " << float(end_time - start_time) / CLOCKS_PER_SEC
             << " seconds to generate " << N
             << " uniformly distributed random numbers." << std::endl;
 
   // oneMKL random number generator on GPU device
   start_time = clock(); // Start timer
 
   oneapi::mkl::rng::mcg31m1 engine(
       Q, seed); // Initialize RNG engine, set RNG distribution
   oneapi::mkl::rng::uniform<float, oneapi::mkl::rng::uniform_method::standard>
       rng_distribution(0.0, 1.0);
   oneapi::mkl::rng::generate(rng_distribution, engine, N, test2)
       .wait(); // Generate RNG sequence
 
   end_time = clock(); // Stop timer
   std::cout << "oneMKL took " << float(end_time - start_time) / CLOCKS_PER_SEC
             << " seconds to generate " << N
             << " uniformly distributed random numbers." << std::endl;
 
   // Show first ten random numbers from each method
   std::cout << std::endl
             << "oneDPL"
             << "\t"
             << "oneMKL" << std::endl;
   for (int i = 0; i < 10; i++)
     std::cout << test1[i] << " " << test2[i] << std::endl;
 
   // Show last ten random numbers from each method
   std::cout << "..." << std::endl;
   for (size_t i = N - 10; i < N; i++)
     std::cout << test1[i] << " " << test2[i] << std::endl;
 
   // Cleanup
   sycl::free(test1, Q.get_context());
   sycl::free(test2, Q.get_context());
 }

必要的 oneDPL 和 oneMKL 函数分别包含在 <oneapi/dpl/random> 和 <oneapi/mkl/rng.hpp> 中。 oneDPL 和 oneMKL 示例执行相同的操作序列：从时钟获取随机数种子，初始化随机数引擎，选择所需的随机数分布，然后生成随机数。 oneDPL 代码使用 SYCL kernel 显式执行设备部署。在 oneMKL 代码中， mkl::rng 函数隐式处理设备部署。