817 čitanja
817 čitanja

Ovaj novi jezik mogao bi ubiti NVIDIA GPU Monopoly

po Thomas Cherickal29m2025/07/11
Read on Terminal Reader

Predugo; Čitati

Multi-Level Intermediate Representation (MLIR) i Mojo programski jezik (Mojo) predstavljaju temeljnu promjenu paradigme u načinu na koji dizajniramo, optimiziramo i raspoređujemo softver za bilo koji hardver.
featured image - Ovaj novi jezik mogao bi ubiti NVIDIA GPU Monopoly
Thomas Cherickal HackerNoon profile picture
0-item
1-item
2-item


Sve slike AI-generiran od strane autora besplatno s NightCafe Studio - vidjeti podnožje za vezu.

Sve slike AI-generiran od strane autora besplatno s NightCafe Studio - vidjeti podnožje za vezu.

Sve slike AI-generiran od strane autora besplatno s NightCafe Studio - vidjeti podnožje za vezu.

Vrijeme visokih performansi računanja definirano je jednim imenom:Čudo je.

Čudo je.

NVIDIA platforma otključala je snagu GPU-a, postajući de facto standard.

Više od desetljeća, programiranje GPU-a značilo je programiranje u CUDA-i.

Ova dominacija, međutim, stvorila je kavez, zaključavajući napredak u jednog prodavatelja.

Ova dominacija, međutim, stvorila je kavez, zaključavajući napredak u jednog prodavatelja.

Ova dominacija, međutim, stvorila je kavez, zaključavajući napredak u jednog prodavatelja.

Ali danas, sredinom 2025. godine - stvari se mijenjaju.

Ali danas, sredinom 2025. godine - stvari se mijenjaju.

The computing world is now undergoing a radical transformation towards heterogeneity.

Vidimo širenje specijaliziranog hardvera:

  • Intel Gaudi Series:

    Intel's Gaudi processors are designed specifically for deep learning training and inference, offering a competitive alternative to Nvidia's GPUs.


  • AMD Instinct MI Series:

    AMD's MI series of GPUs is designed for high-performance computing and AI workloads, providing an alternative to Nvidia's data center GPUs.


  • Groq Tensor Streaming Processor (TSP):

    Groq's TSP architecture is designed for low-latency inference and high throughput, particularly for large language models.


  • Google TPUs (Tensor Processing Units):

    Google's TPUs are custom-designed chips optimized for machine learning workloads, particularly in Google's cloud infrastructure.


  • AWS Trainium:

    AWS Trainium is a chip designed for machine learning training, offering high performance and cost-effectiveness.

I sve više i više startupa koje grade prilagođene silikonske čipove pojavljuju se svaki dan.

I sve više i više startupa koje grade prilagođene silikonske čipove pojavljuju se svaki dan.

Ovaj novi, raznoliki krajolik zahtijeva novu filozofiju programiranja.

Multi-Level Intermediate Representation (MLIR) i Mojo programski jezik

Arcane glyphs? Pretty sure that's not Mojo code...

This is not just another competitor; they represent a fundamental paradigm shift.

Ovo je revolucija u načinu na koji dizajniramo, optimiziramo i implementiramo softver za bilo koji hardver.

This is a revolution in how we design, optimize, and deploy software for any hardware.

This article will deeply explore the architectural chasm between CUDA and MLIR.


  1. Upotrijebit ćemo potpune primjere radnog kodeksa kako bismo pružili konkretnu, praktičnu usporedbu.
  2. Razmotrit ćemo zašto je MLIR prekretnica u odnosu na svog poštovanog prethodnika, LLVM.
  3. Mi ćemo tvrditi da je Mojo vrhunsko dugoročno rješenje.
  4. Analizirat ćemo zašto je ovaj novi stack mijenjač igre za troškove i brzinu.

Ovaj utjecaj se proširuje na kritične nove domene kao što suGenerative AI, Quantum ComputingI čakBlockchain.

Također ćemo gledati u budućnost, pokrivajućimining ASICs,Neuromorphic Computingispecialized hardwareza rijetke tokove podataka koje GPU-i loše rukuju.

Ovo je priča o kraju jednog doba i početku novog.

Ovo je priča o kraju jednog doba i početku novog.

Da bismo shvatili veličinu ove promjene, moramo prvounderstand the four key players.

1. CUDA: The Powerful, Proprietary Incumbent

CUDA: Moćni, vlasnički incumbent

CUDA stands for Compute Unified Device Architecture.

To je NVIDIA paralelno računanje platforma i model programiranja.

Omogućuje programerima da pišu kod poput C++, koji se zove jezgre, koji se pokreću na NVIDIA GPU-ima.

CUDA's Strengths:

Čačevićeve snage:

Its ecosystem of libraries is mature and unmatched:

  • Mathematical Libraries:
    • cuBLAS: For basic linear algebra subprograms (BLAS).
    • cuRAND: For random number generation.
    • cuFFT: For Fast Fourier Transforms.
    • cuSPARSE: For sparse matrix operations.
    • cuTENSOR: For tensor operations.
    • cuSOLVER: For dense and sparse direct solvers.
  • Parallel Algorithm Libraries:
    • nvGRAPH: For graph algorithms.
    • Thrust: For parallel algorithms and data structures.
  • Communication Libraries:
    • NVSHMEM: For partitioned global address space (PGAS) programming.
    • NCCL: For multi-GPU and multi-node collective communication.
  • Deep Learning Libraries:
    • cuDNN: For deep neural network computations.
    • TensorRT: For optimized deep learning inference.
    • Riva: For conversational AI.
    • DALI: For data loading and augmentation for deep learning.

Omogućuje izravnu kontrolu na niskoj razini nad hardverom, što omogućuje vrhunske performanse za stručnjake.

Njegova duga povijest izgradila je masivnu zajednicu s ogromnom dokumentacijom i potporom.

Its long history has built a massive community with vast documentation and support.

CUDA's Fatal Flaw: The Cage

