SYCLの環境構築

本記事は以前はてなブログに投稿していた記事の内容を加筆修正して移行したものです。

SYCLとは？

CPU,GPU,FPGAのような様々な異るデバイスを扱うための統一規格です。

例えばGPUでは、Nvidia製のGPUで実行するコードを開発するにはCUDAを使うのが一般的ですが、そのコードはAMDやIntelのGPUでは実行できません。しかし、SYCLを用いて開発することで全てのデバイスで実行することができるようになります。

また、デバイスで実行されるコードを別に用意する必要のあるOpenCLとは違い、ホストで実行されるコードと同じC++のソースコードに記述することが可能です。

参考

環境構築

なぜソースからビルドするのか？

去年の年内まではArch Linuxで intel-one-api-dpcpp-cpp さえインストールしていればうまく動いていたのですが、年度末くらいに久しく触ろうとしたらOpenCLを認識できない等のトラブルが発生してビルドまではできるけど、実行ができないという状況になっていました。

色々とトラブルシューティングを試みたのでが、結局解決できなかったのでコンパイラ等のSDKをソースから自分でビルドする方法でやってみたところうまくできたのでblogに残しておこうと思います。

コンパイラ、SDKをソースからビルドする

intelのリポジトリのガイドを見るととりあえずGit,CMake,Python,Ninja,hwloc,C++のコンパイラが必要だそうなのでインストールされていなかったhwlocをインストールしました。

1

sudo pacman -S hwloc

まず適当なディレクトリを作成してそこにソースをクローンします。

1
2
3


mkdir sycl_workspace
cd sycl_workspace
git clone https://github.com/intel/llvm -b sycl # 結構時間がかかる!

続いてconfigureしてからビルドを実行します(お決まりのコース!)。ここで注意が必要なのは--native_cpuがないとCPUで実行できるコードをコンパイルできなくなります。

またLanguage Serverとしてclangdを使いたいので--ci-defaultも必要です。というのも普通の(普通にインストールできる)C++用のclangdではSYCLのコードの解析ができないためです。

1
2
3


python llvm/buildbot/configure.py --ci-default --native_cpu
python llvm/buildbot/compile.py -j 11 # 40分くらいはかかる
python llvm/buildbot/compile.py -t clangd -j 11 # 15分くらいはかかる

ビルドが終了するとllvm/build/install内にbin,lib,includeが作られ必要なファイル類が置かれるのでinstallの中身を適当な場所に置いて使うときにPATH,LD_LIBRARY_PATHを設定します。 clangdはllvm/build/bin内にのみ置かれますが、適当なディレクトリに置いたbinにcopyしておきます。

実際にコードをビルド&実行してみる

実行にはlevel-zero-loaderが必要で、更にintel CPU内蔵のGPUでも実行したければ intel-compute-runtimeも必要なのでインストールします。

1

sudo pacman -S intel-compute-runtime level-zero-loader

デバイスが認識されているかチェック

確認用にまずデバイスが認識されているかを見る以下のコードをビルド & 実行してみました。

check_devices.cpp

 1
 2
 3
 4
 5
 6
 7
 8
 9
10
11
12
13
14
15
16
17
18


#include <sycl/sycl.hpp>

auto main() -> int {
  std::vector<sycl::device> devices = sycl::device::get_devices();

  for (const auto &device : devices) {
    std::cout << "Device: " << device.get_info<sycl::info::device::name>()
              << "\n";
    std::cout << "Vendor: " << device.get_info<sycl::info::device::vendor>()
              << "\n";
    std::cout << "Max Compute Units: "
              << device.get_info<sycl::info::device::max_compute_units>()
              << "\n";
    std::cout << "----------------------------------\n";
  }

  return 0;
}

今回は$HOME/sycl_localにツールチェーンとライブラリ、ヘッダを置いたので、まず以下のようにしたPATH,LD_LIBRARY_PATHの設定を行いました。

1
2


export PATH="$HOME/sycl_local/bin:$PATH"
export LD_LIBRARY_PATH="$HOME/sycl_local/lib:${LD_LIBRARY_PATH:-}"

ビルドは以下のように行ないました。

1

clang -fsysl -fsysl-targets=native_cpu check_devices.cpp -o check_devices

これを実行してデバイス名が表示されればOKです。私の環境では以下のようになりました。 CPU,GPUが認識されていますね。

 1
 2
 3
 4
 5
 6
 7
 8
 9
10
11
12


Device: Intel(R) UHD Graphics 770
Vendor: Intel(R) Corporation
Max Compute Units: 32
----------------------------------
Device: SYCL Native CPU
Vendor: Intel(R) Corporation
Max Compute Units: 24
----------------------------------
Device: Intel(R) UHD Graphics 770
Vendor: Intel(R) Corporation
Max Compute Units: 32
----------------------------------

