SCALE - AMD GPUで修正なしにCUDAを実行

 (docs.scale-lang.com)
12 ポイント 投稿者 GN⁺ 2024-07-16 | 1件のコメント | WhatsAppで共有
  • SCALEは、CUDAアプリケーションをAMD GPU向けにネイティブコンパイルできるようにするGPGPUプログラミングツールキット
  • CUDAプログラムやビルドシステムを修正する必要はなく、さらに多くのGPUベンダーとCUDA APIのサポートを開発中

どのように動作するのか?

  • SCALEは、他のクロスプラットフォームGPGPUソリューションと比べていくつかの主要な革新を持つ
    • CUDAプログラムをそのまま受け入れる。他の言語へ移植する必要はない。これは、プログラムがインラインPTX asm を使用する場合でも同様
    • SCALEコンパイラは、nvcc と同じコマンドラインオプションとCUDA方言を受け入れ、ドロップイン置き換えとして動作する
    • NVIDIA CUDA Toolkitのインストールを「装う」ことで、既存のビルドツールやスクリプトがそのまま動作する

どのプロジェクトがテストされたのか?

  • オープンソースのCUDAプロジェクトをコンパイルし、テストを実行してSCALEを検証
  • 現在、次のオープンソースプロジェクトがナイトリー自動テストに含まれており、完全に合格している
    • NVIDIA Thrust, Blender Cycles, AMGX, llama-cpp, faiss, xgboost, GOMC, stdgpu, hashcat

どのGPUがサポートされるのか?

  • 次のGPUターゲットがサポートされ、ナイトリーテストに含まれている
    • AMD gfx1030 (Navi 21, RDNA 2.0)
    • AMD gfx1100 (Navi 31, RDNA 3.0)
  • 次のGPUターゲットは一時的な手動テストを経ており、「動作しているようだ」
    • AMD gfx1010
    • AMD gfx1101
  • 次のGPUサポートに向けて作業中
    • AMD gfx900 (Vega 10, GCN 5.0)
  • 特定のAMD GPUアーキテクチャのサポートを早急に望む場合は連絡してほしい

SCALEの構成要素

  • AMD GPU向けにnvcc方言のCUDAをコンパイルできる nvcc 互換コンパイラ。PTX asmを含む
  • AMD GPU向けのCUDAランタイムおよびドライバAPI実装
  • ROCmライブラリに委譲して「CUDA-X」APIを提供するオープンソースのラッパーライブラリ。cuBLAS や cuSOLVER のようなライブラリがこの方法で処理される

SCALEと他のソリューションの違い

  • 新しいGPGPUソフトウェアの書き方を提供するのではなく、SCALEは広く使われているCUDA言語で書かれたプログラムをAMD GPU向けに直接コンパイルできるようにする
  • SCALEはNVIDIA CUDAとの完全な互換性を目標としている。ユーザーは複数のコードベースを維持したり性能を妥協したりせずに、複数のGPUベンダーをサポートできるべきだと考えている
  • SCALEの言語はNVIDIA CUDAの スーパーセット であり、nvcc から離れたいユーザーに対して、GPUコードの記述をより簡単かつ効率的にする任意の言語拡張を提供する
  • SCALEは進行中の作業である。利用を妨げるAPIの欠落がある場合は連絡してほしい。開発優先順位を調整するとのこと

GN⁺のまとめ

  • SCALEは、CUDAアプリケーションをAMD GPU向けにネイティブコンパイルできるようにする重要なツールキット
  • 既存のCUDAプログラムを修正することなくAMD GPUで実行できるため、開発者にとって大きな利点がある
  • NVIDIA CUDAとの完全な互換性を目指しており、複数のGPUベンダーをサポートするうえで有利
  • 進行中のプロジェクトであり、必要なAPIが欠落している場合は開発チームに連絡して優先順位を調整できる
  • 類似した機能を持つプロジェクトとして、ROCmとHIPがある

関連記事

1件のコメント

 
GN⁺ 2024-07-16
Hacker Newsの意見

  • 多くの人はAMDが翻訳レイヤーをサポートすべきだと考えているが、それは悪いアイデアだという意見がある

    • CUDAはベンダー中立に設計されておらず、Nvidiaは技術的・法的に困難を作り出す可能性がある
    • 例えば、cuDNNやcuBLASをこの上で動かすことはライセンス契約に違反する可能性がある
    • こうしたNvidiaライブラリは、AMDが再実装してサポートしなければならないAPI境界の一部になるだろう

  • バグ互換性を追求するのは愚かだという意見がある

    • 重要なCUDAユーザーはオープンソースである
    • AMDはpytorchやllama.cppのようなアップストリームプロジェクトに直接サポートを実装できる
    • サポートがあればコミュニティが保守できる

  • ハードウェアに大きく依存するコードがAMDで「そのまま動く」ことがあり得るのか理解できないという意見がある

    • ほとんどの本格的なCUDAコードは、レジスタファイルや共有メモリサイズ、wgmma命令、最適なテンソルコアのメモリおよびレジスタレイアウト、テンソルメモリアクセラレータ命令などを意識している

  • 事実なら印象的だが、オープンソースではなく、どのように動作するのか正確な詳細が不足しているという意見がある

    • 最近のプロジェクトにオープンソース、少なくともソース利用可能であることを期待する理由はよく分からない

  • Nvidiaの高い評価の主因は、AMDがGPUをMLに有用にすることへ投資していないためだという意見がある

    • AMDが独占禁止措置を恐れているのか、あるいはハードウェアのアプローチに競争力を制限する何かがあるのかもしれない
    • 同社は暗号資産マイニング向けGPU需要の急増時と現在のAIブームによる需要急増時に、数十億ドル規模の機会を逃したように見える

  • AMDの失策があまりに大きいため、このようなプロジェクトを称賛したくなるという意見がある

    • 特にLinuxでは、ノートPCの機能が物理的には存在するのに使えず、非常にフラストレーションがたまる

  • 数年前にSpectral Computeで働いていたことがある

    • 非常に賢く有能な技術チームだった
    • 当時はAMD向けを対象にしていただけでなく、基本的なLLVM ptxバックエンドやNVCCを上回っていた

  • CUDAを少し書いたことがある

    • AMDカード向けのコードを書くための基本セットアップは何なのか気になる

  • このプロジェクトは素晴らしいという意見がある

    • AMDがNvidiaと直接競争するようになることを期待している

  • 現在の制限事項についてのページがあるのは良いが、大半の人が「CUDA」として説明しているものは実際のCUDA機能のごく一部だという意見がある

    • ワープシャッフル、アトミック演算、DPX、TMA、MMAなどの高度な機能について比較表があるとよい
    • PTX命令をRDNA対応命令、またはそれをエミュレートする命令一覧にマッピングする表が理想的だ

  • 技術的には可能なので本物かもしれないという意見がある

    • インラインPTXを解析してAMDGPUにマッピングするのは非常に大変だろう
    • インラインPTXを使わないCUDAソースをAMDGPU向けに動かすことは、HIPに置き換えるのと似ている
    • 一部の詳細は疑わしいかもしれない。例えば、アトミックモデルが一致しない、あるいはVoltaが異なる命令ポインタモデルを持っている可能性がある
    • しかし、正しく行うことはできる
    • AMDはこれをやらないだろう
    • CUDAは一般にそれほど優れたものではなく、法務チームが問題を起こすだろう
    • しかし、他の人たちなら十分に実現できる