Vendor Lock-In: CUDA code runs only on NVIDIA GPUs.

Samo

To povezuje programere i cijele industrije u jednog, skupog dobavljača hardvera.

To suzbija konkurenciju i ograničava slobodu odabira najboljeg hardvera za posao.

Problem dvostrukog jezika: velika gužva u AI-u i znanstvenom računanju.

Istraživači prototip u jeziku na visokoj razini kao što je Python zbog svoje jednostavnosti i brzine iteracije.

No, za proizvodnju, kod koji je od ključne važnosti za performanse mora biti potpuno prepisan u C++/CUDA na niskoj razini.

But for production, performance-critical code must be completely rewritten in low-level C++/CUDA.

To stvara bolnu i skupu isključenost, usporavajući put od istraživanja do razmještanja.

Kompleksnost programiranja:

CUDA je moćan, ali zloglasno složen i zloglasan.

Izvođač je prisiljen biti ručni menadžer memorije, prijenos podataka između CPU-a (host) i GPU-a (naprava).

Razvijatelj mora biti i programer hardvera, upravlja blokovima nitova, mrežama i sinhronizacijom.

Ova složenost je strma krivulja učenja i čest izvor suptilnih bugova.

2. LLVM: The Foundation and Its "Semantic Gap”

2. LLVM: The Foundation and Its "Semantic Gap”

Projekt LLVM je zbirka modularnih i ponovnih kompilacijskih tehnologija.

Njegov jezik je LLVM Intermediate Representation (IR), jezik niske razine.

LLVM je postao standard za moderne backend kompilatore, posebno za CPU-e.

Frontend kompilera (kao što je Clang za C++) prevodi izvorni kod u LLVM IR.

LLVM backend zatim optimizira ovaj IR i pretvara ga u strojni kod za određeni CPU.

Ova modularnost bila je revolucionarna za svoje vrijeme.

Međutim, LLVM je dizajniran za CPU-centric svijet.

Njegova IR je previše niska za novi svijet heterogenog hardvera.

To gubi ključne informacije na visokoj razini iz izvornog koda, problem poznat kao "semantički jaz".


Na primjer, prilikom sastavljanja TensorFlow modela, znanje da je operacija Convolution gubi se.


LLVM IR vidi samo generičku zbirku krugova i aritmetičkih uputa.


To sprječava kompilator da provodi snažne, domene specifične optimizacije.


Više ne razumije namjeru programera na visokoj razini.


To je suština problema „semantičke razlike“.


I taj je problem ono što je MLIR riješio.

It loses crucial high-level information from the source code, a problem known as the "semantic gap."


For example, when compiling a TensorFlow model, the knowledge that an operation is a Convolution is lost.


LLVM IR only sees a generic collection of loops and arithmetic instructions.


This prevents the compiler from performing powerful, domain-specific optimizations.


It no longer understands the programmer's high-level intent.


This is the essence of the “semantic gap problem.”


And this problem is what MLIR has Solved.

3. MLIR: The Universal Translator for Hardware

MLIR: Univerzalni prevoditelj za hardver

MLIR je rođen u Googleu iz potrebe za kompilacijom TensorFlow za CPU-e, GPU-e i njihove TPU-e.

Shvatili su da LLVM-ov jedinstveni IR niske razine nije dovoljan.

MLIR-ov proboj je jedinstvena infrastruktura za definiranje i sastavljanje višestrukih IR-ova.

Ovi kompozibilni IR-ovi nazivaju se dijalekti.

Dialekti

MLIR je kao univerzalni prevoditelj, tečan u svemu, od koncepata na visokoj razini do detalja stroja na niskoj razini.

Dijalekt visoke razine može izravno predstavljati koncepte specifične za domenu.

For example, a "TensorFlow dialect" has an operation for tf.conv2d.

A "Linear Algebra dialect" has an operation for linalg.matmul.

To zadržava kritične semantičke informacije koje LLVM odbacuje.

To omogućuje snažnu strategiju kompilacije nazvanuPostupno snižavanje• • •

Postupno snižavanje


  1. Kompiler započinje s predstavljanjem dijalekta na visokoj razini.
  2. Izvršava optimizacije na visokoj razini, specifične za domene na ovom dijalektu.
  3. Zatim postupno "spušta" kod kroz niz međusobnih dijalekata.
  4. Svaki intermedijski dijalekt provodi svoje specifične optimizacije.
  5. Konačno, postiže dijalekt niske razine, kao što je LLVM IR dijalekt, za konačnu proizvodnju strojnog koda.
The compiler starts with a high-level dialect representation.Konačno, postiže dijalekt niske razine, kao što je LLVM IR dijalekt, za konačnu proizvodnju strojnog koda.

This process preserves high-level context for as long as possible.

This enables vastly superior optimizations for any hardware target.

MLIR je manjkajuća veza između jezika na visokoj razini i raznolikosti silicija.

MLIR is the missing link between high-level languages and diverse silicon.

MLIR je manjkajuća veza između jezika na visokoj razini i raznolikosti silicija.

4. Mojo: The User-Friendly Face of MLIR's Power

Mojo: Korisnički prijateljsko lice moći MLIR-a

Ako je MLIR moćan, složen motor, Mojo je elegantno, intuitivno korisničko sučelje.

Mojo je stvorio Chris Lattner, izvorni arhitekt LLVM-a i jezika Swift.

Mojo je stvorio Chris Lattner, izvorni arhitekt LLVM-a i jezika Swift.

Dizajniran je iz prvih načela kako bi bio savršen jezik za MLIR eru.

U tom smislu, to je danas najteže tehnološki napredni jezik.

Čak i Rust se temelji na LLVM-u i ima sve nedostatke LLVM-a.

Even Rust is based on LLVM and has all of LLVM’s shortcomings.

Mojo je danas jedini glavni programski jezik zasnovan na MLIR-u.

Mojo is the only major programming language today based on MLIR.

Mojo's Key Features:

