PCクラスタワークショップ
HPC-OSS部会セッション
AMD ROCmソフトウェアについて
日本AMD株式会社大原久樹 (Hisaki.Ohara@amd.com)

[Public]

Optimized AI software stack
AI Models and Algorithms
AI Ecosystem optimized for AMD• Commitment to Open-Source+ Other Frameworks

Libraries
Compilers and Tools• No Code Change Execution
A proven software stack

Runtime• Optimized for Generative AI

AMD Instinct GPU

Leadership performance

Use of third party marks/logos/products is for informational purposes only and no endorsement of or by AMD is intended or implied GD-83
2|

AMD | PUBLIC | MARCH 2024

MIT/BSD License[Public]

Apache License

AMD ROCm

Software Stack

GPL License

HPC Focused

AI Solutions &Services

AI Focused

Business Services
rocALUTION

Fine Tunes Models

Hugging Face

Accelerated AI
Math &Communication
Libraries

MIOpenrocSOLVER

rocPRIM

MIGraphX

rocWMMArocSPARSE

rocThrustrocBLASLt

Composable
KernelEcosystem
rocBLAS

rocRAND

rocFFT

RCCL

Compiler

Profile/Tracer

Debugger

HIPIFY

TritonOpen Software
Platform

Compilers& Tools

Drivers
Runtimes

RedHat, SLES & Ubuntu Device Drivers and Run-timeGPU
Use of third party marks/logos/products is for informational purposes only and no endorsement of or by AMD is intended or implied GD-83
3|

AMD | PUBLIC | MARCH 2024

[Public]

What is Heterogeneous-Compute Interface (HIP)?
• C++ runtime API and kernel language with CUDA-like APIs
• HIP allows developers to create portable applications for AMD and NVIDIA GPUs

• cudaFunction style functions expressed as hipFunction

CUDA

4|HIP

▪ cudaMalloc (&d_x, N*sizeof(float));▪ hipMalloc (&d_x, N*sizeof(float));cudaMemcpy(d_x, x, N*sizeof(float),cudaMemcpyHostToDevice);

hipMemcpy(d_x, x, N*sizeof(float),hipMemcpyHostToDevice);

cudaDeviceSynchronize();

hipDeviceSynchronize();

AMD | PUBLIC | MARCH 2024

[Public]

HIP – Seamless Porting from CUDA to ROCm
Code Conversion Workflow Diagram
Simple CUDA to HIP Conversion Tool•
Hipify-perl: simple and fast•
Hipify-clang: high-quality translation

Kernels Comparison

CUDA Vector Add

HIP Vector Add

CUDA RT +NVCC

HIP Code

AMD HCC C++ Compiler

SW

NVIDIA NVCC Compiler

Kernels are syntactically the same

5|ISA

ISA

AMD GPUNVIDIA GPU

AMD | PUBLIC | MARCH 2024

[Public]

HIPIFY - Automatic CUDA to HIP
A set of tools that translate CUDA code into HIP C++• hipify-perl• Simple string replacement technique• Recommended for quick scans of projects• hipify-clang• Requires clang compiler to parse source files• High quality translation

Limitations• Inline NVIDIA PTX assembly• CUDA intrinsic functions• Hard-coded dependencies on warp size, sharedmemory size
• Code geared toward the limited size of register file on

NVIDIA hardware• Unsupported libraries, e.g., CUTLASS to CK6|

AMD | PUBLIC | MARCH 2024

[Public]Example: Qulacs (quantum simulator) Porting to HIP

https://github.com/qulacs/qulacs/pull/638
7|

AMD | PUBLIC | MARCH 2024

[Public]

8|

AMD | PUBLIC | MARCH 2024

[Public]

Contribution to LLVM
hisohara:~/Projects/ROCm/llvm-project$ git log --since="2023-0626" --pretty=format:'%ae’|
awk -F '@' '{print $2}'|sort|uniq -c|sort -nr|head -15
6261 gmail.com
4219 google.com
2958 users.noreply.github.com
2415 amd.com
1939 arm.com
1897 redhat.com
1181 apple.com
1146 sifive.com
921 redking.me.uk
894 outlook.com
863 maskray.me
744 intel.com
672 nvidia.com
517 fhahn.com
464 igalia.com

9|

AMD | PUBLIC | MARCH 2024

[Public]

10|

AMD | PUBLIC | MARCH 2024

[Public]

11|

AMD | PUBLIC | MARCH 2024

[Public]

Expanding Software Developer Ecosystem

Hugging Face

62,000+ CI/CD tests running nightly
Fully integrated optimum library

From ‘port-to’ to ‘develop-on’
with latest platforms

Use of third party marks/logos/products is for informational purposes only and no endorsement of or by AMD is intended or implied GD-83
12|

AMD | PUBLIC | MARCH 2024

Tensor Flow

Dynamo Instructor

JAX

OpenAI Triton

ONNX Runtime

OpenXLA

DeepSpeed

MLIR | IREE

Increasing open-source
contributions and expanding footprint[Public]

Transitioning AI Workloads to AMD GPUs
NvidiaAMD ROCmSW

Drop-in(Out-of-the-box)
SupportML Frameworks
PythonML Kernel
Development

TRITONKERNELS→

C++, Triton IR

CUDA
KERNELS→

ML Libraries
C++

CUDA Triton
Backend

cuBLAS,
cuSparse,
cuFFT,NCCL,
cuDNN…

Drop-in

Port / Optimize

NVCC

MIRROR
Equivalent Libraries

Use of third party marks/logos/products is for informational purposes only and no endorsement of or by AMD is intended or implied GD-83
13|AMD | PUBLIC | MARCH 2024

TRITONKERNELSCUDA
KERNELS→

ROCM TritonBackend→ HIPIFY → HIPCC

rocBLAS,rocSparse,
rocFFT,
RCCL,
MIOpen…

[Public]

AMD +: LLM Out-of-the-Box Acceleration with AMD GPU

1

Supports all transformers models and tasks

2

Zero code change even keep “cuda”

3Latest LLM optimization features

import torch# Support all transformers models and tasks on AMD
from transformers import AutoTokenizer, AutoModelForCausalLM

1model_id = "TheBloke/Llama-2-7B-fp16"
tokenizer = AutoTokenizer.from_pretrained(model_id)# Zero code change between Nvidia and AMD
device = torch.device("cuda:0" if torch.cuda.is_available() else "cpu")

2

3with torch.device(device):# Latest innovations and features are available (flash_attention_2)
model = AutoModelForCausalLM.from_pretrained(model_id, torch_dtype=torch.float16, attn_implementation="flash_attention_2")
token_ids = tokenizer([”…"], padding=False, return_tensors="pt").to(device)generated_ids = model.generate(**token_ids, max_new_tokens=50, early_stopping=True)
print(tokenizer.batch_decode(generated_ids))
https://huggingface.co/blog/huggingface-and-optimum-amd
14|

AMD | PUBLIC | MARCH 2024