Šī jaunā valoda varētu nogalināt NVIDIA GPU monopolu

Visi Attēli AI ģenerēts autors bez maksas ar NightCafe Studio - skatiet pēdas saiti.

Augstas veiktspējas datoru laikmets ir definēts ar vienu nosaukumu:Tā brīnums.

NVIDIA platforma atbloķēja GPU jaudu, kļūstot par de facto standartu.

Vairāk nekā desmit gadus GPU programmēšana nozīmēja programmēšanu CUDA.

Tomēr šī dominance ir radījusi būri, bloķējot progresu vienā piegādātājs.

Bet šodien, 2025. gada vidū - lietas mainās.

Bet šodien, 2025. gada vidū - lietas mainās.

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

Mēs redzam specializētās aparatūras izplatīšanos:

• 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.

Un arvien vairāk jaunuzņēmumu, kas izveido pielāgotus silīcija mikroshēmas, katru dienu parādās.

Šī jaunā, daudzveidīgā ainava prasa jaunu programmēšanas filozofiju.

Multi-Level Intermediate Representation (MLIR) un Mojo programmēšanas valoda

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

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. Mēs izmantosim pilnus, darba kodeksa piemērus, lai sniegtu konkrētu, praktisku salīdzinājumu.
2. Mēs izpētīsim, kāpēc MLIR ir pārsteigums salīdzinājumā ar tā cienījamo priekšgājēju LLVM.
3. Mēs apgalvosim, ka Mojo ir augstākais ilgtermiņa risinājums.
4. Mēs analizēsim, kāpēc šis jaunais kaudze ir izmaksu un ātruma spēļu mainītājs.

This impact extends to critical emerging domains such as Generative AI, Quantum ComputingUn patBlockchain.

Mēs skatīsimies arī uz nākotni, aptverotmining ASICs,Neuromorphic Computingunspecialized hardwarepar retām datu plūsmām, ko GPU apstrādā slikti.

Tas ir stāsts par viena laikmeta beigu un jauna laikmeta rītausmu.

Tas ir stāsts par viena laikmeta beigu un jauna laikmeta rītausmu.

Lai saprastu šīs izmaiņas lielumu, mums vispirms irunderstand the four key players.

1. CUDA: The Powerful, Proprietary Incumbent

CUDA stands for Compute Unified Device Architecture.

Tā ir NVIDIA paralēlās datortehnikas platforma un programmēšanas modelis.

Tas ļauj izstrādātājiem rakstīt C++ līdzīgu kodu, ko sauc par kodoliem, kas darbojas NVIDIA GPU.

CUDA's Strengths:

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.

Tas nodrošina tiešu, zema līmeņa kontroli pār aparatūru, ļaujot ekspertiem sasniegt maksimālo veiktspēju.

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.

Tas apvieno izstrādātājus un veselas nozares vienā, dārgā aparatūras piegādātājs.

Tas nomāc konkurenci un ierobežo brīvību izvēlēties labāko aparatūru darbam.

Divu valodu problēma: nozīmīgs trūkums AI un zinātniskajā datorā.

Pētnieki prototipu augsta līmeņa valodā, piemēram, Python par tās vienkāršību un iterācijas ātrumu.

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

Tas rada sāpīgu un dārgu atvienošanos, palēninot ceļu no pētniecības līdz izvietošanai.

Programmēšanas sarežģītība:

CUDA ir spēcīgs, bet notoriāli sarežģīts un verbozs.

Izstrādātājs ir spiests būt manuālais atmiņas pārvaldnieks, kas pārsūta datus starp CPU (uzņēmēju) un GPU (ierīci).

Izstrādātājs ir arī aparatūras plānotājs, kas pārvalda pavedienu blokus, režģus un sinhronizāciju.

Šī sarežģītība ir strauja mācīšanās līkne un bieža smalku kļūdu avots.

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

LLVM projekts ir modulāru un atkārtoti izmantojamu kompilatoru tehnoloģiju kolekcija.

Tās kodols ir LLVM starpposma pārstāvība (IR), zema līmeņa, montāžas līdzīga valoda.

LLVM kļuva par standartu mūsdienu kompilatoru aizmugures, jo īpaši CPU.

Compiler frontend (piemēram, Clang C++) tulko avota kodu LLVM IR.