A Superset of Python

  • Mojo ima za cilj punu kompatibilnost s postojećim Python ekosustavom.
  • This is a killer feature!
  • It allows developers to import and use any Python library like NumPy, Pandas, or Matplotlib.
  • U potpunosti zaobiđe problem "hladnog početka" s kojim se novi jezici suočavaju koristeći se ogromnim ekosustavom Pythona.

Značajke programiranja pravih sustava:

  • Za razliku od Pythona, Mojo je kompilirani jezik sa snažnim statičkim tipkanjem.
  • This eliminates entire classes of runtime errors and enables C++-level performance optimizations.
  • It introduces modern memory management concepts like ownership and borrowing (from Rust) for memory safety without the overhead of a garbage collector.

First-Class MLIR Integration:

  • Mojo exposes the full power of MLIR directly to the developer.
  • Programeri mogu napisati Pythonic kod na visokoj razini za većinu svojih aplikacija.
  • Kada je potrebna maksimalna učinkovitost, oni mogu koristiti specifične MLIR dijalekte i napisati jezgre niske razine.
  • Crucially, this can all be done within the same file, in the same language.
Crucially, this can all be done within the same file, in the same language.

Mojo elegantno rješava "probleme s dva jezika".

Mojo elegantly solves the "two-language problem."

Full Code Examples and Analysis

Cjeloviti kod primjeri i analiza

Theory is one thing; practice is another.

Sljedeći primjeri cjelovitog radnog kodeksa -

Will demonstrate the profound differences between the two paradigms.

Example 1: Matrix Multiplication

Example 1: Matrix Multiplication

Ovo je "Hello, World!" visokoučinkovitog računanja i jasno otkriva ključnu filozofiju svake platforme.

The Full CUDA Implementation

This is a complete, compilable CUDA program for matrix multiplication.

(CUDA C++)

// Filename: matmul.cu
// To compile: nvcc matmul.cu -o matmul_cuda

#include <iostream>
#include <vector>
#include <cuda_runtime.h>

// Helper to check for CUDA errors
#define CUDA_CHECK(err) { \
    cudaError_t err_code = err; \
    if (err_code != cudaSuccess) { \
        std::cerr << "CUDA Error: " << cudaGetErrorString(err_code) << " at line " << __LINE__ << std::endl; \
        exit(EXIT_FAILURE); \
    } \
}

// CUDA Kernel for Matrix Multiplication (Device Code)
__global__ void matrixMulKernel(float* C, const float* A, const float* B, int N) {
    // Calculate the global row and column index of the element
    int row = blockIdx.y * blockDim.y + threadIdx.y;
    int col = blockIdx.x * blockDim.x + threadIdx.x;

    // Boundary check to avoid accessing out-of-bounds memory
    if (row < N && col < N) {
        float p_value = 0.0f;
        // Each thread computes one element of the result matrix C
        for (int k = 0; k < N; ++k) {
            p_value += A[row * N + k] * B[k * N + col];
        }
        C[row * N + col] = p_value;
    }
}

// Main function (Host Code)
int main() {
    const int N = 256;
    const int size = N * N * sizeof(float);

    // Step 1. Allocate host memory
    std::vector<float> h_A(N * N);
    std::vector<float> h_B(N * N);
    std::vector<float> h_C(N * N);

    // Initialize host matrices
    for (int i = 0; i < N * N; ++i) {
        h_A[i] = static_cast<float>(rand()) / RAND_MAX;
        h_B[i] = static_cast<float>(rand()) / RAND_MAX;
    }

    // Step 2. Allocate device memory
    float *d_A, *d_B, *d_C;
    CUDA_CHECK(cudaMalloc((void**)&d_A, size));
    CUDA_CHECK(cudaMalloc((void**)&d_B, size));
    CUDA_CHECK(cudaMalloc((void**)&d_C, size));

    // Step 3. Copy matrices from host to device
    std::cout << "Copying data from host to device..." << std::endl;
    CUDA_CHECK(cudaMemcpy(d_A, h_A.data(), size, cudaMemcpyHostToDevice));
    CUDA_CHECK(cudaMemcpy(d_B, h_B.data(), size, cudaMemcpyHostToDevice));

    // Step 4. Define kernel launch configuration
    // Use 16x16 threads per block, a common choice
    dim3 threadsPerBlock(16, 16);
    // Calculate the number of blocks needed in each dimension
    dim3 numBlocks((N + threadsPerBlock.x - 1) / threadsPerBlock.x, (N + threadsPerBlock.y - 1) / threadsPerBlock.y);

    // Step 5. Launch the kernel on the device
    std::cout << "Launching kernel..." << std::endl;
    matrixMulKernel<<<numBlocks, threadsPerBlock>>>(d_C, d_A, d_B, N);
    CUDA_CHECK(cudaGetLastError());
    CUDA_CHECK(cudaDeviceSynchronize()); // Wait for the kernel to finish

    // Step 6. Copy the result matrix back from device to host
    std::cout << "Copying result from device to host..." << std::endl;
    CUDA_CHECK(cudaMemcpy(h_C.data(), d_C, size, cudaMemcpyDeviceToHost));

    // Step 7. Free device memory
    CUDA_CHECK(cudaFree(d_A));
    CUDA_CHECK(cudaFree(d_B));
    CUDA_CHECK(cudaFree(d_C));

    std::cout << "CUDA Matrix Multiplication finished successfully." << std::endl;
    // (Optional: Add verification step here)

    return 0;
}

Analysis of the CUDA Code:

Analiza CUDA kodeksa:

The code is dominated by boilerplate and low-level management.

Korak 1, 2, 3, 6 i 7 su isključivo za upravljanje memorijom preko CPU / GPU granice.

To je dosadno, sklono pogreškama i zatvara osnovni algoritam.

Našglobal keyword, blockIdx, threadIdx, and the <<<...>>> syntax are CUDA-specific hardware abstractions.

Ovaj kod je temeljno i trajno vezan uz NVIDIA hardversku arhitekturu.

