OpenAI Triton: Menulis Kernel GPU Tanpa C++/CUDA Murni
I1 — GPU Kernels: Dari CUDA ke Triton & FlashAttention
Menulis Kernel GPU Berkinerja Tinggi Menggunakan Python
Sebelum adanya Triton, menulis kernel GPU kustom membutuhkan keahlian mendalam dalam C++, CUDA, manajemen sinkronisasi thread (__syncthreads()), dan pencegahan konflik memori (shared memory bank conflict).
Paradigma Block-Level Triton
Triton mengabstraksikan thread hardware menjadi operasi berbasis Blok Tensor:
import triton
import triton.language as tl
@triton.jit
def add_kernel(x_ptr, y_ptr, output_ptr, n_elements, BLOCK_SIZE: tl.constexpr):
pid = tl.program_id(axis=0)
block_start = pid * BLOCK_SIZE
offsets = block_start + tl.arange(0, BLOCK_SIZE)
mask = offsets < n_elements
# Load data dari VRAM ke register secara paralel
x = tl.load(x_ptr + offsets, mask=mask)
y = tl.load(y_ptr + offsets, mask=mask)
# Operasi komputasi
output = x + y
# Tulis hasil kembali ke VRAM
tl.store(output_ptr + offsets, output, mask=mask)
Compiler Triton secara otomatis menangani alokasi register optimal, memory coalescing, dan pemanfaatan instruksi Tensor Cores!
I1