CUDA C++ で Flash Attention を実装する冒険

この記事では、著者がCUDA C++でFlash Attentionを実装し、最適化していく過程を詳細に説明しています。基本的な実装から始めて、共有メモリのスウィズリング、2段階パイプライン、より効率的なldmatrixの使用などの手法を用いてカーネルを段階的に改良しました。反復的なプロファイリングと最適化により、最終的な実装はハードウェアの理論限界に近いパフォーマンスを実現しています。また、オンラインsoftmaxの実装の詳細や、共有メモリのバンク競合の解決策についても深く掘り下げており、CUDA C++開発者にとって貴重な知見を提供しています。