Stvarni algoritam - tri ugrađene krugove - je mali dio ukupnog koda.

Programerovo mentalno opterećenje troši se na upravljanje hardverom, a ne na sam problem.

The programmer's mental overhead is spent on hardware management, not on the problem itself.

The Full Mojo Implementation

Potpuna implementacija Mojo

This Mojo version achieves the same result with breathtaking simplicity and power.

(može se)

# Filename: matmul.mojo
# To run: mojo matmul.mojo

from memory import DType, Tensor
from random import rand
from time import now

fn matmul_naive(C: Tensor[DType.float32], A: Tensor[DType.float32], B: Tensor[DType.float32]):
    """A naive, high-level implementation of matrix multiplication."""
    let N = A.dim(0)
    let M = A.dim(1)
    let P = B.dim(1)

    for i in range(N):
        for j in range(P):
            var sum: Float32 = 0.0
            for k in range(M):
                sum += A.load(i, k) * B.load(k, j)
            C.store(i, j, sum)

fn main():
    let N = 256
    
    # 1. Allocate and initialize tensors.
    # Mojo's Tensor handles memory allocation automatically.
    # The compiler will place it in the most appropriate memory space.
    var A = Tensor[DType.float32](N, N)
    var B = Tensor[DType.float32](N, N)
    var C = Tensor[DType.float32](N, N)

    for i in range(N):
        for j in range(N):
            A.store(i, j, rand[DType.float32]())
            B.store(i, j, rand[DType.float32]())

    print("Starting Mojo Matrix Multiplication...")
    
    let start_time = now()
    
    # 2. Call the function.
    # The MLIR-based compiler optimizes this high-level code.
    # It can automatically tile, vectorize, and parallelize this code
    # for the target hardware (CPU, GPU, etc.).
    matmul_naive(C, A, B)

    let end_time = now()
    let duration_ms = (end_time - start_time) / 1_000_000.0

    print("Mojo Matrix Multiplication finished successfully.")
    print("Execution time:", duration_ms, "ms")
    # (Optional: Print a corner of the result matrix to verify)
    print("Result C[0,0]:", C.load(0,0))
}

And that is all!

The Mojo Approach is Far Superior

Mojo pristup je daleko superiorniji

Programmability and Focus:

  • Mojo kod je čist i izravno izražava algoritam.
  • Programer se usredotočuje na što (matematički), a ne na kako (prijenos memorije).
  • There is no manual cudaMalloc, cudaMemcpy, or cudaFree.
  • Cijela ta klasa pogrešaka je nestala.

Abstrakcija s izvedbom:

  • The simple nested loops are not what gets executed.
  • Kompiler na osnovi MLIR-a obavlja sofisticirane transformacije.
  • To pretvara ovaj jednostavan kod u visoko optimiziranu jezgru.
  • Može automatski primijeniti pletenje, vektorizaciju i paralelizaciju.
  • Programer može dodati savjete (kao što je @vectorize ili @parallelize) kako bi vodio kompilator, postižući kontrolu bez složenosti.

Portability (The Ultimate Advantage):

  • To je ključna točka.
  • Ista matmul.mojo datoteka može se ponovno kompilirati kako bi se pokrenula na NVIDIA GPU-u, AMD GPU-u, Intel CPU-u s AVX512 ili Google TPU-u.
  • The logic remains the same; the compiler backend changes.
  • CUDA kod bi zahtijevao potpuno, skupo prepisivanje za svaki novi hardverski cilj.
  • Mojo nudi "prenosivost performansi", razbijanje zaključavanja dobavljača i potvrđivanje budućnosti koda.
Mojo offers "performance portability," breaking vendor lock-in and future-proofing the code.

MLIR-based Mojo je nesumnjivo postavljen da zamijeni LLVM-based CUDA, a programeri će uživati u promjeni!

MLIR-based Mojo is undeniably set to replace LLVM-based CUDA, and developers will enjoy the change!

Za više informacija o Mojo, pogledajte članak ispod:

Example 2: Gen AI and the Transformer Attention Mechanism

Primjer 2: Gen AI i mehanizam pažnje transformatora

The "attention" mechanism is the heart of models like GPT-4 and is a major computational bottleneck.

Optimizing it is critical.

Optimizing it is critical.

The CUDA Implementation (Conceptual FlashAttention)

FlashAttention je prekretnički algoritam koji ručno i stručno orkestrira kretanje podataka između GPU-ovog spora glavnog memorije (HBM) i brzog memorije na čipu (SRAM) kako bi se smanjile prepreke.

Pravi kod je tisuće redaka dug i nevjerojatno složen.

The real code is thousands of lines long and incredibly complex.

The links to the components of the full algorithm implementation are given below:

https://github.com/Dao-AILab/flash-attention/blob/main/csrc/flash_attn/src/flash_fwd_kernel.h

https://github.com/Dao-AILab/flash-attention/blob/main/csrc/flash_attn/flash_api.cpp

Together, they are almost 3000 lines long.

The repository contains thousands of files.

The learning curve and the onboarding curve are both steep.

Pojednostavljena verzija (generirana AI) je navedena u nastavku:

(na primjer u C++)

// This is a simplified conceptual view of a FlashAttention-style CUDA kernel.
// The actual implementation is far more complex.

