GPU මිණුම් පරීක්ෂණ, 70B මතක ප්‍රමාණනය සහ AI සම්පාදක

නවීන AI පද්ධති ඉංජිනේරුකරණය කර්නල් මට්ටමේ මිණුම් සලකුණු පරීක්ෂණ, ධාරිතා සහ කලාප පළල ගණිතය සහ බහු-අදියර සම්පාදක පහළ-මට්ටම්කරණය එකට ගෙන එයි. පහත දැක්වෙන්නේ ක්ෂේත්‍ර තුනේම සම්පූර්ණ විස්තරයකි.

1. සම්පූර්ණ මිණුම් පරීක්ෂණ ස්ක්‍රිප්ට් එක: Eager PyTorch එදිරිව torch.compile එදිරිව අභිරුචි Triton

GPU ක්‍රියාත්මක කිරීම අසමමුහුර්ත බැවින්, නිවැරදි CUDA මිණුම් පරීක්ෂණ සඳහා සැබෑ කාල ඔරලෝසු කියවීමට පෙර උණුසුම් කිරීමේ වාර (JIT සම්පාදනය ඇරඹීමට සහ මතක හැඹිලි වෙන් කිරීමට) සහ torch.cuda.synchronize() (හෝ CUDA Events) අවශ්‍ය වේ.

පහත දැක්වෙන්නේ 4096 × 8192 FP16 න්‍යාසයක් පුරා ඒකාබද්ධ පේළි-අනුව Softmax වැඩබරකදී Eager PyTorch, torch.compile (TorchInductor) සහ අභිරුචි Triton Kernel සසඳන ස්වාධීන Python ස්ක්‍රිප්ට් එකකි.

import torch
import triton
import triton.language as tl

# 1. Custom Fused Triton Kernel
@triton.jit
def fused_softmax_kernel(
    output_ptr, input_ptr,
    input_row_stride, output_row_stride,
    n_cols,
    BLOCK_SIZE: tl.constexpr
):
    row_idx = tl.program_id(0)
    row_start_ptr = input_ptr + row_idx * input_row_stride
    col_offsets = tl.arange(0, BLOCK_SIZE)
    mask = col_offsets < n_cols

    # Single read from HBM into SRAM/registers
    row = tl.load(row_start_ptr + col_offsets, mask=mask, other=-float('inf')).to(tl.float32)
    row_minus_max = row - tl.max(row, axis=0)
    numerator = tl.exp(row_minus_max)
    denominator = tl.sum(numerator, axis=0)
    softmax_output = (numerator / denominator).to(tl.float16)

    # Single write back to HBM
    output_row_start_ptr = output_ptr + row_idx * output_row_stride
    tl.store(output_row_start_ptr + col_offsets, softmax_output, mask=mask)

def triton_softmax(x: torch.Tensor) -> torch.Tensor:
    n_rows, n_cols = x.shape
    BLOCK_SIZE = triton.next_power_of_2(n_cols)
    y = torch.empty_like(x)
    fused_softmax_kernel[(n_rows,)](
        y, x, x.stride(0), y.stride(0), n_cols, BLOCK_SIZE=BLOCK_SIZE
    )
    return y

# 2. Unfused Eager PyTorch Baseline (Forces multiple HBM round-trips)
def naive_pytorch_softmax(x: torch.Tensor) -> torch.Tensor:
    x_fp32 = x.float()
    x_max = x_fp32.max(dim=-1, keepdim=True)[0]
    exp_x = torch.exp(x_fp32 - x_max)
    return (exp_x / exp_x.sum(dim=-1, keepdim=True)).half()

# 3. TorchInductor Compiled Function
compiled_softmax = torch.compile(naive_pytorch_softmax, mode="max-autotune")

