複数のベンダーの GPU プログラミング モデルを知ることが必要悪である理由は何ですか?

序章

並列プログラミングとアクセラレータ プログラミングに関する前回の2 回の投稿では、アクセラレータ プログラミングと並列プログラミングの基本と、コードの正確性を確保するために必要なプログラミングの概念について説明しました。アクセラレーターのプログラミングを学ぶプログラマーにとって興味深い質問は、「NVIDIA、AMD、および Intel アクセラレーター用にプログラムするために、CUDA、ROCm、および SYCL を学ぶ必要があるか?」ということです。

このシリーズを計画するとき、私は各アーキテクチャのハードウェア実行モデルとメモリ モデルの類似点と相違点を掘り下げることを期待していました。ハードウェア スレッドでの実行をグループ化するために CUDA、ROCm、および SYCL が使用する、ワープ、ウェーブフロント、およびワークグループについて説明するつもりでした。しかし、考えれば考えるほど、ブログの投稿には少し多すぎるように思えます。これらのトピックに興味がある場合は、Codeplay の CUDA 開発者向けの素敵な SYCL を参照してください。

代わりに、今日はコードなしから GPU での実行、任意の GPU での実行にできるだけ早く移行する方法に焦点を当てます。アクセラレーターのパフォーマンス上の利点をできるだけ早く確認することほど楽しいものはありません。私にとっては、Codeplay oneAPI for NVIDIAおよびoneAPI for AMD (beta)のリリースと重なるため、これは楽しいことです。

SYCL と複数のベンダーの GPU

SYCL は、NVIDIA、AMD、および Intel GPU のコンパイラ サポートを備えたプログラミング モデルです。コードを SYCL で記述し、それらのベンダー GPU でビルドして実行できます。以前は、GPU ターゲットに応じて、さまざまなコミュニティ コンパイラを使用するか、Intel オープン ソース コンパイラをビルドする必要がありました。最新の Codeplay ツールチェーンのリリースにより、ビルド済みの単一のツールチェーンを介してコードをすばやく簡単に実行できるようになりました。詳細については、Codeplay の CTO である Ruyman Reyes の新しいリリースに関するブログをご覧ください。

あなたが尋ねるかもしれない質問の 1 つは、実際にコードを一度ビルドしてから、実行時にどの GPU で実行するかを決定できるかということです。答えはイエスです！これは、このサポートの優れた点です。一度コードをビルドすると、実行時にターゲットを手動で選択するか、ランタイム ライブラリに自動的に選択させることができます。

ここまでたどり着いたので、これはあなたが興味を持っていると思います。私の目標は、選択した GPU で実行できるようにすることです。私の目的のために、Ubuntu 22.04 で 3 つのベンダー GPU すべてをテストし、APT 経由でインストールを行っています。

Codeplay ツールチェーンのインストール

まず、いくつかの基本的なシステム パッケージをセットアップします。

sudo apt update
sudo apt -y install cmake pkg-config build-essential

# download the key to system keyring
wget -O- https://apt.repos.intel.com/intel-gpg-keys/GPG-PUB-KEY-INTEL-SW-PRODUCTS.PUB \
| gpg --dearmor | sudo tee /usr/share/keyrings/oneapi-archive-keyring.gpg > /dev/null

# add signed entry to apt sources and configure the APT client to use Intel repository:
echo "deb [signed-by=/usr/share/keyrings/oneapi-archive-keyring.gpg] https://apt.repos.intel.com/oneapi all main" | sudo tee /etc/apt/sources.list.d/oneAPI.list

sudo apt-update
sudo apt install intel-basekit

ターゲットにする/ターゲットにする各 GPU に必要なベース ドライバー/フレームワークがあることを確認してください。ここでも、Ubuntu の手順をリンクしています。

Intel GPU ドライバーのインストール

異なる OS での Intel GPU の手順については、こちらを参照してください。Ubuntu 22.04 で Intel Arc を使用しているため、次の手順に従いました。

NVIDIA インストール用の oneAPI

ここにアクセスして、CUDA のインストール方法に関する最新の公式手順を入手してください。CUDA を起動して実行するために使用するものをインライン化しています。

1. CUDA キーリング、CUDA をインストールして再起動する