template<typename Kernel_traits>
__global__ void flash_attention_fwd_kernel(Flash_fwd_params params) {

    // 1. Incredibly complex setup code
    // Calculates dozens of pointers and indices for HBM and shared memory (SRAM)
    const int block_row_idx = blockIdx.x;
    const int head_idx = blockIdx.y;
    // ... many more calculations ...

    // 2. Explicitly allocate shared memory tiles for Q, K, V
    // The developer must manage this limited resource manually.
    extern __shared__ char smem[];
    float* sQ = (float*)smem;
    float* sK = sQ + kTileM * kTileK;
    float* sV = sK + kTileN * kTileK;

    // 3. Main loop over the sequence, manually loading blocks
    for (int k_block_idx = 0; k_block_idx < params.k_num_blocks; ++k_block_idx) {

        // Manually orchestrate asynchronous loads from HBM into SRAM
        // to hide memory latency. This is extremely difficult to get right.
        load_qkv_block_from_hbm(params, ...);
        __syncthreads(); // Hard synchronization barrier

        // Manually perform matrix multiplication in fast SRAM
        compute_sram_matmul(sQ, sK, ...);

        // Recompute softmax "online" to avoid writing the huge intermediate
        // attention score matrix back to slow HBM. This is the core trick.
        compute_online_softmax(...);
        __syncthrows();

        // Update the output block
        update_output_block(sV, ...);
    }

    // 4. Manually write the final output block back to HBM
    store_output_to_hbm(params, ...);
}

Analysis of the CUDA/FlashAttention Approach:

Analysis of the CUDA/FlashAttention Approach:
  • It is a masterpiece of manual, hardware-specific engineering.
  • It achieves incredible performance by treating the GPU like a manually programmed machine.
  • To čini kod praktički nečitljiv, neodrživ i neprenosiv.
  • Only a handful of world-class experts can write or modify such code.
  • To predstavlja vrhunac performansi unutar zatvorenog ekosustava, ali i vrhunac složenosti i rigidnosti.

The Conceptual Mojo Implementation

Konceptualna implementacija Mojo

The Mojo version expresses the same algorithmic idea(tiling, online softmax) na visokoj razini, delegirajući orkestriranje hardvera kompilatoru MLIR.

(Moj prijatelj :)

from memory import DType, Tensor
from algorithm import parallelize

struct AttentionParams:
    var is_causal: Bool
    # ... other model parameters

# This function is a high-level, portable description of the FlashAttention algorithm.
fn flash_attention[T: DType](Q: Tensor[T], K: Tensor[T], V: Tensor[T], params: AttentionParams) -> Tensor[T]:
    # Define problem dimensions from input tensors
    let num_batches = Q.dim(0)
    let num_heads = Q.dim(2)
    let seqlen_q = Q.dim(1)
    let seqlen_k = K.dim(1)
    
    # Define tunable tiling parameters. The compiler can use these as hints.
    alias BLOCK_M: Int = 128
    alias BLOCK_N: Int = 64

    # The output tensor
    var O = Tensor[T](Q.dims)

    # The @parallelize decorator tells the compiler to map this function
    # over the available hardware parallelism (e.g., CUDA thread blocks or CPU cores).
    @parallelize(num_batches * num_heads)
    fn compute_head(batch_idx: Int, head_idx: Int):
        
        # Define per-worker accumulators. The compiler will map these
        # to the fastest available memory (e.g., registers or SRAM).
        var o_i = Tensor[T](seqlen_q, V.dim(3))
        var l_i = Tensor[T](seqlen_q) # Stores the denominator of the softmax
        var m_i = Tensor[T](seqlen_q) # Stores the max of each row for stable softmax
        o_i.zero()
        l_i.fill(0.0)
        m_i.fill(-50000.0) # Negative infinity

        # Loop over blocks of the Key/Value sequence
        for j in range(0, seqlen_k, BLOCK_N):
            # 1. Load tiles of K and V.
            # The compiler is responsible for generating the optimal code
            # to move this data from main memory to fast memory.
            let k_j = K.load_tile[BLOCK_N](batch_idx, j, head_idx)
            let v_j = V.load_tile[BLOCK_N](batch_idx, j, head_idx)
            
            # Loop over blocks of the Query sequence
            for i in range(0, seqlen_q, BLOCK_M):
                # 2. Load tile of Q.
                let q_i = Q.load_tile[BLOCK_M](batch_idx, i, head_idx)
                
                # 3. Compute attention scores for the tile. This is a simple matmul.
                let s_ij = q_i @ k_j.transpose()
                
                # Causal masking for decoder models like GPT
                if params.is_causal:
                    # Algorithmic logic, no hardware specifics
                    apply_causal_mask(s_ij, i, j)

                # 4. Perform the "online softmax" update.
                # This is pure mathematical logic, not memory management.
                let m_ij = row_max(s_ij)
                let p_ij = exp(s_ij - m_ij)
                let l_ij = row_sum(p_ij)
                
                let m_new = max(m_i, m_ij)
                let l_new = exp(m_i - m_new) * l_i + exp(m_ij - m_new) * l_ij

                # Update output tile
                o_i = (l_i / l_new * exp(m_i - m_new)) * o_i + (exp(m_ij - m_new) / l_new) * (p_ij @ v_j)

                # Update softmax stats
                l_i = l_new
                m_i = m_new

        # 5. Store the final output. The compiler manages the write-back.
        O.store_tile(batch_idx, head_idx, o_i)
    
    compute_head()
    return O

Jedna datoteka

Manje od 100 mjesta.

No brain-racking dependencies.

Of course, this is just the algorithm, but in the repository, the same algorithm took nearly 3000 LOC with CUDA!

Of course, this is just the algorithm, but in the repository, the same algorithm took nearly 3000 LOC with CUDA!

So now you understand the difference:

So now you understand the difference:

Mojo is Game-Changing for AI:

Mojo je Game-Changing za AI:

Separation of Concerns:

  • The Mojo code describes the algorithm.
  • The CUDA code describes a manual hardware implementation.
  • This is a profound difference.
  • The Mojo programmer can focus on improving the algorithm:
  • While the MLIR compiler focuses on mapping it to silicon.

Research Velocity and Maintainability:

  • An AI researcher can easily understand and modify this Mojo code to test a new idea.
  • Modifikacija CUDA koda bio bi ogroman, vremenski projekt inženjeringa koji zahtijeva rijetku vještinu.
  • This dramatically accelerates the research and development cycle.