LLVM backend pēc tam optimizē šo IR un pārvērš to mašīnkodā konkrētam CPU.

Šī modularitāte bija revolucionāra savam laikam.

Tomēr LLVM tika izstrādāts CPU centrālai pasaulei.

Tās IR ir pārāk zems līmenis jaunajai heterogēnas aparatūras pasaulei.

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 dzimis Google no nepieciešamības kompilēt TensorFlow CPU, GPU un to TPU.

Viņi saprata, ka LLVM vienīgais, zema līmeņa IR nebija pietiekams.

MLIR sasniegums ir vienota infrastruktūra vairāku IR definēšanai un sastādīšanai.

Šos kompostējamos IR sauc par dialektiem.

MLIR is like a universal translator, fluent in everything from high-level concepts to low-level machine details.

Augsta līmeņa dialekts var tieši pārstāvēt domēna specifiskus jēdzienus.

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

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

Tas saglabā kritisko semantisko informāciju, ko LLVM noraida.

Tas ļauj spēcīgu kompilācijas stratēģiju, ko sauc parPakāpenisks samazinājums• * *

1. Kompilators sākas ar augsta līmeņa dialekta pārstāvniecību.
2. Tas veic augsta līmeņa, domēna specifiskas optimizācijas šajā dialektā.
3. Tad tas pakāpeniski "mazina" kodu, izmantojot virkni starpposma dialektu.
4. Katrs starpposma dialekts veic savas specifiskās optimizācijas.
5. Visbeidzot, tas sasniedz zema līmeņa dialektu, piemēram, LLVM IR dialektu, galīgajai mašīnkodes ģenerācijai.

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

This enables vastly superior optimizations for any hardware target.

MLIR ir trūkstošā saikne starp augsta līmeņa valodām un daudzveidīgu silīciju.

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

Ja MLIR ir spēcīgs, sarežģīts dzinējs, Mojo ir gluda, intuitīva lietotāja saskarne.

Mojo izveidoja Chris Lattner, sākotnējais LLVM un Swift valodas arhitekts.

Mojo izveidoja Chris Lattner, sākotnējais LLVM un Swift valodas arhitekts.

Tas ir izstrādāts no pirmajiem principiem, lai būtu ideāla valoda MLIR laikmetam.

Šajā ziņā tā ir mūsdienās tehnoloģiski visattīstītākā valoda.

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

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

Mojo's Key Features:

Python pārveidošana

• Mojo mērķis ir pilnīga saderība ar esošo Python ekosistēmu.
• Tā ir slepkava īpašība!
• Tas ļauj izstrādātājiem importēt un izmantot jebkuru Python bibliotēku, piemēram, NumPy, Pandas vai Matplotlib.
• Tas pilnībā apiet "aukstā sākuma" problēmu, ar kuru saskaras jaunas valodas, piesaistot Python milzīgo ekosistēmu.

True Systems programmēšanas funkcijas:

• Unlike Python, Mojo is a compiled language with strong static typing.
• Tas novērš veselas darbības laika kļūdu klases un ļauj veikt C++ līmeņa veiktspējas optimizācijas.
• It introduces modern memory management concepts like ownership and borrowing (from Rust) for memory safety without the overhead of a garbage collector.

Pirmās klases MLIR integrācija:

• Mojo exposes the full power of MLIR directly to the developer.
• Programmētāji var rakstīt augsta līmeņa, Pythonic kodu lielākajai daļai savu pieteikumu.
• Ja ir nepieciešama maksimāla veiktspēja, viņi var lejupielādēt konkrētus MLIR dialektus un rakstīt zema līmeņa kodoli.
• Svarīgi ir tas, ka to visu var izdarīt vienā un tajā pašā failā, tajā pašā valodā.

Mojo elegantly solves the "two-language problem."

Full Code Examples and Analysis

Theory is one thing; practice is another.

Nākamie pilni darba kodeksa piemēri -

Tas parādīs dziļās atšķirības starp abām paradigmām.

Example 1: Matrix Multiplication

Tas ir augstas veiktspējas datortehnikas "Hello, World!" un tas skaidri atklāj katras platformas pamatfilosofiju.

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:

Kodeksu dominē boilerplate un zema līmeņa vadība.