wget https://developer.download.nvidia.com/compute/cuda/repos/ubuntu2204/x86_64/cuda-keyring_1.0-1_all.deb
sudo dpkg -i cuda-keyring_1.0-1_all.deb

sudo apt-get update
sudo apt-get install cuda-11-7
sudo apt-get install nvidia-gds
sudo reboot

export PATH=/usr/local/cuda-11.7/bin${PATH:+:${PATH}}
export LD_LIBRARY_PATH=/usr/local/cuda-11.7/lib64\
                         ${LD_LIBRARY_PATH:+:${LD_LIBRARY_PATH}}

sh oneapi-for-nvidia-gpus-2023.0.0-linux.sh

. /opt/intel/oneapi/setvars.sh --include-intel-llvm
export PATH=/usr/local/cuda-11.7/bin${PATH:+:${PATH}}
export LD_LIBRARY_PATH=/usr/local/cuda-11.7/lib64\
                         ${LD_LIBRARY_PATH:+:${LD_LIBRARY_PATH}}

ここにアクセスして、ROCm のインストール方法に関する最新の公式手順を入手できます。また、ここにインライン化して、作業を少し簡単にします。

1. AMD amdgpu-install プログラムを APT 経由でインストールします

sudo apt-get update
wget https://repo.radeon.com/amdgpu-install/5.4.1/ubuntu/jammy/amdgpu-install_5.4.50401-1_all.deb
sudo apt-get install ./amdgpu-install_5.4.50401-1_all.deb

amdgpu-install --usecase="dkms,graphics,opencl,hip,hiplibsdk"

sh oneapi-for-amd-gpus-2023.0.0-linux.sh

. /opt/intel/oneapi/setvars.sh --include-intel-llvm

export PATH=/PATH_TO_ROCM_ROOT/bin:$PATH
export LD_LIBRARY_PATH=/PATH_TO_ROCM_ROOT/lib:$LD_LIBRARY_PATH

オープン ソースの oneAPI サンプル GitHub リポジトリで試すことができるさまざまなコード サンプルが多数あります。

Codeplay の例から、彼らがこの simple-sycl-app.cpp を作成したことがわかります。

#include <sycl/sycl.hpp>

int main() {
  // Creating buffer of 4 ints to be used inside the kernel code
  sycl::buffer<sycl::cl_int, 1> Buffer(4);

  // Creating SYCL queue
  sycl::queue Queue;

  // Size of index space for kernel
  sycl::range<1> NumOfWorkItems{Buffer.size()};

  // Submitting command group(work) to queue
  Queue.submit([&](sycl::handler &cgh) {
    // Getting write only access to the buffer on a device
    auto Accessor = Buffer.get_access<sycl::access::mode::write>(cgh);
    // Executing kernel
    cgh.parallel_for<class FillBuffer>(
        NumOfWorkItems, [=](sycl::id<1> WIid) {
          // Fill buffer with indexes
          Accessor[WIid] = (sycl::cl_int)WIid.get(0);
        });
  });

  // Getting read only access to the buffer on the host.
  // Implicit barrier waiting for queue to complete the work.
  const auto HostAccessor = Buffer.get_access<sycl::access::mode::read>();

  // Check the results
  bool MismatchFound = false;
  for (size_t I = 0; I < Buffer.size(); ++I) {
    if (HostAccessor[I] != I) {
      std::cout << "The result is incorrect for element: " << I
                << " , expected: " << I << " , got: " << HostAccessor[I]
                << std::endl;
      MismatchFound = true;
    }
  }

  if (!MismatchFound) {
    std::cout << "The results are correct!" << std::endl;
  }

  return MismatchFound;
}

通常、さまざまなターゲット用にビルドするには、次のようなコマンド ラインを使用できます。

NVIDIA:

clang++ -fsycl -fsycl-targets=nvptx64-nvidia-cuda simple-sycl-app.cpp -o simple-sycl-app

clang++ -fsycl -fsycl-targets=amdgcn-amd-amdhsa -Xsycl-target-backend --offload-arch=<ARCH> simple-sycl-app.cpp -o simple-sycl-app

clang++ -fsycl -fsycl-targets=spir64 simple-sycl-app.cpp -o simple-sycl-app