# 4. Accurate CUDA Event Benchmarking Harness
def benchmark_fn(fn, x, warmup=25, rep=100):
    for _ in range(warmup):
        fn(x)
    torch.cuda.synchronize()

    start_events = [torch.cuda.Event(enable_timing=True) for _ in range(rep)]
    end_events = [torch.cuda.Event(enable_timing=True) for _ in range(rep)]

    for i in range(rep):
        start_events[i].record()
        fn(x)
        end_events[i].record()

    torch.cuda.synchronize()
    times_ms = [s.elapsed_time(e) for s, e in zip(start_events, end_events)]
    median_ms = sorted(times_ms)[rep // 2]

    # Effective Memory Bandwidth (GB/s): 1 read + 1 write of FP16 tensor
    bytes_transferred = 2 * x.numel() * x.element_size()
    gb_per_sec = (bytes_transferred / (median_ms * 1e-3)) / 1e9
    return median_ms, gb_per_sec

if __name__ == "__main__":
    assert torch.cuda.is_available(), "CUDA GPU required."
    x = torch.randn((4096, 8192), device="cuda", dtype=torch.float16)

    # Verify numerical accuracy against native PyTorch
    ref = torch.softmax(x, dim=-1)
    assert torch.allclose(triton_softmax(x), ref, atol=1e-2)
    assert torch.allclose(compiled_softmax(x), ref, atol=1e-2)

    for name, fn in [
        ("Naive Eager PyTorch", naive_pytorch_softmax),
        ("torch.compile (Inductor)", compiled_softmax),
        ("Custom Triton Kernel", triton_softmax),
    ]:
        ms, bw = benchmark_fn(fn, x)
        print(f"{name:<26} | Latency: {ms:.4f} ms | Effective BW: {bw:.1f} GB/s")

• ප්‍රතිඵල පෙන්වන දේ: H100 එකක (උපරිම HBM කලාප පළල 3,350 GB/s), Naive Eager PyTorch සාමාන්‍යයෙන් ළඟා වන්නේ ~600–800 GB/s ඵලදායී කලාප පළලකට පමණි. එයට හේතුව එය වෙනම කර්නල් 5ක් දියත් කර අතරමැදි ටෙන්සර් නැවත නැවත VRAM තුළ ලිවීමයි. torch.compile සහ අභිරුචි Triton Kernel යන දෙකම සම්පූර්ණ දාමය එක් වාරයකට ඒකාබද්ධ කර, නිරන්තරයෙන් 2,600–3,000+ GB/s වෙත ළඟා වේ (න්‍යායික දෘඪාංග උපරිමයේ 80–90%).

2. දෘඪාංග ප්‍රමාණන සැලැස්ම: නිෂ්පාදන පරිසරයේ 70B LLM සේවාවක් සැපයීම

පොකුරක ප්‍රමාණය නිවැරදිව තීරණය කිරීමට ස්ථිතික බර මතකය, ගතික KV හැඹිලි මතකය සහ මතක කලාප පළලෙන් සීමා වන විකේතන ධාරිතාව ගණනය කළ යුතුය.

ඉලක්ක වැඩබර උපකල්පන

• මාදිලිය: Llama-3-ශෛලියේ 70B Dense Model (ස්තර 80ක්, GQA හරහා KV ශීර්ෂ 8ක්, ශීර්ෂ මානය 128).

• නිරවද්‍යතාව: FP8 බර (පරාමිතියකට බයිට් 1ක්) සහ FP8 KV හැඹිලිය (මූලද්‍රව්‍යයකට බයිට් 1ක්).

• සමගාමීත්වය සහ සන්දර්භය: සමගාමී පරිශීලකයන් 64ක්, එක් අයකුට ටෝකන 8,192ක සන්දර්භ කවුළුවක් බැගින්.

• සලකා බලන දෘඪාංගය: NVIDIA H100 80GB SXM (80 GB HBM3, 3.35 TB/s මතක කලාප පළල).

පියවර 1: ස්ථිතික මාදිලි බර සඳහා VRAM ගණනය කරන්න

බර VRAM=70× 10^(9) පරාමිති× බයිට් 1 (FP8)≈ 70 GB

පියවර 2: ගතික KV හැඹිලි VRAM ගණනය කරන්න

සෑම ටෝකනයක් සඳහාම එක් එක් ස්තරය, මානය 128 වන GQA ශීර්ෂ 8ක් පුරා Key දෛශිකයක් සහ Value දෛශිකයක් ගබඩා කරයි:

ටෝකනයකට KV බයිට් = 2 (K සහ V) × ස්තර 80 × KV ශීර්ෂ 8 × මාන 128 × බයිට් 1 = බයිට් 163,840 ≈ 160 KB/ටෝකන

සමගාමී පරිශීලකයන් 64කට එක් අයකුට ටෝකන 8,192 බැගින් (මුළු සක්‍රිය ටෝකන 524,288):

මුළු KV හැඹිලි VRAM=ටෝකන 524,288× බයිට් 163,840≈ 85.9 GB

පියවර 3: GPU ගණන සහ Tensor Parallelism තීරණය කරන්න

අවශ්‍ය මුළු VRAM=70 GB (බර)+85.9 GB (KV හැඹිලිය)+~15 GB (CUDA Graphs/සක්‍රියතා)≈ 171 GB

• අවම දෘඪාංග: NVLink හරහා සම්බන්ධ කළ 4× H100 80GB GPUs (මුළු HBM 320 GB), Tensor Parallelism = 4 (TP=4) සමඟ ධාවනය වේ.

• එක් GPU එකක මතක භාවිතය: එක් GPU එකක් 17.5 GB බෙදා වෙන් කළ බර + 21.5 GB KV හැඹිලිය + ~4 GB ධාවන කාල අතිරේක භාවිතය රඳවා ගනී = 80 GB න් ~43 GB භාවිත වේ. එමඟින් තදබදය හදිසියේ වැඩි වීම සඳහා ප්‍රමාණවත් අමතර ඉඩක් ඉතිරි වේ.

පියවර 4: ටෝකන ජනන ධාරිතාව (විකේතන වේගය) ඇස්තමේන්තු කරන්න

විකේතනයේදී පරිශීලකයන් 64 දෙනාටම එක් පියවරක් ජනනය කිරීමට, මුළු 70 GB මාදිලි බර සහ 85.9 GB KV හැඹිලිය HBM වෙතින් එක් වරක් කියවිය යුතුය (පියවරකට කියවන මුළු දත්ත 155.9 GB). මෙය GPU 4හි ඒකාබද්ධ මතක කලාප පළල හරහා සිදු වේ (4 × 3.35 TB/s = 13.4 TB/s, ~75% බස් කාර්යක්ෂමතාවක් උපකල්පනය කළ විට = 10.05 TB/s ඵලදායී):

එක් විකේතන පියවරකට කාලය (TPOT)≈ (155.9 GB)/(10,050 GB/s)≈ ටෝකනයකට 15.5 ms

එක් පරිශීලකයකුගේ වේගය≈ (1000 ms)/(15.5 ms)≈ ටෝකන 64.5/තත්පරය/පරිශීලක | මුළු පද්ධති ධාරිතාව=64× 64.5≈ ටෝකන 4,128/තත්පරය

3. AI සම්පාදක ස්ටැක්: Python සිට PTX Assembly දක්වා පහළ මට්ටම්වලට පරිවර්තනය

නවීන AI සම්පාදක බහු-අදියර Intermediate Representation (IR) නළමාර්ගයක් හරහා ඉහළ-මට්ටමේ ගතික Python කේතය සෘජුව දෘඪාංගයේ ධාවනය වන GPU යන්ත්‍ර උපදෙස් බවට පරිවර්තනය කරයි.

• MLIR (Multi-Level Intermediate Representation) නවීන AI සම්පාදකවල ප්‍රමුඛ වන්නේ ඇයි.

• ඓතිහාසිකව, සම්මත LLVM වැනි සම්පාදක ක්‍රියා කළේ පහළ-මට්ටමේ අදිශ ලූප සහ දර්ශක මත පමණක් වන අතර ඉහළ-මට්ටමේ න්‍යාස ව්‍යුහය පිළිබඳ දෘෂ්ටිය අහිමි විය.
• MLIR මඟින් සම්පාදකවලට (Triton, XLA සහ Mojo වැනි) වියුක්තකරණ මට්ටම් කිහිපයක් එකවර පවත්වා ගැනීමට ඉඩ දෙයි. ඒවා ඉහළ උපභාෂාවලදී 2D න්‍යාස ටයිල් සහ හවුල් මතක සැකසුම් ගැන තර්ක කර, පසුව ක්‍රමයෙන් LLVM IR, PTX (Parallel Thread Execution) අතථ්‍ය assembly සහ අවසානයේ SASS දක්වා පහළ මට්ටම්වලට පරිවර්තනය කරයි (ඔබගේ නිශ්චිත NVIDIA GPU පරම්පරාව ක්‍රියාත්මක කරන නිවැරදි ද්විමය යන්ත්‍ර කේතය).