1., 2., 3., 6. un 7. soļi ir paredzēti tikai atmiņas pārvaldībai pāri CPU/GPU robežai.

Tas ir garlaicīgi, pakļauti kļūdām un aptver galveno algoritmu.

Tāsglobalatslēgvārds, blockIdx, threadIdx un <<<...>>> sintakse ir CUDA specifiskas aparatūras abstrakcijas.

This code is fundamentally and permanently tied to NVIDIA's hardware architecture.

Faktiskais algoritms - trīs nested loki - ir neliela daļa no kopējā koda.

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

The Full Mojo Implementation

Šī Mojo versija sasniedz to pašu rezultātu ar aizraujošu vienkāršību un spēku.

(Mūziķis)

# 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))
}

Un tas ir viss!

The Mojo Approach is Far Superior

Programmēšana un koncentrēšanās:

• Mojo kods ir tīrs un tieši izsaka algoritmu.
• Programmētājs koncentrējas uz ko (matemātiku), nevis kā (atmiņas pārnesumus).
• There is no manual cudaMalloc, cudaMemcpy, or cudaFree.
• Visas šīs kļūdas ir pazudušas.

Abstraction with Performance:

• The simple nested loops are not what gets executed.
• MLIR balstītais kompilators veic sarežģītas transformācijas.
• Tas pārvērš šo vienkāršo kodu ļoti optimizētā kodolā.
• It can apply tiling, vectorization, and parallelization automatically.
• Programmētājs var pievienot padomus (piemēram, @vectorize vai @parallelize), lai vadītu kompilatoru, panākot kontroli bez sarežģītības.

Portability (The Ultimate Advantage):

• Tas ir izšķirošais punkts.
• To pašu matmul.mojo failu var pārkompilēt, lai to darbinātu ar NVIDIA GPU, AMD GPU, Intel CPU ar AVX512 vai Google TPU.
• Loģika paliek tāda pati; kompilators backend mainās.
• CUDA kods prasītu pilnīgu, dārgu pārrakstīšanu katram jaunam aparatūras mērķim.
• Mojo piedāvā "izpildes pārnesamību", pārtraucot piegādātāja bloķēšanu un nākotnes drošību.

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

Lai uzzinātu vairāk par Mojo, skatiet zemāk esošo rakstu:

Example 2: Gen AI and the Transformer Attention Mechanism

Uzmanības mehānisms ir tādu modeļu kā GPT-4 sirds, un tas ir liels aprēķinu šķērslis.

Optimizing it is critical.

The CUDA Implementation (Conceptual FlashAttention)

FlashAttention ir vēsturisks algoritms, kas manuāli un profesionāli orķestrē datu kustību starp GPU lēno galveno atmiņu (HBM) un tās ātru uz mikroshēmas atmiņu (SRAM), lai samazinātu šķēršļus.

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

Kopā tie ir gandrīz 3000 līnijas garš.

The repository contains thousands of files.

Mācīšanās līkne un onboarding līkne ir gan strauja.

Zemāk ir sniegta vienkāršota versija (AI ģenerēta):

