|
|
import torch |
|
|
import time |
|
|
import triton |
|
|
import triton.language as tl |
|
|
|
|
|
@triton.jit |
|
|
def vortex_final_form_kernel(X, Out, N, BLOCK_SIZE: tl.constexpr): |
|
|
pid = tl.program_id(0) |
|
|
offsets = pid * BLOCK_SIZE + tl.arange(0, BLOCK_SIZE) |
|
|
mask = offsets < N |
|
|
|
|
|
|
|
|
x = tl.load(X + offsets, mask=mask) |
|
|
|
|
|
|
|
|
|
|
|
res = x |
|
|
for i in range(50): |
|
|
res = tl.maximum(res * 1.01, 0.0) |
|
|
|
|
|
|
|
|
tl.store(Out + offsets, res, mask=mask) |
|
|
|
|
|
def run_final(): |
|
|
N = 1024 * 1024 * 256 |
|
|
print("--- BLITZ VORTEX: FINAL FORM (H200) ---") |
|
|
X = torch.randn(N, device="cuda") |
|
|
Out = torch.empty_like(X) |
|
|
|
|
|
|
|
|
torch.cuda.synchronize() |
|
|
start = time.time() |
|
|
for _ in range(5): |
|
|
curr = X |
|
|
for _ in range(50): curr = torch.clamp(curr * 1.01, min=0.0) |
|
|
torch.cuda.synchronize() |
|
|
eager_ms = (time.time() - start) / 5 * 1000 |
|
|
|
|
|
|
|
|
grid = (triton.cdiv(N, 16384),) |
|
|
torch.cuda.synchronize() |
|
|
start = time.time() |
|
|
for _ in range(5): vortex_final_form_kernel[grid](X, Out, N, BLOCK_SIZE=16384) |
|
|
torch.cuda.synchronize() |
|
|
vortex_ms = (time.time() - start) / 5 * 1000 |
|
|
|
|
|
print(f"Eager Latency: {eager_ms:.4f}ms") |
|
|
print(f"Blitz Latency: {vortex_ms:.4f}ms") |
|
|
print(f"SILICON ART SPEEDUP: {eager_ms/vortex_ms:.2f}x") |
|
|
|
|
|
if __name__ == "__main__": |
|
|
run_final() |
|
|
|