Todas las Imágenes generadas por AI por el autor de forma gratuita con NightCafe Studio - véase el footer para el enlace.
Todas las Imágenes generadas por AI por el autor de forma gratuita con NightCafe Studio - véase el footer para el enlace.
Todas las Imágenes generadas por AI por el autor de forma gratuita con NightCafe Studio - véase el footer para el enlace.La era de la computación de alto rendimiento ha sido definida por un único nombre:El milagro.
El milagro.La plataforma de NVIDIA desbloqueó el poder de las GPU, convirtiéndose en el estándar de facto.
Durante más de una década, programar una GPU significaba programar en CUDA.
Esta dominación, sin embargo, ha creado una jaula, bloqueando el progreso en un único vendedor.
Esta dominación, sin embargo, ha creado una jaula, bloqueando el progreso en un único vendedor.
Esta dominación, sin embargo, ha creado una jaula, bloqueando el progreso en un único vendedor.Pero hoy, a mediados de 2025 - las cosas están cambiando.
Pero hoy, a mediados de 2025 - las cosas están cambiando.
The computing world is now undergoing a radical transformation towards heterogeneity.
Estamos viendo una proliferación de hardware especializado:
-
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.
Y cada vez más startups que construyen chips de silicio personalizados aparecen todos los días.
Y cada vez más startups que construyen chips de silicio personalizados aparecen todos los días.Este nuevo paisaje diverso requiere una nueva filosofía de programación.
La Representación Intermedia Multi-Level (MLIR) y el lenguaje de programación Mojo
This is not just another competitor; they represent a fundamental paradigm shift.
Esta es una revolución en la forma en que diseñamos, optimizamos y implementamos software para cualquier hardware.
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.
- Utilizaremos ejemplos completos de código de trabajo para proporcionar una comparación concreta y práctica.
- Descubriremos por qué MLIR es un avance sobre su venerable predecesor, LLVM.
- Argumentaremos que Mojo es la solución superior a largo plazo.
- Analizaremos por qué esta nueva pila es un cambiador de juego en términos de costo y velocidad.
Este impacto se extiende a dominios emergentes críticos comoGenerative AI, Quantum ComputingY inclusoBlockchain.
También vamos a mirar hacia el futuro, cubriendomining ASICs, deNeuromorphic Computing, yspecialized hardwarepara flujos de datos escasos que las GPUs manejan mal.
Esta es la historia del fin de una era y el amanecer de una nueva.
Esta es la historia del fin de una era y el amanecer de una nueva.
Para comprender la magnitud de este cambio, debemos primerounderstand the four key players.
1. CUDA: The Powerful, Proprietary Incumbent
CUDA: El poderoso, propietario incumbenteCUDA stands for Compute Unified Device Architecture.
Es la plataforma de computación paralela y el modelo de programación de NVIDIA.
Permite a los desarrolladores escribir código similar a C++, llamado núcleos, que se ejecuta en GPUs NVIDIA.
CUDA's Strengths:
Las fortalezas de Cuda: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.
Proporciona un control directo y de bajo nivel sobre el hardware, lo que permite el rendimiento máximo para los expertos.
Su larga historia ha construido una comunidad masiva con amplia documentación y apoyo.
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.
SóloEsto vincula a desarrolladores y a industrias enteras a un único proveedor de hardware caro.
Esto suprime la competencia y limita la libertad de elegir el mejor hardware para el trabajo.
El problema de dos idiomas: una gran brecha en la IA y la computación científica.
Los investigadores prototipo en un lenguaje de alto nivel como Python por su simplicidad y velocidad de iteración.
Pero para la producción, el código crítico al rendimiento debe ser completamente reescrito en C++/CUDA de bajo nivel.
But for production, performance-critical code must be completely rewritten in low-level C++/CUDA.
Esto crea una desconectividad dolorosa y costosa, ralentizando el camino de la investigación a la implementación.
Complejidad de la programación:
CUDA es poderoso pero notoriamente complejo y verboso.
El desarrollador está obligado a
un gestor de memoria manual, que transfiere datos entre la CPU (host) y la GPU (dispositivo).
El desarrollador también debe ser un programador de hardware, gestionando bloques de hilos, redes y sincronización.
Esta complejidad es una curva de aprendizaje empinada y una fuente frecuente de errores sutiles.
2. LLVM: The Foundation and Its "Semantic Gap”
LLVM: La Fundación y su “Gap Semántico”El Proyecto LLVM es una colección de tecnologías de compiladores modulares y reutilizables.
Su núcleo es el LLVM Intermediate Representation (IR), un lenguaje de bajo nivel, similar a la asamblea.
LLVM se convirtió en el estándar para los backend de compiladores modernos, especialmente para CPU.
Un frontend de compilador (como Clang para C++) traduce el código fuente en LLVM IR.
El backend LLVM optimiza este IR y lo convierte en código de máquina para una CPU específica.
Esta modularidad fue revolucionaria para su tiempo.
Sin embargo, LLVM fue diseñado para un mundo centrado en la CPU.
Su IR es demasiado bajo para el nuevo mundo de hardware heterogéneo.
Se pierde información crucial de alto nivel del código fuente, un problema conocido como "la brecha semántica".
Por ejemplo, al compilar un modelo TensorFlow, se pierde el conocimiento de que una operación es una Convolución.
LLVM IR sólo ve una colección genérica de loop y instrucciones aritméticas.
Esto impide que el compilador realice poderosas optimizaciones específicas de dominio.
Ya no entiende la intención de alto nivel del programador.
Esta es la esencia del “problema de la brecha semántica”.
Y este problema es lo que MLIR ha resuelto.
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: el traductor universal para hardwareMLIR nació en Google de la necesidad de compilar TensorFlow para CPU, GPU y sus TPU.
Se dieron cuenta de que la única IR de bajo nivel de LLVM no era suficiente.
El avance de MLIR es una infraestructura unificada para definir y componer múltiples IRs.
Estos IRs compuestos se llaman dialectos.
DialectosMLIR es como un traductor universal, fluido en todo, desde conceptos de alto nivel a detalles de máquina de bajo nivel.
Un dialecto de alto nivel puede representar conceptos específicos de dominio directamente.
For example, a "TensorFlow dialect" has an operation for tf.conv2d.
A "Linear Algebra dialect" has an operation for linalg.matmul.
Esto retiene la información semántica crítica que el LLVM descarta.
Esto permite una poderosa estrategia de compilador llamadaDisminución progresiva* * El
Disminución progresiva
- El compilador comienza con una representación de dialectos de alto nivel.
- Realiza optimizaciones específicas de dominio de alto nivel en este dialecto.
- Luego, progresivamente "desciende" el código a través de una serie de dialectos intermedios.
- Cada dialecto intermediario realiza sus propias optimizaciones específicas.
- Finalmente, llega a un dialecto de bajo nivel, como el dialecto LLVM IR, para la generación de código de máquina final.
This process preserves high-level context for as long as possible.
This enables vastly superior optimizations for any hardware target.
MLIR es el vínculo que falta entre los idiomas de alto nivel y la diversidad del silicio.
MLIR es el vínculo que falta entre los idiomas de alto nivel y la diversidad del silicio.
MLIR es el vínculo que falta entre los idiomas de alto nivel y la diversidad del silicio.4. Mojo: The User-Friendly Face of MLIR's Power
Mojo: La cara amigable del poder de MLIRSi MLIR es el motor poderoso y complejo, Mojo es la interfaz de usuario elegante e intuitiva.
Mojo fue creado por Chris Lattner, el arquitecto original de LLVM y el lenguaje Swift.
Mojo fue creado por Chris Lattner, el arquitecto original de LLVM y el lenguaje Swift.
Está diseñado desde los primeros principios para ser el lenguaje perfecto para la era MLIR.
En este sentido, es el lenguaje más tecnológicamente avanzado hoy en día.
Incluso Rust se basa en LLVM y tiene todas las deficiencias de LLVM.
Even Rust is based on LLVM and has all of LLVM’s shortcomings.
Mojo es el único lenguaje de programación basado en MLIR.
Mojo is the only major programming language today based on MLIR.
Mojo's Key Features:
Superset de Python
- Mojo aims for full compatibility with the existing Python ecosystem.
- ¡Este es un personaje asesino!
- Permite a los desarrolladores importar y usar cualquier biblioteca de Python como NumPy, Pandas o Matplotlib.
- Completamente elude el problema del "inicio frío" que enfrentan los nuevos idiomas al aprovechar el vasto ecosistema de Python.
Características de la Programación de Sistemas Reales:
- A diferencia de Python, Mojo es un lenguaje compilado con un fuerte tipo estático.
- Esto elimina clases enteras de errores de tiempo de ejecución y permite optimizar el rendimiento a nivel C++.
- Introduce conceptos modernos de gestión de la memoria como la propiedad y el préstamo (de Rust) para la seguridad de la memoria sin la sobrecarga de un colector de basura.
Integración MLIR de primera clase:
- Mojo expone el poder total de MLIR directamente al desarrollador.
- Los programadores pueden escribir código Pythonic de alto nivel para la mayoría de sus aplicaciones.
- Cuando se requiere el máximo rendimiento, pueden descender a utilizar dialectos MLIR específicos y escribir núcleos de bajo nivel.
- Lo que es crucial, todo esto se puede hacer dentro del mismo archivo, en el mismo idioma.
Mojo resuelve elegantemente el "problema de dos idiomas".
Mojo elegantly solves the "two-language problem."
Full Code Examples and Analysis
Ejemplos y análisis de código completoLa teoría es una cosa, la práctica es otra.
Los siguientes ejemplos de código de trabajo completo -
Demostrará las profundas diferencias entre los dos paradigmas.
Example 1: Matrix Multiplication
Example 1: Matrix MultiplicationThis is the "Hello, World!" of high-performance computing, and it clearly reveals the core philosophy of each platform.
The Full CUDA Implementation
Este es un programa completo, compilable CUDA para la multiplicación de matriz.
(en el caso de 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:
Análisis del Código CUDA:The code is dominated by boilerplate and low-level management.
Los pasos 1, 2, 3, 6 y 7 son puramente para gestionar la memoria a través del límite de la CPU / GPU.
Esto es aburrido, propenso al error y obscurece el algoritmo principal.
El
Este código está fundamentalmente y permanentemente ligado a la arquitectura de hardware de NVIDIA.
El algoritmo real, tres circuitos anclados, es una pequeña fracción del código total.
The programmer's mental overhead is spent on hardware management, not on the problem itself.
The programmer's mental overhead is spent on hardware management, not on the problem itself.
The Full Mojo Implementation
Implementación completa de MojoEsta versión de Mojo logra el mismo resultado con una simplicidad y poder impresionantes.
(en español)
# 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
El enfoque de Mojo es mucho superiorProgrammability and Focus:
- The Mojo code is clean and expresses the algorithm directly.
- El programador se centra en lo que (la matemática), no el cómo (las transferencias de memoria).
- No hay manual cudaMalloc, cudaMemcpy, o cudaFree.
- Toda esa clase de errores ha desaparecido.
Abstracción con el rendimiento:
- Las simples cintas niñadas no son lo que se ejecuta.
- The MLIR-based compiler performs sophisticated transformations.
- Esto convierte este código simple en un núcleo altamente optimizado.
- It can apply tiling, vectorization, and parallelization automatically.
- El programador puede agregar pistas (como @vectorize o @parallelize) para guiar al compilador, logrando control sin complejidad.
Portability (The Ultimate Advantage):
- Este es el punto crucial.
- El mismo archivo matmul.mojo puede ser re-compilado para ejecutarse en una GPU de NVIDIA, una GPU de AMD, una CPU de Intel con AVX512 o una TPU de Google.
- The logic remains the same; the compiler backend changes.
- El código CUDA requeriría una reescritura completa y costosa para cada nuevo objetivo de hardware.
- Mojo ofrece "portabilidad de rendimiento", rompiendo el bloqueo del vendedor y garantizando el futuro del código.
El Mojo basado en MLIR está indiscutiblemente configurado para reemplazar el CUDA basado en LLVM, y los desarrolladores disfrutarán del cambio!
MLIR-based Mojo is undeniably set to replace LLVM-based CUDA, and developers will enjoy the change!
Para más información en Mojo, consulte el artículo de abajo:
Example 2: Gen AI and the Transformer Attention Mechanism
Ejemplo 2: Gen AI y el Mecanismo de Atención de TransformadoresEl mecanismo de "atención" es el corazón de modelos como el GPT-4 y es una gran barrera computacional.
La optimización es crítica.
Optimizing it is critical.
The CUDA Implementation (Conceptual FlashAttention)
FlashAttention es un algoritmo de referencia que orquesta manualmente y de manera experta el movimiento de datos entre la memoria principal lenta (HBM) de la GPU y su memoria rápida en chip (SRAM) para reducir los obstáculos.
El código real es de miles de líneas de largo y increíblemente complejo.
The real code is thousands of lines long and incredibly complex.
Los enlaces a los componentes de la implementación completa del algoritmo se dan a continuación:
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
Juntos, son casi 3000 líneas de longitud.
The repository contains thousands of files.
La curva de aprendizaje y la curva de embarque son ambas abruptas.
Una versión simplificada (generada por AI) se da a continuación:
(en el caso de 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:
Análisis del enfoque CUDA/FlashAttention:- It is a masterpiece of manual, hardware-specific engineering.
- Consigue un rendimiento increíble tratando la GPU como una máquina programada manualmente.
- Esto hace que el código sea prácticamente ilegible, no manejable y no portátil.
- Sólo un puñado de expertos de clase mundial pueden escribir o modificar dicho código.
- Representa el pico de rendimiento dentro de un ecosistema cerrado, pero también el pico de complejidad y rigidez.
The Conceptual Mojo Implementation
Implementación conceptual de MojoLa versión de Mojo expresa lo mismoalgorithmic idea(tiling, softmax en línea) a un alto nivel, delegando la orquestación de hardware al compilador MLIR.
(Mojo:)
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
Un solo archivo.
Menos de 100 puntos.
No brain-racking dependencies.
Por supuesto, este es sólo el algoritmo, pero en el repositorio, el mismo algoritmo tomó casi 3000 LOC con CUDA!
Of course, this is just the algorithm, but in the repository, the same algorithm took nearly 3000 LOC with CUDA!
Ahora ya sabes la diferencia:
So now you understand the difference:
Mojo is Game-Changing for AI:
Mojo está cambiando el juego para AI:Separation of Concerns:
- El código Mojo describe el algoritmo.
- El código CUDA describe una implementación manual de hardware.
- Esta es una diferencia profunda.
- El programador de Mojo puede centrarse en mejorar el algoritmo:
- Mientras que el compilador MLIR se centra en mapearlo al silicio.
Research Velocity and Maintainability:
- An AI researcher can easily understand and modify this Mojo code to test a new idea.
- Modificar el código CUDA sería un proyecto de ingeniería masivo y demorado que requeriría un raro conjunto de habilidades.
- Esto acelera drásticamente el ciclo de investigación y desarrollo.
Hardware Freedom:(El más importante)
- Este código Mojo no está vinculado a NVIDIA.
- It can be compiled to run on:
- AMD GPUs
- Google TPUs
- Intel Gaudi
- Custom AI chips.
- Any architecture there is!
- Los dialectos de MLIR se pueden ampliar para soportar cualquier nuevo hardware:
- Making the Mojo code truly future-proof.
Esto rompe el monopolio de NVIDIA en la IA de alto rendimiento y disminuirá los costes.
This breaks the NVIDIA monopoly on high-performance AI and will drive down costs.
Specialized Hardware and Future Domains
Hardware especializado y dominios futurosLas limitaciones del modelo CUDA se vuelven aún más evidentes cuando miramos más allá de las cargas de trabajo densas tradicionales al futuro de la computación.
The limitations of the CUDA model become even more apparent when we look beyond traditional dense workloads to the future of computing.
MLIR/Mojo está diseñado para este futuro.
MLIR/Mojo is designed for this future.
Blockchain, Mining, and ASICs
Blockchain, Minería y ASICLas Blockchains Proof-of-Work como Bitcoin requieren un enorme poder de hash.
El objetivo es encontrar una "nonce" que, cuando se hash con otros datos, produzca un resultado por debajo de un determinado objetivo.
Esta es una búsqueda de fuerza bruta, perfecta para hardware paralelo.
Inicialmente, los mineros usaron CPUs, luego GPUs para su paralelismo superior.
El código CUDA para un minero SHA-256 es de bajo nivel, centrado en operaciones bitwise y enteras.
Sin embargo, para un algoritmo estable y inalterable como SHA-256, el hardware final es un ASIC.
However, for a stable, unchanging algorithm like SHA-256, the ultimate hardware is an ASIC.
Un ASIC (Application-Specific Integrated Circuit) es un chip diseñado para un único propósito: implementar un algoritmo en el hardware.
Un ASIC (Application-Specific Integrated Circuit) es un chip diseñado para un único propósito: implementar un algoritmo en el hardware.
An SHA-256 ASIC has the hashing logic literally baked into the silicon.
Es miles de veces más eficiente que una GPU para esa tarea.
Aquí termina la historia de CUDA, pero la historia de MLIR/Mojo se vuelve aún más interesante.
Aquí termina la historia de CUDA, pero la historia de MLIR/Mojo se vuelve aún más interesante.El proceso de diseño de un chip se llama síntesis de alto nivel (HLS).
Las herramientas HLS convierten una descripción de alto nivel de un algoritmo en un lenguaje de descripción de hardware de bajo nivel (como Verilog o VHDL) utilizado para fabricar el chip.
MLIR, a través de proyectos como CIRCT (Circuit IR for Compilers and Tools), está diseñado para ser la columna vertebral de la próxima generación de HLS.
MLIR, through projects like CIRCT (Circuit IR for Compilers and Tools), is designed to be the backbone of next-generation HLS.
- Un desarrollador podría escribir un algoritmo de hash en Mojo.
- Para la minería de GPU, la compilarían usando el backend de GPU.
- Para crear un ASIC, podrían compilar exactamente el mismo código Mojo usando un backend de HLS.
- La infraestructura MLIR bajaría la lógica de Mojo de alto nivel a Verilog.
Esto unifica toda la pila, desde el software de alto nivel hasta el diseño de silicio personalizado.
Permite el prototipo rápido y la implementación de nuevos algoritmos en el hardware más eficiente posible, ya sea una GPU o un ASIC completamente nuevo.
CUDA no tiene respuesta a esto.
CUDA no tiene respuesta a esto.
It is a software-only solution for a single vendor's programmable hardware.
Neuromorphic Computing and Sparse Data
Computación Neuromórfica y Sparse DataLas GPUs de NVIDIA son maestros de SIMT: Single Instruction, Multiple Thread.
NVIDIA GPUs are masters of SIMT: Single Instruction, Multiple Thread.
Esto significa que son increíblemente eficientes cuando miles de hilos están ejecutando la misma instrucción en datos diferentes (por ejemplo, una multiplicación de matriz).
Sin embargo, son muy ineficientes en cargas de trabajo con ramificación pesada o acceso a datos irregular.
This is because of "thread divergence."
Si los hilos en un grupo (un "warp") toman diferentes ramas de una declaración si/else, el hardware debe ejecutar ambos caminos serialmente, con los hilos en el camino inactivo simplemente apagados.
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.
AmbosEsto mata el rendimiento para muchos problemas importantes.
Neuromorphic Computing:
Este es un paradigma informático inspirado en el cerebro.
Los chips neuromórficos, como el Loihi de Intel, no se basan en relojes y matemáticas de matriz densa.
Son conducidos por eventos.
They are event-driven.
“Neurones” disparan un “spike” sólo cuando su potencial de entrada cruza un umbral.
Estos picos viajan a otras "sinapses", que pueden entonces causar que otras neuronas se incendien.
Este es un proceso extremadamente escaso, ramificado y asíncrono.
Tratar de simular esto en una GPU es horriblemente ineficiente debido a la constante divergencia de filamentos.
Trying to simulate this on a GPU is horrifically inefficient due to constant thread divergence.
MLIR es la solución perfecta para esto.
MLIR es la solución perfecta para esto.
MLIR es la solución perfecta para esto.
- Se puede crear un "dialecto neuromórfico" dentro de MLIR.
- Este dialecto tendría operaciones de primera clase para Spike, Synapse, NeuronUpdate.
- Un desarrollador podría escribir un algoritmo neuromórfico en Mojo usando estos conceptos de alto nivel.
- El compilador MLIR, con un backend para un chip neuromórfico específico como Loihi, traduciría estos conceptos en las instrucciones nativas, impulsadas por eventos del chip.
This allows for a portable, high-level programming model for a completely non-traditional form of computing.
El modelo CUDA no es relevante en este dominio.
The CUDA model is not relevant in this domain.
Sparse and Graph Data:
Datos Sparse y Graph:Muchos problemas del mundo real implican datos escasos: redes sociales, motores de recomendación y simulaciones científicas.
Representarlos como matrizes densas es un desperdicio.
Representarlos como matrizes densas es un desperdicio.
Procesarlos en GPUs conduce a patrones irregulares de acceso a la memoria, lo que derrota las optimizaciones de coalescimiento de la memoria de la GPU y reduce el rendimiento.
Again, MLIR provides the answer.
- Un "dialecto gráfico" o "dialecto tensor escaso" puede representar estas estructuras de datos nativamente.
- El compilador puede entonces aplicar optimizaciones especializadas para manejar la escasez.
- Por ejemplo, puede reordenar los nodos para mejorar la localización de la memoria o utilizar formatos de almacenamiento comprimidos.
-
Esto permite que un algoritmo de alto nivel escrito en Mojo sea compilado de manera eficiente para datos escasos en cualquier hardware.
This allows a high-level algorithm written in Mojo to be efficiently compiled for sparse data on any hardware.
Esto es algo que es extremadamente difícil hoy en día.
Y al lado del imposible con CUDA.
Y al lado del imposible con CUDA.
Quantum Computing Simulation
Simulación de Computación CuánticaSimulating a quantum computer on a classical computer is essential for developing and testing quantum algorithms.
El método más común es la simulación vectorial de estado.
El estado de un sistema cuántico N-qubit es representado por un vector de números complejos 2^N.
Para sólo 50 qubits, este vector tiene 2^50 (más de un cuadrillón) de elementos, que requieren petabytes de memoria.
For just 50 qubits, this vector has 2^50 (over a quadrillion) elements, requiring petabytes of memory.
Un algoritmo cuántico es una secuencia de "puertas".
Cada puerta es equivalente a multiplicar el vector de estado masivo por una matriz muy grande y muy escasa.
Esta es una carga de trabajo que es tanto computacionalmente intensa como de memoria-ancho de banda.
NVIDIA ha invertido mucho aquí con su biblioteca cuQuantum, una solución basada en CUDA de alto rendimiento.
cuQuantum es muy rápido en GPUs NVIDIA, pero tiene las limitaciones clásicas de CUDA:
- Vendor Lock-In: Su simulación cuántica está ligada al hardware de NVIDIA.
- Optimización de bajo nivel: El compilador sólo ve multiplicaciones de matriz-vector.
- No hay ventaja de dominio: no tiene optimizaciones para la mecánica cuántica, basándose en LLVM (la brecha semántica).
The MLIR/Mojo Advantage for Quantum Simulation:
Las ventajas de MLIR/Mojo para la simulación cuántica:El enfoque MLIR permite un nivel mucho mayor de inteligencia en el compilador.
- Un "dialecto cuántico" puede ser definido en MLIR.
- Este dialecto no representaría las puertas como matrices; las representaría como sus objetos cuánticos: Hadamard, CNOT, Toffoli.
- Un desarrollador escribiría su circuito cuántico en Mojo usando estos objetos de alto nivel.
- El compilador MLIR puede entonces realizar optimizaciones cuánticas específicas antes de que se genere alguna matriz.
Por ejemplo, el compilador sabría que aplicar una puerta Hadamard (H) dos veces en una fila es una operación de identidad y puede eliminarse completamente.
Sabría que ciertas secuencias de puertas se pueden "fusionar" en una única puerta más eficiente.
Por ejemplo, el compilador sabría que aplicar una puerta Hadamard (H) dos veces en una fila es una operación de identidad y puede eliminarse completamente.
Sabría que ciertas secuencias de puertas se pueden "fusionar" en una única puerta más eficiente.
Esta es una clase entera de optimización que es invisible para el compilador CUDA, que sólo ve matrices genéricas, gracias a LLVM.
This is an entire class of optimization that is invisible to the CUDA compiler, which only sees generic matrices, thanks to LLVM.
Después de realizar estas simplificaciones algebraicas de alto nivel, el compilador MLIR bajaría el circuito simplificado a una secuencia optimizada de operaciones de matriz escasas para el hardware objetivo.
Debido a que todo esto está construido en MLIR, el mismo circuito cuántico de alto nivel escrito en Mojo podría ser compilado para ejecutarse en una GPU NVIDIA, una GPU AMD o un clúster de CPU.
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.
Esto proporciona un mayor rendimiento (debido a una optimización más inteligente) y completa libertad de hardware.
Nvidia está invirtiendo mucho en hardware de simulación cuántica y la pila de software.
Pero su plataforma CUDA-Q todavía está basada en LLVM.
El Mojo basado en MLIR no solo puede ofrecer optimización avanzada, sino que también ofrece una programación más simple.
MLIR-based Mojo can not just offer advanced optimization - it also offers simpler programming.
Final Verdict: Today vs. The Inevitable Future
Veredicto Final: Hoy vs. El futuro inevitableThe Verdict Today (2025):
The Verdict Today (2025):- CUDA es el rey de la colina, y la colina es grande.
- Su ecosistema maduro, extensas bibliotecas y una comunidad masiva son activos poderosos.
- Para un equipo que ya ha invertido en hardware de NVIDIA y necesita enviar un producto de inmediato, CUDA es la elección pragmática.
- La inercia de una década de dominio es una fuerza poderosa.
- Mojo sigue siendo joven.
- Su ecosistema está creciendo a una velocidad increíble, pero todavía no puede coincidir con la ampliación de las bibliotecas probadas en combate de CUDA.
The Verdict for the Long Run:
La sentencia para la larga carrera:- El futuro es heterogéneo.
- No es una suposición, es una realidad.
- El aumento del silicio de IA personalizado y la renovada competencia de AMD e Intel han convertido el bloqueo de los proveedores en un riesgo empresarial y técnico inaceptable.
- Los problemas del futuro, como los datos escasos, la IA neuromórfica, la minería de blockchain y la computación cuántica, no encajan bien en el rígido modelo SIMT de las GPU de hoy.
- MLIR es la única arquitectura existente y respaldada por la industria diseñada para resolver este problema.
- Su adopción por Google, Apple, Intel, AMD y ARM es una señal clara de su papel central en el futuro de los compiladores.
- Mojo es el único lenguaje construido (aún) para aprovechar este poder.
El Mojo:
- Resolver el problema de los dos idiomas
- Combina usabilidad con rendimiento
- Ofrece una puerta de entrada a todo el ecosistema MLIR.
La transición de CUDA a un mundo basado en MLIR será gradual, pero es inevitable.
Es un cambio fundamental de un modelo cerrado y centrado en el hardware a un futuro abierto y definido por el software.
Las deficiencias de Mojo
- Mojo sigue en desarrollo.
- Aún no tiene clases.
- Sus bibliotecas de terceros son pocas, pero crecen a un ritmo increíble.
- Tiene aplicaciones en todas partes donde se utiliza Python, pero necesita evolucionar con Python.
- El lenguaje entero aún no es de código abierto, aunque los expertos dicen que esto pronto cambiará.
- No es compatible con Windows (aún).
- Y requiere portar a los sistemas Android, iOS y Edge IOT.
¿Será el ganador a largo plazo?
I believe it will, and developers will be happier with Mojo than CUDA.
Conclusión
CUDA construyó el impresionante palacio de la computación de alto rendimiento de hoy.
CUDA built the impressive palace of today's high-performance computing.
Pero es una jaula.
But it is a cage.
MLIR y Mojo están entregando a cada desarrollador la clave para desbloquearlo y construir el futuro sobre cualquier base que elijan.
MLIR and Mojo are handing every developer the key to unlock it and build the future on any foundation they choose.
Y esa fundación está destinada a ser MLIR y Mojo.
Y esa fundación está destinada a ser MLIR y Mojo.
The simplest reason - the budget.
El presupuesto .Por eso, a menos que Nvidia pivote, y pronto:
¡Este será el final del dominio de Nvidia - a menos que ellos también abrazan MLIR!
This will be the end of the dominance of Nvidia - unless they embrace MLIR as well!
Referencias
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 fue utilizado para el resumen y la investigación para este artículo.
Google AI Studio fue utilizado para el resumen y la investigación para este artículo.
Todas las imágenes fueron generadas por el autor con NightCafe Studio de forma gratuita, disponible en el enlace de abajo:
Todas las imágenes fueron generadas por el autor con NightCafe Studio de forma gratuita, disponible en el enlace de abajo:
https://creator.nightcafe.studio/