サンプルコード(ベクトル加算)のビルド&実行

最後にGPGPUの入門で良く見るベクトル加算c[i] = a[i] + b[i]の例の以下をCPU/GPUで実行して動作確認しました。

vector_add.cpp

 1
 2
 3
 4
 5
 6
 7
 8
 9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67


#include <iostream>
#include <sycl/buffer.hpp>
#include <sycl/sycl.hpp>
#include <vector>

template <class T> auto show_vec_1d(const std::vector<T> &array) -> void {
  for (const auto &v : array) {
    std::cout << v << " ";
  }
  std::cout << "\n";
}

auto check_devices(const sycl::device &dev) -> void {
  std::cout << "vendor: " << dev.get_info<sycl::info::device::vendor>() << "\n";
  std::cout << "device: " << dev.get_info<sycl::info::device::name>() << "\n";
  std::cout << "device version: " << dev.get_info<sycl::info::device::version>()
            << "\n";
  std::cout << "driver version: "
            << dev.get_info<sycl::info::device::driver_version>() << "\n";
  std::cout << "mx compute units: "
            << dev.get_info<sycl::info::device::max_compute_units>() << "\n";
  std::cout << "local mem size: "
            << dev.get_info<sycl::info::device::local_mem_size>() << "\n";
  std::cout << "has fp64 :" << dev.has(sycl::aspect::fp64) << "\n";
  std::cout << "has fp16 :" << dev.has(sycl::aspect::fp16) << "\n";
}

auto main(void) -> int {
  std::vector<int> vec_a{1, 2, 3, 4, 5}, vec_b{10, 20, 30, 40, 50};
  auto vec_size = (std::size_t)vec_a.size();

  // impl naive
  std::vector<int> vec_c_naive(vec_size, 0);
  for (auto j = 0; j < vec_size; ++j) {
    vec_c_naive[j] = vec_a[j] + vec_b[j];
  }
  show_vec_1d(vec_c_naive);

  // impl SYCL
  std::vector<int> vec_c_sycl(vec_size, 0);

  // Queue
  // auto q = sycl::queue{sycl::cpu_selector_v, sycl::async_handler{}};
  // use gpu !
  auto q = sycl::queue{sycl::gpu_selector_v, sycl::async_handler{}};
  auto dev = q.get_device();
  std::cout << "information of device now running: \n";
  check_devices(dev);

  {
    sycl::buffer<int, 1> device_a(vec_a.data(), sycl::range<1>(vec_size));
    sycl::buffer<int, 1> device_b(vec_b.data(), sycl::range<1>(vec_size));
    sycl::buffer<int, 1> device_c(vec_c_sycl.data(), sycl::range<1>(vec_size));
    q.submit([&](sycl::handler &cgh) {
      sycl::accessor acc_a(device_a, cgh, sycl::read_only);
      sycl::accessor acc_b(device_b, cgh, sycl::read_only);
      sycl::accessor acc_c(device_c, cgh, sycl::write_only);

      cgh.parallel_for(sycl::range<1>(vec_size), [=](sycl::item<1> id) {
        acc_c[id] = acc_a[id] + acc_b[id];
      });
    });
  }
  show_vec_1d(vec_c_sycl);

  return 0;
}

ここではCPU/GPUの切り替えは簡易的にsycl::queueの構築の際にsycl::cpu_selector_vにするコード、 sycl::gpu_selector_vにするコードとしてコメントアウトで切り変えてビルドしまいた。

また、ビルドはそれぞれ以下のように行いました。

1
2


clang -fsycl -fsycl-targets=native_cpu vector_add.cpp -o vector_add_cpu # CPU向けにビルド
clang -fsycl vector_add.cpp -o vector_add_gpu # GPU向けにビルド

それぞれを実行すると以下のような結果が得られました。それぞれ正しい結果が得られていますね。

 1
 2
 3
 4
 5
 6
 7
 8
 9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25


./vector_add_cpu # CPUで実行
11 22 33 44 55
information of device now running:
vendor: Intel(R) Corporation
device: SYCL Native CPU
device version: 0.1
driver version: 0.0.0
mx compute units: 24
local mem size: 32768
has fp64 :1
has fp16 :1
11 22 33 44 55

./vector_add_gpu # GPUで実行
11 22 33 44 55
information of device now running:
vendor: Intel(R) Corporation
device: Intel(R) UHD Graphics 770
device version: 12.2.0
driver version: 1.6.32961
mx compute units: 32
local mem size: 65536
has fp64 :0
has fp16 :1
11 22 33 44 55

感想

ソースからビルドする時に--native_cpuが必要だったりとハマったポイントはありましたが何とかSYCLの環境構築ができました。 syclacademyというチュートリアルを一応一周してはいるのでraytracingなんかをSYCLで書いてみたいと思います。