Hardware Freedom: (The Most Important)

  • This Mojo code is not tied to NVIDIA.
  • It can be compiled to run on:
    • AMD GPUs
    • Google TPUs
    • Intel Gaudi
    • Custom AI chips.
    • Any architecture there is!
  • Dijalekti MLIR-a mogu se proširiti kako bi podržali bilo koji novi hardver:
  • Učinite Mojo kod stvarno budućnost-proof.

This breaks the NVIDIA monopoly on high-performance AI and will drive down costs.

This breaks the NVIDIA monopoly on high-performance AI and will drive down costs.

Specialized Hardware and Future Domains

Specijalizirani hardver i buduće domene

I said I wanted a futuristic image. The AI art generator delivered. Cool!

The limitations of the CUDA model become even more apparent when we look beyond traditional dense workloads to the future of computing.

The limitations of the CUDA model become even more apparent when we look beyond traditional dense workloads to the future of computing.

MLIR/Mojo is designed for this future.

MLIR/Mojo is designed for this future.

Blockchain, Mining, and ASICs

Blockchain, rudarstvo i ASIC

Proof-of-Work blokčeini poput Bitcoina zahtijevaju ogromnu hashing moć.

Cilj je pronaći "nonce" koji, kada je hashed s drugim podacima, proizvodi rezultat ispod određenog cilja.

Ovo je brute-force pretraživanje, savršeno za paralelni hardver.

U početku, rudari su koristili CPU-e, a zatim GPU-e za njihovu vrhunsku paralelnost.

The CUDA code for a SHA-256 miner is low-level, focused on bitwise and integer operations.

Međutim, za stabilan, nepromjenjiv algoritam kao što je SHA-256, konačni hardver je ASIC.

However, for a stable, unchanging algorithm like SHA-256, the ultimate hardware is an ASIC.


An ASIC (Application-Specific Integrated Circuit) is a chip designed for one single purpose - to implement an algorithm in hardware.

An ASIC (Application-Specific Integrated Circuit) is a chip designed for one single purpose - to implement an algorithm in hardware.

An SHA-256 ASIC has the hashing logic literally baked into the silicon.

To je tisuće puta učinkovitije od GPU-a za taj jedan zadatak.

To je mjesto gdje CUDA priča završava, ali MLIR / Mojo priča postaje još zanimljivija.

To je mjesto gdje CUDA priča završava, ali MLIR / Mojo priča postaje još zanimljivija.

Proces dizajna čipa naziva se High-Level Synthesis (HLS).

HLS alat pretvara opis algoritma na visokoj razini u opisni jezik hardvera na niskoj razini (kao što je Verilog ili VHDL) koji se koristi za proizvodnju čipa.


MLIR, kroz projekte kao što je CIRCT (Circuit IR za kompilere i alate), dizajniran je da bude leđa sljedeće generacije HLS-a.

MLIR, through projects like CIRCT (Circuit IR for Compilers and Tools), is designed to be the backbone of next-generation HLS.


  1. A developer could write a hashing algorithm in Mojo.
  2. For GPU mining, they would compile it using the GPU backend.
  3. Za stvaranje ASIC-a, mogli su sastaviti točno isti Mojo kod koristeći HLS backend.
  4. Infrastruktura MLIR-a smanjila bi Mojo logiku na visokoj razini u Verilog.
Isti Mojo kod

This unifies the entire stack, from high-level software to custom silicon design.

It allows for rapid prototyping and deployment of new algorithms onto the most efficient hardware possible, be it a GPU or a brand new ASIC.

CUDA has no answer to this.

Cuda na to nema odgovora.

It is a software-only solution for a single vendor's programmable hardware.

Neuromorphic Computing and Sparse Data

Neuromorphic Computing and Sparse Data

NVIDIA GPU-i su majstori SIMT-a: Jedinstvena uputa, višestruka nit.

NVIDIA GPUs are masters of SIMT: Single Instruction, Multiple Thread.

This means they are incredibly efficient when thousands of threads are all executing the same instruction on different data (e.g., a matrix multiplication).

Međutim, oni su vrlo neučinkoviti na radnim opterećenjima s teškim razgraničenjem ili nepravilnim pristupom podacima.

This is because of "thread divergence."

Ako nitke u grupi ( "warp") uzimaju različite grane izjave if/else, hardver mora izvršiti oba puta serijalno, a nitke u neaktivnoj stazi jednostavno su isključene.

If threads in a group (a "warp") take different branches of an if/else statement, the hardware must execute both paths serially, with threads in the inactive path simply turned off.

Obojica

To ubija performanse za mnoge važne probleme.

Neuromorphic Computing:

This is a brain-inspired computing paradigm.

Neuromorfni čipovi, kao što je Intelov Loihi, nisu zasnovani na satima i matricijskoj matematici.

Oni su događajno utemeljeni.

They are event-driven.

"Neuroni" pucaju "spike" samo kada njihov ulazni potencijal prelazi prag.

These spikes travel to other "synapses," which may then cause other neurons to fire.

Ovo je iznimno rijedak, teški i asincroni proces.

Pokušavanje simuliranja toga na GPU-u je užasno neučinkovito zbog konstantne divergencije nitova.

Trying to simulate this on a GPU is horrifically inefficient due to constant thread divergence.


MLIR je savršeno rješenje za to.

MLIR je savršeno rješenje za to.

MLIR je savršeno rješenje za to.


  1. U okviru MLIR-a može se stvoriti "neuromorfni dijalekt".
  2. Ovaj dijalekt bi imao operacije prve klase za Spike, Synapse, NeuronUpdate.
  3. Razvijatelj bi mogao napisati neuromorfni algoritam u Mojo-u koristeći ove koncepte na visokoj razini.
  4. MLIR kompilator, s backendom za specifični neuromorfni čip kao što je Loihi, prevodio bi te koncepte u čipove prirodne, događajno utemeljene upute.

To omogućuje prijenosni model programiranja na visokoj razini za potpuno netradicionalni oblik računanja.

Model CUDA nije relevantan u ovom području.

The CUDA model is not relevant in this domain.

Sparse and Graph Data:

Sparse i grafički podaci:

Mnogi stvarni problemi uključuju rijetke podatke: društvene mreže, motore za preporuke i znanstvene simulacije.

Predstavljati ih kao guste matrice je rasipno.

Representing these as dense matrices is wasteful.

Obrada ih na GPU-ima dovodi do nepravilnih uzoraka pristupa memorije, što pobjeđuje optimizacije memorijskog spajanja GPU-a i smanjuje performanse.

Again, MLIR provides the answer.

  1. "Grafski dijalekt" ili "prljavi tenzorski dijalekt" može nativno predstavljati ove podatkovne strukture.
  2. Kompiler zatim može primijeniti specijalizirane optimizacije za rješavanje sparsnosti.
  3. Na primjer, može preuredati čvorove kako bi poboljšao lokaciju memorije ili koristiti komprimirane formate skladištenja.

To omogućuje da se algoritam na visokoj razini napisan u Mojo učinkovito kompilira za rijetke podatke na bilo kojem hardveru.

This allows a high-level algorithm written in Mojo to be efficiently compiled for sparse data on any hardware.

To je nešto što je danas iznimno teško.

And next to impossible with CUDA.

I pored nemoguće s CUDA.

Quantum Computing Simulation

Quantum Computing Simulation

Simulating a quantum computer on a classical computer is essential for developing and testing quantum algorithms.

The most common method is state vector simulation.

Stanje kvantnog sustava N-qubit predstavlja vektor 2^N kompleksnih brojeva.

For just 50 qubits, this vector has 2^50 (over a quadrillion) elements, requiring petabytes of memory.

For just 50 qubits, this vector has 2^50 (over a quadrillion) elements, requiring petabytes of memory.

A quantum algorithm is a sequence of "gates."

Svaka vrata su ekvivalentna množenju masivnog vektorskog stanja s vrlo velikom, vrlo rijetkom matricom.

Ovo je radno opterećenje koje je računalno intenzivno i vezano uz memorije.

NVIDIA je puno uložila ovdje sa svojom cuQuantum knjižnicom, rješenjem koje se temelji na CUDA-u visokih performansi.

cuQuantum je vrlo brz na NVIDIA GPU-ima, ali ima klasična CUDA ograničenja:

  1. Uključivanje dobavljača: Vaša kvantna simulacija vezana je za NVIDIA hardver.
  2. Low-Level Optimization: The compiler sees only matrix-vector multiplications.
  3. Nema prednosti domene: nema optimizacije za kvantnu mehaniku, a temelji se na LLVM-u (semantička praznina).

The MLIR/Mojo Advantage for Quantum Simulation:

MLIR/Mojo prednosti za kvantnu simulaciju:

MLIR pristup omogućuje mnogo višu razinu inteligencije u kompilatoru.

  1. "Kvantni dijalekt" može se definirati u MLIR-u.
  2. Ovaj dijalekt ne bi predstavio vrata kao matrice; on bi ih predstavio kao njihove kvantne objekte: Hadamard, CNOT, Toffoli.
  3. Razvijatelj bi napisao svoj kvantni krug u Mojo-u koristeći ove objekte na visokoj razini.
  4. MLIR kompilator zatim može izvršiti kvantno specifične optimizacije prije nego što se čak i generiraju matrice.
quantum-specific optimizations


Na primjer, kompilator bi znao da je primjena vrata Hadamard (H) dva puta za redom operacija identiteta i može se potpuno eliminirati.

Znao bi da se određene sekvence vrata mogu "spojiti" u jednu, učinkovitiju vrata.

Na primjer, kompilator bi znao da je primjena vrata Hadamard (H) dva puta za redom operacija identiteta i može se potpuno eliminirati.

Znao bi da se određene sekvence vrata mogu "spojiti" u jednu, učinkovitiju vrata.


Ovo je cijela klasa optimizacije koja je nevidljiva CUDA kompilatoru, koji vidi samo generičke matrice, zahvaljujući LLVM-u.

This is an entire class of optimization that is invisible to the CUDA compiler, which only sees generic matrices, thanks to LLVM.

Nakon što je izvršio te algebričke pojednostavnjenja na visokoj razini, MLIR kompilator bi zatim spustio pojednostavnjeno krugove u optimiziranu sekvencu rijetkih operacija matrice za ciljni hardver.

Budući da je sve to izgrađeno na MLIR-u, isti visokokvalitetni kvantni krug napisan u Mojo-u mogao bi biti kompiliran za rad na NVIDIA GPU-u, AMD GPU-u ili CPU-u.

Because this is all built on MLIR, the same high-level quantum circuit written in Mojo could be compiled to run on an NVIDIA GPU, an AMD GPU, or a CPU cluster.

To pruža i veću učinkovitost (zbog pametnije optimizacije) i potpunu slobodu hardvera.

Nvidia ulaže puno u kvantnu simulaciju hardvera i softvera.

But its CUDA-Q platform is still LLVM-based.

Mojo koji se temelji na MLIR-u ne može samo ponuditi naprednu optimizaciju - također nudi jednostavnije programiranje.

MLIR-based Mojo can not just offer advanced optimization - it also offers simpler programming.

Final Verdict: Today vs. The Inevitable Future

Final Verdict: Today vs. The Inevitable Future

Futurism is the in-thing!

The Verdict Today (2025):

Presuda danas (2025):
  1. CUDA je kralj brda, a brdo je veliko.
  2. Njegov zreli ekosustav, opsežne knjižnice i ogromna zajednica su moćna imovina.
  3. Za tim koji već ulaže u NVIDIA hardver i treba odmah isporučiti proizvod, CUDA je pragmatičan izbor.
  4. Inercija desetljeća dominacije je moćna sila.
  5. Mojo je još uvijek mlad.
  6. Njegov ekosustav raste nevjerojatnom brzinom, ali još uvijek ne može odgovarati širini CUDA-inih borbeno testiranih knjižnica.

The Verdict for the Long Run:

Presuda za dugotrajnu vožnju:
  1. The future is heterogeneous.
  2. To nije nagađanje, to je stvarnost.
  3. Porast prilagođenog AI silicija i obnovljena konkurencija od AMD-a i Intel-a učinili su zaključavanje dobavljača neprihvatljivim poslovnim i tehničkim rizikom.
  4. Problemi budućnosti - krhki podaci, neuromorfna AI, rudarenje blokčeina i kvantno računanje - ne uklapaju se uredno u rigidni SIMT model današnjih GPU-a.
  5. MLIR je jedina postojeća, industrijski podržana arhitektura dizajnirana za rješavanje ovog problema.
  6. Njegovo usvajanje od strane Googlea, Applea, Intela, AMD-a i ARM-a jasan je znak njegove središnje uloge u budućnosti kompilatora.
  7. Mojo je jedini jezik koji je izgrađen (još) kako bi iskoristio tu moć.

Mojo :

  • Riješite dvosmjerni problem
  • Kombinacija učinkovitosti s učinkovitosti
  • Omogućuje pristup cijelom MLIR ekosustavu.
Riješite dvosmjerni problemCombines usability with performanceOmogućuje pristup cijelom MLIR ekosustavu.

Prijelaz iz CUDA-e u svijet koji se temelji na MLIR-u bit će postupan, ali neizbježan.

To je temeljni pomak od zatvorenog, hardversko-centricnog modela do otvorene, softversko definirane budućnosti.

Nedostaci Mojo

  1. Mojo je još uvijek u razvoju.
  2. Još uvijek nema razreda.
  3. Njegove treće strane knjižnice su malo, ali rastu nevjerojatnom brzinom.
  4. Ima aplikacije gdje se Python koristi - ali mora evoluirati s Pythonom.
  5. Cijeli jezik još nije otvoren izvor, iako stručnjaci kažu da će se to uskoro promijeniti.
  6. Ne podržava Windows (još uvijek).
  7. I to zahtijeva portiranje na Android, iOS i Edge IOT sustave.

No, hoće li to biti pobjednik u dugoročnom razdoblju?

I believe it will, and developers will be happier with Mojo than CUDA.

Zaključak

CUDA je izgradila impresivnu palaču današnjeg računalstva visokih performansi.

CUDA built the impressive palace of today's high-performance computing.

Ali to je kavez.

But it is a cage.

MLIR i Mojo daju svakom programeru ključ za otključavanje i izgradnju budućnosti na bilo kojoj osnovi koju odaberu.

MLIR and Mojo are handing every developer the key to unlock it and build the future on any foundation they choose.

I ta je zaklada namijenjena da bude MLIR i Mojo.

I ta je zaklada namijenjena da bude MLIR i Mojo.

The simplest reason - the budget.

U proračunu .

To je razlog zašto, osim ako Nvidia pivo, a uskoro:

Ovo će biti kraj dominacije Nvidie - osim ako i oni ne prihvate MLIR!

This will be the end of the dominance of Nvidia - unless they embrace MLIR as well!

I asked for professional wear - does the AI think engineers work in medical labs? Crazy!


referenciji

Official Project Pages

  • MLIR (Multi-Level Intermediate Representation)
    • Text description: The official homepage for the MLIR project, hosted by LLVM. This is the canonical source for documentation, talks, and the project's overall mission statement.
    • https://mlir.llvm.org/
  • Mojo Programming Language
    • The official documentation for the Mojo programming language from Modular, the company that created it. This is the primary resource for learning the language.[2]
    • https://docs.modular.com/mojo/
  • NVIDIA CUDA Toolkit
  • LLVM Compiler Infrastructure Project
    • The main homepage for the LLVM project, which provides an overview of the entire ecosystem, including Clang, LLDB, and other sub-projects. MLIR is a part of this larger project.
    • https://llvm.org/
  • Chris Lattner's Homepage
    • The personal homepage of Chris Lattner, the creator of LLVM, Clang, Swift, MLIR, and Mojo. It provides his work history and links to his talks and papers, offering direct insight into the creation of these technologies.
    • https://nondot.org/sabre/

AI and Attention Mechanism (FlashAttention)

  • FlashAttention Original Paper (arXiv)
    • The original scientific paper, "FlashAttention: Fast and Memory-Efficient Exact Attention with IO-Awareness," which introduced the algorithm. This is the primary source for understanding the technical details and performance benefits.
    • https://arxiv.org/abs/2205.14135
  • FlashAttention-2 Paper (arXiv)
    • The follow-up paper describing FlashAttention-2, which details further optimizations for parallelism and work partitioning to achieve even greater speedups on modern GPUs.
    • https://arxiv.org/abs/2307.08691
  • FlashAttention GitHub Repository

Quantum Computing Simulation

Specialized Hardware (Neuromorphic & ASICs)

  • Intel Neuromorphic Computing Overview
  • CIRCT (Circuit IR Compilers and Tools) Project
    • The official homepage for the CIRCT project, an LLVM/MLIR incubator looking to apply compiler technology to hardware design, including High-Level Synthesis (HLS) for FPGAs and ASICs.
    • https://circt.llvm.org/
  • CIRCT GitHub Repository
    • The official GitHub repository for the CIRCT project, containing the source code, dialects, and tools for hardware compiler design.
    • https://github.com/llvm/circt

Google AI Studio was used for the outline and the research for this article. You can find it here:

https://aistudio.google.com/

Google AI Studio koristio se za osmišljavanje i istraživanje ovog članka.

https://aistudio.google.com/

Sve slike su generirane od strane autora s NightCafe Studio besplatno, dostupno na poveznici ispod:

https://creator.nightcafe.studio/

Sve slike su generirane od strane autora s NightCafe Studio besplatno, dostupno na poveznici ispod:

https://creator.nightcafe.studio/


Trending Topics

blockchaincryptocurrencyhackernoon-top-storyprogrammingsoftware-developmenttechnologystartuphackernoon-booksBitcoinbooks