(CUDA 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:

• It is a masterpiece of manual, hardware-specific engineering.
• It achieves incredible performance by treating the GPU like a manually programmed machine.
• Tas padara kodu praktiski neizlasāmu, neuzturamu un nepārvietojamu.
• Tikai daži pasaules klases eksperti var rakstīt vai mainīt šādu kodu.
• Tas ir veiktspējas maksimums slēgtā ekosistēmā, bet arī sarežģītības un stingrības maksimums.

The Conceptual Mojo Implementation

The Mojo version expresses the same Algoritma idejas(tiling, online softmax) augstā līmenī, deleģējot aparatūras orķestri MLIR kompilatoram.

(Dzīve ir tāda :)

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

One file.

Mazāk par 100 vietām.

Nav smadzeņu atkarības.

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:

Mojo is Game-Changing for AI:

Separation of Concerns:

• The Mojo code describes the algorithm.
• CUDA kods apraksta manuālo aparatūras ieviešanu.
• 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:

• AI pētnieks var viegli saprast un mainīt šo Mojo kodu, lai pārbaudītu jaunu ideju.
• CUDA koda modificēšana būtu milzīgs, laikietilpīgs inženiertehniskais projekts, kam nepieciešama reta prasmju kopa.
• This dramatically accelerates the research and development cycle.

Hardware Freedom:(Viens no svarīgākajiem

• 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!
• MLIR dialektus var paplašināt, lai atbalstītu jebkuru jaunu aparatūru:
• Making the Mojo code truly future-proof.

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

Specialized Hardware and Future Domains

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.

Blockchain, Mining, and ASICs

Darba pierādījumu blokķēdēm, piemēram, Bitcoin, ir nepieciešama milzīga hashing jauda.

The goal is to find a "nonce" that, when hashed with other data, produces a result below a certain target.

Šī ir brute force meklēšana, kas ir ideāli piemērota paralēlai aparatūrai.

Sākotnēji kalnrači izmantoja CPU, tad GPU, lai nodrošinātu augstāko paralēli.

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

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

ASIC (Application-Specific Integrated Circuit) ir mikroshēma, kas paredzēta vienam mērķim - algoritma ieviešanai aparatūrā.

ASIC (Application-Specific Integrated Circuit) ir mikroshēma, kas paredzēta vienam mērķim - algoritma ieviešanai aparatūrā.

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

Tas ir tūkstošiem reižu energoefektīvāks nekā GPU vienam uzdevumam.

This is where the CUDA story ends, but the MLIR/Mojo story gets even more interesting.

The process of designing a chip is called High-Level Synthesis (HLS).

HLS rīki pārvērš augsta līmeņa algoritma aprakstu zema līmeņa aparatūras apraksta valodā (piemēram, Verilog vai VHDL), ko izmanto mikroshēmas izgatavošanai.

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

1. Izstrādātājs varētu rakstīt hashing algoritmu Mojo.
2. GPU ieguves gadījumā viņi to sastādītu, izmantojot GPU aizmuguri.
3. Lai izveidotu ASIC, viņi varētu sastādīt tieši to pašu Mojo kodu, izmantojot HLS backend.
4. The MLIR infrastructure would lower the high-level Mojo logic into Verilog.

Tas apvieno visu kaudzi, sākot no augsta līmeņa programmatūras līdz pielāgotam silīcija dizainam.

Tas ļauj ātri prototipēt un ieviest jaunus algoritmus uz visefektīvāko iespējamo aparatūru, neatkarīgi no tā, vai tas ir GPU vai pavisam jauns ASIC.

CUDA has no answer to this.

Cuda nav atbildes uz šo jautājumu.

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

Neuromorphic Computing and Sparse Data

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

Tas nozīmē, ka tie ir neticami efektīvi, ja tūkstošiem virzienu visi izpilda vienu un to pašu instrukciju uz dažādiem datiem (piemēram, matricas reizināšana).

Tomēr tie ir ļoti neefektīvi darba slodzēs ar smagu atdalīšanu vai neregulāru datu piekļuvi.

Tas ir saistīts ar “trieciena atšķirību”.

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.

Tas nogalina sniegumu daudzām svarīgām problēmām.

Neuromorphic Computing:

Tā ir smadzeņu iedvesmota datortehnika.

Neiromorfie mikroshēmas, piemēram, Intel Loihi, nav balstītas uz pulksteņiem un blīvu matricu matemātiku.

They are event-driven.

"Neurons" fire a "spike" only when their input potential crosses a threshold.

Šie spraudeņi ceļo uz citām "sinapsēm", kas pēc tam var izraisīt citu neironu uzliesmojumu.

Tas ir ārkārtīgi reti, nozarē smags un asimetrisks process.

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

MLIR ir ideāls risinājums šim nolūkam.

MLIR ir ideāls risinājums šim nolūkam.

1. MLIR ietvaros var izveidot "neiromorfisku dialektu".
2. Šajā dialektā būtu pirmās klases operācijas Spike, Synapse, NeuronUpdate.
3. Izstrādātājs varētu uzrakstīt neiromorfisko algoritmu Mojo, izmantojot šos augsta līmeņa jēdzienus.
4. The MLIR compiler, with a backend for a specific neuromorphic chip like Loihi, would translate these concepts into the chip's native, event-driven instructions.

Tas ļauj izveidot pārnēsājamu, augsta līmeņa programmēšanas modeli pilnīgi netradicionālai datortehnikai.

The CUDA model is not relevant in this domain.

The CUDA model is not relevant in this domain.

Sparse and Graph Data:

Daudzas reālās pasaules problēmas ir saistītas ar retiem datiem: sociālajiem tīkliem, ieteikumu dzinējiem un zinātniskām simulācijām.

To pārstāvēšana kā blīvās matricas ir izšķērdīga.

To pārstāvēšana kā blīvās matricas ir izšķērdīga.

Tos apstrādājot GPU, rodas neregulāri atmiņas piekļuves modeļi, kas uzvar GPU atmiņas apvienošanas optimizācijas un samazina veiktspēju.

Again, MLIR provides the answer.

1. A "graph dialect" or "sparse tensor dialect" can represent these data structures natively.
2. Tad kompilators var piemērot specializētas optimizācijas, lai risinātu trūkumus.
3. Piemēram, tas var pārkārtot mezglus, lai uzlabotu atmiņas atrašanās vietu vai izmantotu saspiestus uzglabāšanas formātus.

4. 
   

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

This is something that is extremely difficult today.

Un blakus neiespējamam ar CUDA.

Un blakus neiespējamam ar CUDA.

Quantum Computing Simulation

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

Visizplatītākā metode ir vectora simulācija.

The state of an N-qubit quantum system is represented by a vector of 2^N complex numbers.

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

Kvantālais algoritms ir virkne vārtu.

Katrs vārtsargs ir ekvivalents masveida stāvokļa vektora reizināšanai ar ļoti lielu, ļoti reti sastopamu matricu.

Šī ir darba slodze, kas ir gan skaitļošanas intensīva, gan atmiņas joslas platuma saistīta.

NVIDIA ir daudz ieguldījusi šeit ar savu cuQuantum bibliotēku, augstas veiktspējas CUDA balstītu risinājumu.

cuQuantum ir ļoti ātrs NVIDIA GPU, bet tam ir klasiskie CUDA ierobežojumi:

1. Pārdevēja bloķēšana: Jūsu kvantu simulācija ir saistīta ar NVIDIA aparatūru.
2. Zema līmeņa optimizācija: kompilators redz tikai matricas-vektora reizinājumus.
3. Nav domēna priekšrocības: Tam nav kvantu mehānikas optimizācijas, pamatojoties uz LLVM (semantisko plaisu).

The MLIR/Mojo Advantage for Quantum Simulation:

MLIR pieeja ļauj daudz augstāku izlūkošanas līmeni kompilatorā.

1. "Kvantu dialektu" var definēt MLIR.
2. Šis dialekts nepārstāvētu vārti kā matricas; tas tos pārstāvētu kā kvantu objektus: Hadamard, CNOT, Toffoli.
3. A developer would write their quantum circuit in Mojo using these high-level objects.
4. MLIR kompilators pēc tam var veikt kvantu specifiskas optimizācijas, pirms tiek ģenerētas jebkādas matrices.

Piemēram, kompilators zinātu, ka Hadamard vārtu (H) pielietošana divas reizes pēc kārtas ir identitātes operācija un to var pilnībā novērst.

Tas zinātu, ka noteiktas vārtu secības var “apvienot” vienā, efektīvākos vārtiem.

Piemēram, kompilators zinātu, ka Hadamard vārtu (H) pielietošana divas reizes pēc kārtas ir identitātes operācija un to var pilnībā novērst.

Tas zinātu, ka noteiktas vārtu secības var “apvienot” vienā, efektīvākos vārtiem.

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

Pēc šo augsta līmeņa algebrisko vienkāršošanu veikšanas MLIR kompilators pēc tam samazinātu vienkāršoto apriti uz optimizētu retas matricas operāciju secību mērķa aparatūrai.

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.

Tas nodrošina gan augstāku veiktspēju (smart optimizācijas dēļ), gan pilnīgu aparatūras brīvību.

Nvidia iegulda lielas summas kvantu simulācijas aparatūrā un programmatūrā.

Bet tās CUDA-Q platforma joprojām ir balstīta uz LLVM.

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

Final Verdict: Today vs. The Inevitable Future

The Verdict Today (2025):

1. CUDA ir kalna karalis, un kalns ir liels.
2. Tās nobriedusi ekosistēma, plašas bibliotēkas un masveida kopiena ir spēcīgi aktīvi.
3. Komandai, kas jau ir ieguldījusi NVIDIA aparatūrā un nepieciešama tūlītēja produkta piegāde, CUDA ir pragmatiska izvēle.
4. Desmitgades dominēšanas inercija ir spēcīgs spēks.
5. Mojo joprojām ir jauns.
6. Tās ekosistēma aug ar neticamu ātrumu, bet tā vēl nevar atbilst CUDA cīņā pārbaudīto bibliotēku plašumam.

The Verdict for the Long Run:

1. Nākotne ir neviendabīga.
2. Tas nav pieņēmums, tas ir realitāte.
3. Pielāgota AI silīcija pieaugums un atjaunota konkurence no AMD un Intel ir padarījuši piegādātāju bloķēšanu par nepieņemamu biznesa un tehnisko risku.
4. Nākotnes problēmas - trūcīgie dati, neiromorfiskais AI, blokķēžu ieguves un kvantu datortehnikas - neatbilst mūsdienu GPU stingrajam SIMT modelim.
5. MLIR ir vienīgā esošā, nozares atbalstītā arhitektūra, kas izstrādāta, lai atrisinātu šo problēmu.
6. Tās pieņemšana Google, Apple, Intel, AMD un ARM ir skaidrs signāls par tās centrālo lomu kompilatoru nākotnē.
7. Mojo ir vienīgā valoda, kas izveidota, lai izmantotu šo spēku.

Mūziķi :

• Atrisināt divu valodu problēmu
• Kombinē lietderību ar veiktspēju
• Piedāvā vārti uz visu MLIR ekosistēmu.

Pāreja no CUDA uz MLIR balstītu pasauli būs pakāpeniska, bet tā ir neizbēgama.

Tā ir fundamentāla pāreja no slēgta, aparatūras centrēta modeļa uz atvērtu, programmatūras definētu nākotni.

Mojo trūkumi

1. Mojo joprojām tiek izstrādāta.
2. Viņam pat klases vēl nav.
3. Tās trešo pušu bibliotēkas ir mazas, bet aug neticami strauji.
4. Tam ir lietojumprogrammas visur, kur tiek izmantota Python - bet tai ir jāattīstās ar Python.
5. Visa valoda vēl nav atvērtā koda, lai gan eksperti saka, ka tas drīz mainīsies.
6. Tas neatbalsta Windows (jo vēl).
7. Un tas prasa portēšanu uz Android, iOS un Edge IOT sistēmām.

Bet vai tas būs uzvarētājs ilgtermiņā?

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

Secinājums

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

But it is a cage.

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

Un šis fonds ir paredzēts būt MLIR un Mojo.

Un šis fonds ir paredzēts būt MLIR un Mojo.

The simplest reason - the budget.

Tas ir iemesls, kāpēc, ja vien Nvidia nedarbojas, un drīz:

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

References

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
  • The official portal from NVIDIA for downloading the CUDA Toolkit, which includes the compilers, libraries, and tools necessary for CUDA development.
  • https://developer.nvidia.com/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
  • The official GitHub repository containing the source code for the FlashAttention and FlashAttention-2 CUDA kernels.
  • https://github.com/Dao-AILab/flash-attention

Quantum Computing Simulation

• NVIDIA cuQuantum Official Page
  • NVIDIA's official product page for the cuQuantum SDK, outlining its features for accelerating quantum computing simulations on GPUs.
  • https://developer.nvidia.com/cuquantum
• NVIDIA cuQuantum Documentation
  • The detailed technical documentation for the cuQuantum SDK, providing a high-level overview and API references for the libraries.
  • https://docs.nvidia.com/cuda/cuquantum/index.html

Specialized Hardware (Neuromorphic & ASICs)

• Intel Neuromorphic Computing Overview
  • Intel's official overview of their neuromorphic computing research, which discusses the goals of the program and the Loihi research chips.
  • https://www.intel.com/content/www/us/en/research/neuromorphic-computing.html
• 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 tika izmantots šī raksta apzīmējumam un pētījumam.

https://aistudio.google.com/

Visas bildes tika ģenerētas ar autoru NightCafe Studio bez maksas, pieejama zemāk esošajā saitē:

https://creator.nightcafe.studio/