複数ベンダーの GPU 向けのビルド

NVIDIA GPU と Intel GPU で同時に実行したいとしましょう。実際には、NVIDIA または Intel GPU で実行されるファット バイナリを作成できます。ここでのコツは、複数の SYCL ターゲットを持つようにビルド コマンド ライン パラメーターを適切に設定することです。

clang++ -fsycl -fsycl-targets=nvptx64-nvidia-cuda,spir64 simple-sycl-app.cpp -o simple-sycl-app

tonym@LianLi-Linux:~/dev/2023.0Test$ clang++ -fsycl -fsycl-targets=nvptx64-nvidia-cuda,spir64 simple-sycl-app.cpp -o simple-sycl-app
 
tonym@LianLi-Linux:~/dev/2023.0Test$ SYCL_DEVICE_FILTER=cuda SYCL_PI_TRACE=1 ./simple-sycl-app
SYCL_PI_TRACE[basic]: Plugin found and successfully loaded: libpi_cuda.so [ PluginVersion: 11.15.1 ]
SYCL_PI_TRACE[all]: Selected device: -> final score = 1500
SYCL_PI_TRACE[all]:   platform: NVIDIA CUDA BACKEND
SYCL_PI_TRACE[all]:   device: NVIDIA GeForce RTX 3080 Ti
The results are correct!
tonym@LianLi-Linux:~/dev/2023.0Test$
 
tonym@LianLi-Linux:~/dev/2023.0Test$ SYCL_DEVICE_FILTER=ext_oneapi_level_zero:gpu:0 SYCL_PI_TRACE=1 ./simple-sycl-app
SYCL_PI_TRACE[basic]: Plugin found and successfully loaded: libpi_level_zero.so [ PluginVersion: 11.15.1 ]
SYCL_PI_TRACE[all]: Selected device: -> final score = 1550
SYCL_PI_TRACE[all]:   platform: Intel(R) Level-Zero
SYCL_PI_TRACE[all]:   device: Intel(R) Graphics [0x56a0]
The results are correct!

今後の課題

マルチベンダー GPU の例に AMD の例が含まれていないことに気がつく読者もいるかもしれません。これは、最近 AMD 6800XT GPU を真新しい AMD RX 7900XT GPU に置き換えたためです。残念ながら、現在、RDNA3 カードで使用しているカーネルに関係なく、ROCm は Linux システムに正しくインストールされません。これが更新されたら、私は戻ってきて、3 つのカードすべてで実行される例を使用して説明を提供します。

以前はこのフローのバージョンが 6800XT で機能していましたが、それはここに記載されているオープン ソースの DPC++ フローを使用して Codeplay ソフトウェアがリリースされる前であったことに注意してください。

結論

Codeplay の NVIDIA GPU 用の oneAPI を使用すると、NVIDIA または Intel GPU 用のバイナリを簡単に作成できました。NVIDIA GPU 用の追加の oneAPI をセットアップする時間は、Intel oneAPI Base Toolkit をインストールする時間に加えて約 10 分でした。これには、CUDA と Codeplay ソフトウェアのインストールが含まれます。

新しい Codeplay ソフトウェアは、マルチベンダー GPU オプションを検討している個人や企業にまったく新しいオプションを提供します。オプションがオープン ソースになる前は、コミュニティがサポートするコンパイラでした。これは素晴らしいオプションであり、私が日常の仕事で通常使用しているものです。

ただし、長期的なサポートを検討している人々、つまり企業は生計をツールチェーンに依存するようになり、Codeplay と Intel から公式にサポートされているオプションを利用することで、安心感を得ることができます.

私が読んでいるランダムな技術ニュースを見たい場合は、Twitter でフォローしてください。また、私がホストしている開発者向けの Intel ポッドキャストであるCode Togetherもチェックしてください。

Tony は、Intel のソフトウェア アーキテクト兼テクニカル エバンジェリストです。彼はインテル VTune プロファイラーやその他のパフォーマンス ツールのアーキテクトであり、最近では、Habana のスケーラブルな MLPerf ソリューションを可能にするデータ センター プラットフォームを構築するソフトウェア エンジニアリング チームを率いていました。