<?xml version="1.0" encoding="UTF-8"?>
<rss version="2.0" xmlns:atom="http://www.w3.org/2005/Atom" xmlns:dc="http://purl.org/dc/elements/1.1/">
  <channel>
    <title>DEV Community: Sujal Suyash</title>
    <description>The latest articles on DEV Community by Sujal Suyash (@suzuskiee).</description>
    <link>https://dev.to/suzuskiee</link>
    <image>
      <url>https://media2.dev.to/dynamic/image/width=90,height=90,fit=cover,gravity=auto,format=auto/https:%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Fuser%2Fprofile_image%2F4037917%2F305640a9-5fbd-4c0d-93e2-ceb525b73dc8.webp</url>
      <title>DEV Community: Sujal Suyash</title>
      <link>https://dev.to/suzuskiee</link>
    </image>
    <atom:link rel="self" type="application/rss+xml" href="https://dev.to/feed/suzuskiee"/>
    <language>en</language>
    <item>
      <title>Model Quantization Explained: Shrinking LLMs Without Losing Their Mind</title>
      <dc:creator>Sujal Suyash</dc:creator>
      <pubDate>Mon, 10 Aug 2026 17:20:02 +0000</pubDate>
      <link>https://dev.to/suzuskiee/model-quantization-explained-shrinking-llms-without-losing-their-mind-m2</link>
      <guid>https://dev.to/suzuskiee/model-quantization-explained-shrinking-llms-without-losing-their-mind-m2</guid>
      <description>&lt;h1&gt;
  
  
  Model Quantization Explained: Shrinking LLMs Without Losing Their Mind
&lt;/h1&gt;

&lt;p&gt;As we all know that when we try to load a 70B parameter model, how much time and space it takes just to load the model. To solve this issue, the concept of quantization was introduced.&lt;/p&gt;

&lt;p&gt;This post is my own attempt to understand quantization deeply enough to explain it simply — from &lt;em&gt;what&lt;/em&gt; it actually does, to the math behind &lt;em&gt;how&lt;/em&gt; it preserves accuracy.&lt;/p&gt;

&lt;h2&gt;
  
  
  The Core Idea: Precision, Not Parameters
&lt;/h2&gt;

&lt;p&gt;The most common misconception is that quantization means removing parameters from a model. Actually it isn't. &lt;br&gt;
Concretely: quantization converts high-precision parameters (like 32-bit floats) into lower-precision representations (like 8-bit or 4-bit numbers). This reduces:&lt;/p&gt;

&lt;ul&gt;
&lt;li&gt;
&lt;strong&gt;Memory usage&lt;/strong&gt; (fewer bits per value)&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Inference latency&lt;/strong&gt; (less data to move, and on supported hardware, faster math)&lt;/li&gt;
&lt;/ul&gt;
&lt;h2&gt;
  
  
  Floating Point 101: Sign, Exponent, Mantissa
&lt;/h2&gt;

&lt;p&gt;Before we move forward, you need to know what's actually inside a floating-point number. Every float is basically scientific notation in binary:&lt;br&gt;
&lt;/p&gt;

&lt;div class="highlight js-code-highlight"&gt;
&lt;pre class="highlight plaintext"&gt;&lt;code&gt;value = (-1)^sign × 1.mantissa × 2^exponent
&lt;/code&gt;&lt;/pre&gt;

&lt;/div&gt;



&lt;ul&gt;
&lt;li&gt;
&lt;strong&gt;Sign&lt;/strong&gt; — positive or negative&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Exponent&lt;/strong&gt; — how big the number can get (this controls &lt;em&gt;range&lt;/em&gt;)&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Mantissa&lt;/strong&gt; (significand) — the precise digits within that range (this controls &lt;em&gt;precision&lt;/em&gt;)
More mantissa bits → values closer together can be told apart (finer resolution).
More exponent bits → much larger or smaller magnitudes can be represented before overflow/underflow.&lt;/li&gt;
&lt;/ul&gt;

&lt;p&gt;This is exactly what the "E4M3" / "E5M2" naming in FP8 formats tells you — it's the literal bit split between exponent and mantissa.&lt;/p&gt;

&lt;h2&gt;
  
  
  &lt;img src="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Frob81hhjvw6nw4tfm4m9.png" alt=" " width="788" height="637"&gt;
&lt;/h2&gt;

&lt;h2&gt;
  
  
  Common Quantization Formats
&lt;/h2&gt;

&lt;div class="table-wrapper-paragraph"&gt;&lt;table&gt;
&lt;thead&gt;
&lt;tr&gt;
&lt;th&gt;Format&lt;/th&gt;
&lt;th&gt;Bits&lt;/th&gt;
&lt;th&gt;Exponent&lt;/th&gt;
&lt;th&gt;Mantissa&lt;/th&gt;
&lt;th&gt;Typical Use&lt;/th&gt;
&lt;/tr&gt;
&lt;/thead&gt;
&lt;tbody&gt;
&lt;tr&gt;
&lt;td&gt;FP16&lt;/td&gt;
&lt;td&gt;16&lt;/td&gt;
&lt;td&gt;5&lt;/td&gt;
&lt;td&gt;10&lt;/td&gt;
&lt;td&gt;General inference, good precision, limited range&lt;/td&gt;
&lt;/tr&gt;
&lt;tr&gt;
&lt;td&gt;BF16&lt;/td&gt;
&lt;td&gt;16&lt;/td&gt;
&lt;td&gt;8&lt;/td&gt;
&lt;td&gt;7&lt;/td&gt;
&lt;td&gt;Same range as FP32, coarser precision — popular for training&lt;/td&gt;
&lt;/tr&gt;
&lt;tr&gt;
&lt;td&gt;FP8 E4M3&lt;/td&gt;
&lt;td&gt;8&lt;/td&gt;
&lt;td&gt;4&lt;/td&gt;
&lt;td&gt;3&lt;/td&gt;
&lt;td&gt;Weights &amp;amp; forward-pass activations (precision matters more)&lt;/td&gt;
&lt;/tr&gt;
&lt;tr&gt;
&lt;td&gt;FP8 E5M2&lt;/td&gt;
&lt;td&gt;8&lt;/td&gt;
&lt;td&gt;5&lt;/td&gt;
&lt;td&gt;2&lt;/td&gt;
&lt;td&gt;Gradients in mixed-precision training (range matters more)&lt;/td&gt;
&lt;/tr&gt;
&lt;tr&gt;
&lt;td&gt;NVFP4&lt;/td&gt;
&lt;td&gt;4&lt;/td&gt;
&lt;td&gt;—&lt;/td&gt;
&lt;td&gt;—&lt;/td&gt;
&lt;td&gt;Aggressive compression on newer hardware (e.g. Blackwell)&lt;/td&gt;
&lt;/tr&gt;
&lt;/tbody&gt;
&lt;/table&gt;&lt;/div&gt;

&lt;p&gt;The E4M3 vs E5M2 split exists for a reason: forward-pass weights/activations tend to be well-behaved in magnitude, so the extra mantissa bit in E4M3 helps accuracy. Gradients during training can swing across huge magnitude ranges, so E5M2 trades precision for range to avoid overflowing to infinity or underflowing to zero.&lt;/p&gt;




&lt;h2&gt;
  
  
  Symmetric vs Asymmetric Quantization
&lt;/h2&gt;

&lt;p&gt;At its core, quantization maps a continuous float range onto a small set of integers using a &lt;strong&gt;scale factor&lt;/strong&gt; (and sometimes a &lt;strong&gt;zero-point&lt;/strong&gt;).&lt;/p&gt;

&lt;h3&gt;
  
  
  Symmetric Quantization
&lt;/h3&gt;

&lt;p&gt;Used when the value distribution is roughly centered around zero (common for weights). The float range is mapped symmetrically around zero — no offset needed.&lt;br&gt;
&lt;/p&gt;

&lt;div class="highlight js-code-highlight"&gt;
&lt;pre class="highlight plaintext"&gt;&lt;code&gt;scale = max(abs(W)) / (2^(b-1) - 1)

W_quant   = round(W / scale)
W_dequant = W_quant × scale
&lt;/code&gt;&lt;/pre&gt;

&lt;/div&gt;



&lt;p&gt;Where:&lt;/p&gt;

&lt;ul&gt;
&lt;li&gt;
&lt;code&gt;W&lt;/code&gt; = original full-precision weight tensor&lt;/li&gt;
&lt;li&gt;
&lt;code&gt;b&lt;/code&gt; = target bit-width (e.g., 8 for INT8)&lt;/li&gt;
&lt;li&gt;
&lt;code&gt;max(abs(W))&lt;/code&gt; = the largest absolute value in the tensor — this is the "MaxAbs" calibration you may have seen referenced
Simple, fast, and cheap to compute — but wastes range if the distribution is skewed (e.g., all-positive activations after a ReLU).&lt;/li&gt;
&lt;/ul&gt;

&lt;p&gt;&lt;a href="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Fnx4qa176td36166nvsda.png" class="article-body-image-wrapper"&gt;&lt;img src="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Fnx4qa176td36166nvsda.png" alt=" " width="542" height="534"&gt;&lt;/a&gt;&lt;/p&gt;

&lt;h3&gt;
  
  
  Asymmetric Quantization
&lt;/h3&gt;

&lt;p&gt;Used when the distribution is skewed and doesn't center around zero (common for post-activation values). It introduces a &lt;strong&gt;zero-point&lt;/strong&gt; — an offset that shifts the mapped range to better fit the actual data.&lt;br&gt;
&lt;/p&gt;

&lt;div class="highlight js-code-highlight"&gt;
&lt;pre class="highlight plaintext"&gt;&lt;code&gt;scale      = (max(W) - min(W)) / (2^b - 1)
zero_point = round(-min(W) / scale)

W_quant   = round(W / scale) + zero_point
W_dequant = (W_quant - zero_point) × scale
&lt;/code&gt;&lt;/pre&gt;

&lt;/div&gt;



&lt;p&gt;Asymmetric quantization better utilizes the available integer range for skewed distributions, at the cost of a slightly more expensive computation (the extra zero-point term).&lt;/p&gt;

&lt;p&gt;&lt;a href="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Fq2ygyjd2qrvxwlbpwdxi.png" class="article-body-image-wrapper"&gt;&lt;img src="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Fq2ygyjd2qrvxwlbpwdxi.png" alt=" " width="533" height="506"&gt;&lt;/a&gt;&lt;/p&gt;




&lt;h2&gt;
  
  
  How Accuracy Is Preserved
&lt;/h2&gt;

&lt;p&gt;To maintain the model's accuracy despite the loss of numerical precision, several techniques are used:&lt;/p&gt;

&lt;h3&gt;
  
  
  Quantization-Aware Training (QAT)
&lt;/h3&gt;

&lt;p&gt;QAT simulates quantization &lt;em&gt;during&lt;/em&gt; training so the model learns to be robust to the precision loss it will face later.&lt;/p&gt;

&lt;ol&gt;
&lt;li&gt;
&lt;strong&gt;Fake quantization nodes&lt;/strong&gt; are inserted into the forward pass. These take the real FP32 weight, round it to what it &lt;em&gt;would&lt;/em&gt; look like in low precision, then immediately cast it back to FP32. The weight is still stored in FP32 — only the forward computation "feels" the rounding error.&lt;/li&gt;
&lt;li&gt;This injected error flows into the loss function, same as any other forward-pass computation.&lt;/li&gt;
&lt;li&gt;The catch: rounding has a derivative of zero almost everywhere, so gradients can't flow through it normally. QAT solves this with the &lt;strong&gt;Straight-Through Estimator (STE)&lt;/strong&gt; — it treats the rounding step as if it were the identity function during backpropagation, letting gradients pass through unchanged.&lt;/li&gt;
&lt;li&gt;Over many steps, the optimizer nudges the real weights toward values where quantization rounding hurts the loss the least.
Only after training finishes are the weights actually cast down and stored permanently in low precision.&lt;/li&gt;
&lt;/ol&gt;

&lt;p&gt;&lt;a href="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Ftlkx3zzzt6dsfr9g2va5.png" class="article-body-image-wrapper"&gt;&lt;img src="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Ftlkx3zzzt6dsfr9g2va5.png" alt=" " width="624" height="144"&gt;&lt;/a&gt;&lt;/p&gt;

&lt;h3&gt;
  
  
  Post-Training Quantization (PTQ) with Calibration
&lt;/h3&gt;

&lt;p&gt;PTQ works on an already-trained model by using a small &lt;strong&gt;calibration dataset&lt;/strong&gt;.&lt;/p&gt;

&lt;ol&gt;
&lt;li&gt;Take a small, representative sample of real inputs (for an LLM: a few hundred text sequences from a general or domain-specific corpus).&lt;/li&gt;
&lt;li&gt;Run them through the full-precision model in a normal forward pass (no backpropogation).&lt;/li&gt;
&lt;li&gt;At each layer, record the actual distribution of values seen — weights are fixed, but activation ranges can only be known from real data.&lt;/li&gt;
&lt;li&gt;Use those observed ranges to compute the scale factor (and zero-point, if asymmetric) for that layer.
Representative calibration data matters: if it doesn't resemble real deployment traffic, the resulting scale factors will be miscalibrated — clipping outliers too aggressively, or wasting precision on values that never actually occur.&lt;/li&gt;
&lt;/ol&gt;

&lt;p&gt;Two well-known PTQ methods worth naming:&lt;/p&gt;

&lt;ul&gt;
&lt;li&gt;
&lt;strong&gt;GPTQ&lt;/strong&gt; — uses calibration data to compute layer-wise Hessian (second-order sensitivity) information and solves for quantized weights that minimize reconstruction error.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;AWQ&lt;/strong&gt; — protects "salient" weight channels identified by activation magnitude, rather than treating all weights equally.&lt;/li&gt;
&lt;/ul&gt;

&lt;h3&gt;
  
  
  Hybrid / Mixed-Precision Quantization
&lt;/h3&gt;

&lt;p&gt;Not every layer is equally sensitive to precision loss. The hybrid approach quantizes different layers to different bit-widths based on measured sensitivity:&lt;/p&gt;

&lt;ul&gt;
&lt;li&gt;Less sensitive layers → compressed aggressively (e.g., INT4)&lt;/li&gt;
&lt;li&gt;Critical layers → kept at higher precision (e.g., FP16 or INT8) to preserve overall performance
Sensitivity is often measured using Hessian trace or simple accuracy/perplexity sweeps per layer.&lt;/li&gt;
&lt;/ul&gt;




&lt;h2&gt;
  
  
  Does the Model Go Back to Its Original Form at Inference?
&lt;/h2&gt;

&lt;p&gt;Not exactly, but there's often a dequantization step involved, and which path is used depends on the hardware:&lt;/p&gt;

&lt;p&gt;&lt;strong&gt;Dequantize-then-compute:&lt;/strong&gt; weights sit in memory as INT4/INT8 (saving storage and memory bandwidth), but are upcast back to FP16/BF16 right before the matrix multiply. This saves memory, but the actual math still runs at higher precision — so latency gains are limited.&lt;/p&gt;

&lt;p&gt;&lt;strong&gt;Native low-precision compute:&lt;/strong&gt; on hardware with dedicated support (NVIDIA Tensor Cores — Hopper for FP8, Blackwell for NVFP4), the multiply-accumulate operation happens &lt;em&gt;directly&lt;/em&gt; in low precision, no upcast required. This is where both memory savings &lt;strong&gt;and&lt;/strong&gt; latency/throughput gains show up, because the silicon itself is doing less work per operation.&lt;/p&gt;

&lt;p&gt;So: the stored weights never "revert" — they remain permanently in the lower-precision format. What differs is whether the compute step temporarily upcasts for the math, or the hardware crunches the low-precision numbers directly. Which path gets used depends on the deployment hardware and inference engine (e.g., TensorRT-LLM, vLLM).&lt;/p&gt;




&lt;h2&gt;
  
  
  Wrapping Up
&lt;/h2&gt;

&lt;p&gt;Quantization isn't one trick — it's a toolbox: pick a format based on range vs precision needs, pick symmetric or asymmetric based on your data's distribution, and pick QAT, PTQ, or a hybrid approach based on how much retraining budget you have. The right combination lets you deploy models at a fraction of the memory and latency cost, with accuracy loss that's often barely measurable.&lt;/p&gt;

&lt;p&gt;&lt;em&gt;This post is part of my ongoing series while I learn ML deployment concepts by writing about them. Feedback and corrections welcome!&lt;/em&gt;&lt;/p&gt;

</description>
      <category>machinelearning</category>
      <category>quantizaion</category>
      <category>ai</category>
      <category>llm</category>
    </item>
    <item>
      <title>From Fragment Shaders to Tensor Cores: How CUDA Actually Works ?</title>
      <dc:creator>Sujal Suyash</dc:creator>
      <pubDate>Wed, 29 Jul 2026 17:45:55 +0000</pubDate>
      <link>https://dev.to/suzuskiee/from-fragment-shaders-to-tensor-cores-how-cuda-actually-works--b3k</link>
      <guid>https://dev.to/suzuskiee/from-fragment-shaders-to-tensor-cores-how-cuda-actually-works--b3k</guid>
      <description>&lt;p&gt;Your system has a CPU. It probably also has a GPU. They look similar on a spec sheet - both have "cores," both run at some clock speed, but they were built to solve completely different problems. Understanding &lt;em&gt;why&lt;/em&gt; is the key that unlocks everything else about CUDA.&lt;/p&gt;

&lt;h2&gt;
  
  
  Why does this hardware even exist?
&lt;/h2&gt;

&lt;p&gt;A &lt;strong&gt;CPU&lt;/strong&gt; is latency-optimized. Its job is to finish one complex task as fast as possible, so a large fraction of its transistor budget goes toward things that have nothing to do with raw arithmetic: branch predictors that guess which way an &lt;code&gt;if&lt;/code&gt; statement will go before it's even evaluated, out-of-order execution engines that reorder instructions to avoid stalls, and large multi-level caches (L1, L2, sometimes L3) that keep frequently used data close to the core. A CPU is built to handle unpredictable, branching, sequential logic.&lt;/p&gt;

&lt;p&gt;A &lt;strong&gt;GPU&lt;/strong&gt; is totally opposite of that of CPU. It assumes the workload is predictable, regular, and mathematically dense. The same instruction applied to enormous amounts of data, like every pixel on a screen or every element in a large matrix. Because of that assumption, a GPU strips away most of the branch prediction and out-of-order machinery a CPU relies on, and instead spends its transistor budget on raw arithmetic units -&amp;gt; thousands of small, simple cores instead of a handful of large, complex ones.&lt;/p&gt;

&lt;p&gt;&lt;a href="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Fa08qkdub03zyyu7p9vcc.png" class="article-body-image-wrapper"&gt;&lt;img src="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Fa08qkdub03zyyu7p9vcc.png" alt=" " width="800" height="896"&gt;&lt;/a&gt;&lt;/p&gt;

&lt;p&gt;&lt;strong&gt;CPUs are optimized to finish one task fast. GPUs are optimized to finish millions of similar tasks at once.&lt;/strong&gt;&lt;/p&gt;

&lt;p&gt;Machine learning workloads like matrix multiplications, convolutions, attention mechanisms are, at the arithmetic level, the same handful of operations repeated across enormous tensors. That's exactly the kind of regular, parallel workload GPUs were built for, which is why they became the default hardware for training and running Neural Networks.&lt;/p&gt;

&lt;h2&gt;
  
  
  Before CUDA: computing by pretending to render graphics
&lt;/h2&gt;

&lt;p&gt;It's easy to forget that GPUs weren't originally designed to run arbitrary programs at all. Before 2007, a GPU's entire reason for existing was the graphics pipeline: take a scene made of triangles, run a &lt;strong&gt;vertex shader&lt;/strong&gt; on each vertex to figure out where it lands on screen, then run a &lt;strong&gt;fragment (pixel) shader&lt;/strong&gt; on each resulting pixel to decide its final colour, and write that colour to a framebuffer.&lt;/p&gt;

&lt;p&gt;Researchers noticed something important that a fragment shader is really just a small program that runs once per pixel, in parallel, across the whole screen. If you could trick the GPU into treating your &lt;em&gt;data&lt;/em&gt; as if it were a &lt;em&gt;picture&lt;/em&gt;, you could get that same massive parallelism applied to ordinary numerical computation. This workaround became known as &lt;strong&gt;GPGPU&lt;/strong&gt; —&amp;gt; &lt;em&gt;General Purpose computing on Graphics Processing Units&lt;/em&gt; and it required contorting real math problems into a form the graphics pipeline would accept:&lt;/p&gt;

&lt;ul&gt;
&lt;li&gt;Input data (say, two matrices you wanted to multiply) had to be encoded as &lt;strong&gt;textures&lt;/strong&gt;, because textures were the only large, structured input the GPU knew how to read.&lt;/li&gt;
&lt;li&gt;To actually trigger the computation, you had to render a &lt;strong&gt;full-screen quad&lt;/strong&gt;  two triangles covering the entire viewport, a purely as a pretext to force the GPU to invoke your fragment shader once per output element. The "image" being rendered had no visual meaning; it was a disguise to make the hardware execute your math.&lt;/li&gt;
&lt;li&gt;The fragment shader, written in a shading language meant for lighting and colour, had to be repurposed for arithmetic — reading inputs via texture lookups instead of normal memory reads, and outputting a numeric result as if it were a pixel colour.&lt;/li&gt;
&lt;/ul&gt;

&lt;p&gt;A wildly simplified GLSL fragment shader "computing" &lt;code&gt;a[i] + b[i]&lt;/code&gt; by disguising it as pixel colour math looked something like this:&lt;br&gt;
&lt;/p&gt;

&lt;div class="highlight js-code-highlight"&gt;
&lt;pre class="highlight glsl"&gt;&lt;code&gt;&lt;span class="c1"&gt;// Old-style GPGPU: addition disguised as a fragment shader&lt;/span&gt;
&lt;span class="k"&gt;uniform&lt;/span&gt; &lt;span class="kt"&gt;sampler2D&lt;/span&gt; &lt;span class="n"&gt;textureA&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt; &lt;span class="c1"&gt;// matrix A, encoded as a texture&lt;/span&gt;
&lt;span class="k"&gt;uniform&lt;/span&gt; &lt;span class="kt"&gt;sampler2D&lt;/span&gt; &lt;span class="n"&gt;textureB&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt; &lt;span class="c1"&gt;// matrix B, encoded as a texture&lt;/span&gt;

&lt;span class="kt"&gt;void&lt;/span&gt; &lt;span class="nf"&gt;main&lt;/span&gt;&lt;span class="p"&gt;()&lt;/span&gt; &lt;span class="p"&gt;{&lt;/span&gt;
    &lt;span class="kt"&gt;vec4&lt;/span&gt; &lt;span class="n"&gt;a&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;texture2D&lt;/span&gt;&lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="n"&gt;textureA&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="nb"&gt;gl_TexCoord&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="mi"&gt;0&lt;/span&gt;&lt;span class="p"&gt;].&lt;/span&gt;&lt;span class="n"&gt;xy&lt;/span&gt;&lt;span class="p"&gt;);&lt;/span&gt;
    &lt;span class="kt"&gt;vec4&lt;/span&gt; &lt;span class="n"&gt;b&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;texture2D&lt;/span&gt;&lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="n"&gt;textureB&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="nb"&gt;gl_TexCoord&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="mi"&gt;0&lt;/span&gt;&lt;span class="p"&gt;].&lt;/span&gt;&lt;span class="n"&gt;xy&lt;/span&gt;&lt;span class="p"&gt;);&lt;/span&gt;
    &lt;span class="n"&gt;gl_FragColour&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;a&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;b&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt; &lt;span class="c1"&gt;// the "pixel colour" is actually your result&lt;/span&gt;
&lt;span class="p"&gt;}&lt;/span&gt;
&lt;/code&gt;&lt;/pre&gt;

&lt;/div&gt;



&lt;p&gt;The result didn't come back as an array, rather it came back as an image sitting in a framebuffer, which then had to be read back out and reinterpreted as numbers. Control flow was extremely limited, loops and conditionals were weak or unsupported in early shader models, and there was no way to write to an arbitrary memory location only to the specific pixel a shader invocation happened to own.&lt;/p&gt;

&lt;p&gt;NVIDIA released &lt;strong&gt;Compute Unified Device Architecture (CUDA)&lt;/strong&gt; in 2007 specifically to remove this barrier, exposing the GPU's parallel execution units through a C-like programming model with real memory reads and writes, real control flow, and no requirement to pretend any of it was a picture. This is where &lt;strong&gt;GPU became a general purpose tool&lt;/strong&gt;.&lt;/p&gt;

&lt;h2&gt;
  
  
  How a GPU organizes its execution
&lt;/h2&gt;

&lt;p&gt;CUDA uses a model called &lt;strong&gt;SIMT&lt;/strong&gt; — Single Instruction, Multiple Threads. One instruction is issued, and many threads execute it simultaneously, each on its own piece of data. This is organized into a hierarchy that maps directly onto the physical hardware:&lt;/p&gt;

&lt;p&gt;&lt;a href="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2F33cj2lfvsw5simh6teyv.png" class="article-body-image-wrapper"&gt;&lt;img src="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2F33cj2lfvsw5simh6teyv.png" alt=" " width="800" height="452"&gt;&lt;/a&gt;&lt;/p&gt;

&lt;ul&gt;
&lt;li&gt;
&lt;strong&gt;Thread&lt;/strong&gt; is the smallest unit of execution, with its own program counter and private registers.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Warp&lt;/strong&gt; a group of exactly &lt;strong&gt;32 threads&lt;/strong&gt;, the actual unit the hardware schedules. All 32 threads in a warp are issued the same instruction on the same clock cycle.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Block&lt;/strong&gt; up to 1024 threads scheduled onto one &lt;strong&gt;Streaming Multiprocessor (SM)&lt;/strong&gt;. Threads in a block can synchronize with &lt;code&gt;__syncthreads()&lt;/code&gt; and share fast on-chip memory.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Grid&lt;/strong&gt; every block needed to run the kernel, spread across all available SMs.&lt;/li&gt;
&lt;/ul&gt;

&lt;p&gt;Here's what a minimal CUDA kernel actually looks like — adding two vectors, with one thread handling one element:&lt;br&gt;
&lt;/p&gt;

&lt;div class="highlight js-code-highlight"&gt;
&lt;pre class="highlight cuda"&gt;&lt;code&gt;&lt;span class="k"&gt;__global__&lt;/span&gt; &lt;span class="kt"&gt;void&lt;/span&gt; &lt;span class="nf"&gt;vectorAdd&lt;/span&gt;&lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="kt"&gt;float&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt;&lt;span class="n"&gt;a&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="kt"&gt;float&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt;&lt;span class="n"&gt;b&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="kt"&gt;float&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt;&lt;span class="n"&gt;c&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="kt"&gt;int&lt;/span&gt; &lt;span class="n"&gt;n&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt; &lt;span class="p"&gt;{&lt;/span&gt;
    &lt;span class="kt"&gt;int&lt;/span&gt; &lt;span class="n"&gt;i&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;blockIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="n"&gt;blockDim&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt;
    &lt;span class="k"&gt;if&lt;/span&gt; &lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt; &lt;span class="o"&gt;&amp;lt;&lt;/span&gt; &lt;span class="n"&gt;n&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt; &lt;span class="p"&gt;{&lt;/span&gt;
        &lt;span class="n"&gt;c&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;a&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;b&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt;&lt;span class="p"&gt;];&lt;/span&gt;
    &lt;span class="p"&gt;}&lt;/span&gt;
&lt;span class="p"&gt;}&lt;/span&gt;
&lt;/code&gt;&lt;/pre&gt;

&lt;/div&gt;



&lt;p&gt;Compare that to the GLSL disguise above — real array indexing, real addition, no textures, no framebuffers.&lt;/p&gt;

&lt;h3&gt;
  
  
  Occupancy and latency hiding
&lt;/h3&gt;

&lt;p&gt;An SM typically holds far more warps resident at once than it can execute in a single cycle. When a warp stalls, most often waiting on a slow global memory reads the SM's scheduler switches to a different warp that's ready to go, and the stalled warp resumes once its data arrives. This is called &lt;strong&gt;latency hiding&lt;/strong&gt;, and it's the main reason GPUs don't need CPU-style caches and speculative execution to stay busy. The fraction of an SM's maximum warp capacity actually kept resident is called &lt;strong&gt;occupancy&lt;/strong&gt;.&lt;/p&gt;

&lt;h3&gt;
  
  
  Warp divergence
&lt;/h3&gt;

&lt;p&gt;Because all 32 threads in a warp share a single instruction fetch, they must step through the code together:&lt;br&gt;
&lt;/p&gt;

&lt;div class="highlight js-code-highlight"&gt;
&lt;pre class="highlight cuda"&gt;&lt;code&gt;&lt;span class="k"&gt;if&lt;/span&gt; &lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="n"&gt;data&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;&amp;gt;&lt;/span&gt; &lt;span class="n"&gt;threshold&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt; &lt;span class="p"&gt;{&lt;/span&gt;
    &lt;span class="c1"&gt;// path A&lt;/span&gt;
&lt;span class="p"&gt;}&lt;/span&gt; &lt;span class="k"&gt;else&lt;/span&gt; &lt;span class="p"&gt;{&lt;/span&gt;
    &lt;span class="c1"&gt;// path B&lt;/span&gt;
&lt;span class="p"&gt;}&lt;/span&gt;
&lt;/code&gt;&lt;/pre&gt;

&lt;/div&gt;



&lt;p&gt;If threads within the same warp evaluate this condition differently, the warp &lt;strong&gt;diverges&lt;/strong&gt;. The hardware masks off the threads taking path B, executes path A for the active threads, flips the mask, executes path B, and reconverges once both paths finish. Work that could have run in parallel now runs sequentially . A simple 2 way branch can roughly half effective throughput for that warp.&lt;/p&gt;

&lt;p&gt;&lt;strong&gt;Takeaway:&lt;/strong&gt; uniform, branch-free workloads run efficiently on a GPU. Conditional, data-dependent logic that splits a warp's behaviour is where you lose parallel performance.&lt;/p&gt;

&lt;h2&gt;
  
  
  The real bottleneck: memory, not compute
&lt;/h2&gt;

&lt;p&gt;A modern GPU can typically perform far more arithmetic operations per second than it can pull bytes from memory per second. In most real workloads, memory bandwidth is the limiting factor. This is formalized in the &lt;strong&gt;roofline model&lt;/strong&gt;: a kernel's achievable performance is capped either by peak compute throughput or by peak memory bandwidth, whichever constraint that kernel's ratio of math-to-bytes hits first.&lt;/p&gt;

&lt;p&gt;&lt;a href="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Ffkibgi6fp0gja22v05ql.png" class="article-body-image-wrapper"&gt;&lt;img src="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Ffkibgi6fp0gja22v05ql.png" alt=" " width="800" height="1523"&gt;&lt;/a&gt;&lt;/p&gt;

&lt;p&gt;Efficient CUDA code minimizes round trips to global memory. A common pattern used heavily in matrix multiplication is &lt;strong&gt;tiling&lt;/strong&gt;. Load a chunk of input data from global memory into shared memory once, let every thread in the block reuse that tile for all its arithmetic, and only write the final result back to global memory at the end.&lt;br&gt;
&lt;/p&gt;

&lt;div class="highlight js-code-highlight"&gt;
&lt;pre class="highlight cuda"&gt;&lt;code&gt;&lt;span class="k"&gt;__global__&lt;/span&gt; &lt;span class="kt"&gt;void&lt;/span&gt; &lt;span class="nf"&gt;matMulTiled&lt;/span&gt;&lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="kt"&gt;float&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt;&lt;span class="n"&gt;A&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="kt"&gt;float&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt;&lt;span class="n"&gt;B&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="kt"&gt;float&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt;&lt;span class="n"&gt;C&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="kt"&gt;int&lt;/span&gt; &lt;span class="n"&gt;n&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt; &lt;span class="p"&gt;{&lt;/span&gt;
    &lt;span class="k"&gt;__shared__&lt;/span&gt; &lt;span class="kt"&gt;float&lt;/span&gt; &lt;span class="n"&gt;tileA&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="mi"&gt;16&lt;/span&gt;&lt;span class="p"&gt;][&lt;/span&gt;&lt;span class="mi"&gt;16&lt;/span&gt;&lt;span class="p"&gt;];&lt;/span&gt;
    &lt;span class="k"&gt;__shared__&lt;/span&gt; &lt;span class="kt"&gt;float&lt;/span&gt; &lt;span class="n"&gt;tileB&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="mi"&gt;16&lt;/span&gt;&lt;span class="p"&gt;][&lt;/span&gt;&lt;span class="mi"&gt;16&lt;/span&gt;&lt;span class="p"&gt;];&lt;/span&gt;

    &lt;span class="kt"&gt;int&lt;/span&gt; &lt;span class="n"&gt;row&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;blockIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;y&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="mi"&gt;16&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;y&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt;
    &lt;span class="kt"&gt;int&lt;/span&gt; &lt;span class="n"&gt;col&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;blockIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="mi"&gt;16&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt;
    &lt;span class="kt"&gt;float&lt;/span&gt; &lt;span class="n"&gt;sum&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="mf"&gt;0.0&lt;/span&gt;&lt;span class="n"&gt;f&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt;

    &lt;span class="k"&gt;for&lt;/span&gt; &lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="kt"&gt;int&lt;/span&gt; &lt;span class="n"&gt;t&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="mi"&gt;0&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt; &lt;span class="n"&gt;t&lt;/span&gt; &lt;span class="o"&gt;&amp;lt;&lt;/span&gt; &lt;span class="n"&gt;n&lt;/span&gt; &lt;span class="o"&gt;/&lt;/span&gt; &lt;span class="mi"&gt;16&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt; &lt;span class="n"&gt;t&lt;/span&gt;&lt;span class="o"&gt;++&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt; &lt;span class="p"&gt;{&lt;/span&gt;
        &lt;span class="n"&gt;tileA&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;y&lt;/span&gt;&lt;span class="p"&gt;][&lt;/span&gt;&lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;A&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;row&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="n"&gt;n&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;t&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="mi"&gt;16&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;];&lt;/span&gt;
        &lt;span class="n"&gt;tileB&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;y&lt;/span&gt;&lt;span class="p"&gt;][&lt;/span&gt;&lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;B&lt;/span&gt;&lt;span class="p"&gt;[(&lt;/span&gt;&lt;span class="n"&gt;t&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="mi"&gt;16&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;y&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="n"&gt;n&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;col&lt;/span&gt;&lt;span class="p"&gt;];&lt;/span&gt;
        &lt;span class="n"&gt;__syncthreads&lt;/span&gt;&lt;span class="p"&gt;();&lt;/span&gt;

        &lt;span class="k"&gt;for&lt;/span&gt; &lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="kt"&gt;int&lt;/span&gt; &lt;span class="n"&gt;k&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="mi"&gt;0&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt; &lt;span class="n"&gt;k&lt;/span&gt; &lt;span class="o"&gt;&amp;lt;&lt;/span&gt; &lt;span class="mi"&gt;16&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt; &lt;span class="n"&gt;k&lt;/span&gt;&lt;span class="o"&gt;++&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt;
            &lt;span class="n"&gt;sum&lt;/span&gt; &lt;span class="o"&gt;+=&lt;/span&gt; &lt;span class="n"&gt;tileA&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;y&lt;/span&gt;&lt;span class="p"&gt;][&lt;/span&gt;&lt;span class="n"&gt;k&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="n"&gt;tileB&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;k&lt;/span&gt;&lt;span class="p"&gt;][&lt;/span&gt;&lt;span class="n"&gt;threadIdx&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;];&lt;/span&gt;
        &lt;span class="n"&gt;__syncthreads&lt;/span&gt;&lt;span class="p"&gt;();&lt;/span&gt;
    &lt;span class="p"&gt;}&lt;/span&gt;
    &lt;span class="n"&gt;C&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;row&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="n"&gt;n&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;col&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;sum&lt;/span&gt;&lt;span class="p"&gt;;&lt;/span&gt;
&lt;span class="p"&gt;}&lt;/span&gt;
&lt;/code&gt;&lt;/pre&gt;

&lt;/div&gt;



&lt;p&gt;Each value is fetched from global memory once per tile and reused 16 times from shared memory instead of being re-fetched for every multiplication. This is often the single biggest optimization available for compute-heavy kernels like this one.&lt;/p&gt;

&lt;h3&gt;
  
  
  Memory coalescing
&lt;/h3&gt;

&lt;p&gt;When a warp requests data from global memory, the hardware tries to satisfy that request in as few large, contiguous transactions as possible. This works efficiently if thread 0 requests index 0, thread 1 requests index 1, and so on — &lt;strong&gt;coalesced access&lt;/strong&gt;, where one wide transaction can satisfy the entire warp.&lt;br&gt;
&lt;/p&gt;

&lt;div class="highlight js-code-highlight"&gt;
&lt;pre class="highlight cuda"&gt;&lt;code&gt;&lt;span class="c1"&gt;// Coalesced: consecutive threads read consecutive memory&lt;/span&gt;
&lt;span class="n"&gt;c&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;a&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;b&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt;&lt;span class="p"&gt;];&lt;/span&gt;              &lt;span class="c1"&gt;// i = threadIdx.x + blockIdx.x * blockDim.x&lt;/span&gt;

&lt;span class="c1"&gt;// Uncoalesced: large stride scatters each thread's access far apart&lt;/span&gt;
&lt;span class="n"&gt;c&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="n"&gt;a&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="n"&gt;stride&lt;/span&gt;&lt;span class="p"&gt;]&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;b&lt;/span&gt;&lt;span class="p"&gt;[&lt;/span&gt;&lt;span class="n"&gt;i&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="n"&gt;stride&lt;/span&gt;&lt;span class="p"&gt;];&lt;/span&gt;
&lt;/code&gt;&lt;/pre&gt;

&lt;/div&gt;



&lt;p&gt;If the data is scattered like - a linked list, a fragmented hashmap, or simply a large stride, the hardware can't fetch it in one transaction and instead issues a separate transaction per thread (or small group). This &lt;strong&gt;uncoalesced access&lt;/strong&gt; can cut effective memory bandwidth by an order of magnitude, even though the computation itself hasn't changed — only the layout of the data being read.&lt;/p&gt;

&lt;p&gt;&lt;strong&gt;Takeaway:&lt;/strong&gt; the layout of your data in memory can matter as much as the algorithm itself.&lt;/p&gt;

&lt;h2&gt;
  
  
  From CUDA cores to Tensor Cores
&lt;/h2&gt;

&lt;p&gt;Everything above involves a standard &lt;strong&gt;CUDA core&lt;/strong&gt; which is a simple ALU capable of one scalar operation per clock cycle, per thread. That's already effective for general-purpose parallel work. But deep learning workloads are dominated by one specific operation: matrix multiplication. A fully connected layer is a matrix multiply; a convolution can be reformulated as one; the core operation inside transformer attention is, again, matrix multiplication.&lt;/p&gt;

&lt;p&gt;To accelerate this specifically, NVIDIA introduced the &lt;strong&gt;Tensor Core&lt;/strong&gt; with the Volta architecture in 2017.&lt;/p&gt;

&lt;p&gt;&lt;a href="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Fc3nb461tds7jdwyxuo2h.png" class="article-body-image-wrapper"&gt;&lt;img src="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fdev-to-uploads.s3.us-east-2.amazonaws.com%2Fuploads%2Farticles%2Fc3nb461tds7jdwyxuo2h.png" alt=" " width="800" height="1525"&gt;&lt;/a&gt;&lt;/p&gt;

&lt;p&gt;Where a standard CUDA core computes one scalar operation per cycle, a Tensor Core computes a full mixed-precision multiply-accumulate — &lt;strong&gt;D = A × B + C&lt;/strong&gt; — on small matrix tiles (originally 4×4) in a single clock cycle. Later architectures expanded this substantially: larger effective tile sizes, and support for more numeric precisions — FP16, BF16, TF32, INT8, and structured sparsity, which skips known-zero values to save additional cycles. "Mixed precision" here means the multiplication happens in a lower-precision format for speed, while accumulation happens in a higher-precision format to preserve accuracy.&lt;/p&gt;

&lt;p&gt;This is the central hardware reason GPUs shifted from being primarily graphics chips that happened to be useful for AI, to being AI accelerators that also happen to still render graphics.&lt;/p&gt;

&lt;h2&gt;
  
  
  Putting it together
&lt;/h2&gt;

&lt;ol&gt;
&lt;li&gt;
&lt;strong&gt;GPUs trade single-task speed for massive parallel throughput&lt;/strong&gt; — well suited to the repetitive, regular math in ML workloads.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Before CUDA, that parallelism could only be reached by disguising computation as graphics rendering&lt;/strong&gt; — data encoded as textures, math run inside fragment shaders, results read back out of a framebuffer.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Threads execute in warps of 32, in lockstep, and SMs hide memory latency by switching between resident warps (occupancy)&lt;/strong&gt; — branching that splits a warp's execution path (divergence) reduces performance.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Memory bandwidth, not raw compute, is usually the actual bottleneck&lt;/strong&gt; — structure data so neighbouring threads access neighbouring memory (coalescing), and reuse shared memory via tiling instead of repeatedly hitting global memory.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Tensor Cores exist specifically to accelerate the matrix multiplications neural networks are built from&lt;/strong&gt; — a core hardware reason large-scale AI training is feasible at all.&lt;/li&gt;
&lt;/ol&gt;

&lt;p&gt;Most CUDA optimization techniques you'll encounter fall into one of these five ideas: tiling, shared memory caching, avoiding divergent branches, maximizing occupancy, and choosing the right numeric precision. It's a long way from disguising math as pixel colours to letting a single instruction multiply entire matrix tiles. But the underlying goal never changed: keep thousands of simple cores fed, and keep them all doing the same thing at once.&lt;/p&gt;

</description>
      <category>nvidia</category>
      <category>cuda</category>
      <category>ai</category>
      <category>machinelearning</category>
    </item>
    <item>
      <title>Flowing vs. Thinking: How Liquid Neural Networks Diverge from LLMs</title>
      <dc:creator>Sujal Suyash</dc:creator>
      <pubDate>Mon, 20 Jul 2026 16:41:29 +0000</pubDate>
      <link>https://dev.to/suzuskiee/flowing-vs-thinking-how-liquid-neural-networks-diverge-from-llms-4kal</link>
      <guid>https://dev.to/suzuskiee/flowing-vs-thinking-how-liquid-neural-networks-diverge-from-llms-4kal</guid>
      <description>&lt;p&gt;If you follow the world of Artificial Intelligence, it is easy to assume that scaling up is the only path forward. &lt;strong&gt;Large Language Models (LLMs)&lt;/strong&gt; have dominated the conversation by scaling to hundreds of billions of parameters, acting as massive, discrete reasoning engines.&lt;/p&gt;

&lt;p&gt;But not all problems require a massive library of tokens. In the world of robotics, physical sensors, and continuous time, a different architectural philosophy is thriving: &lt;strong&gt;Liquid Neural Networks (LNNs)&lt;/strong&gt;.&lt;/p&gt;

&lt;p&gt;LLMs process static, discrete symbols, LNNs use highly expressive fluid equations to navigate the chaotic, continuous flow of the physical world. To understand why an LNN can solve complex physics tasks with just a fraction of the compute of an LLM, we have to look at the actual mathematics.&lt;/p&gt;

&lt;h2&gt;
  
  
  The Common Misconception About "Liquid" Adaptation
&lt;/h2&gt;

&lt;p&gt;People often state that LNNs "adapt" according to the input they receive. In the context of LLMs, we usually think of adaptation as shifting attention weights across discrete tokens. When people hear that LNNs adapt dynamically, a common misconception arises: &lt;em&gt;Do LNNs actually change their weights or biases during the forward pass?&lt;/em&gt;&lt;/p&gt;

&lt;p&gt;&lt;strong&gt;The short answer is no.&lt;/strong&gt;&lt;/p&gt;

&lt;p&gt;Just like an LLM, an LNN's parameters — its input weight matrix (

&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;U&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
), recurrent weight matrix (
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;W&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
), and biases (
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;b&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
) — are completely fixed after the training phase is complete.&lt;/p&gt;

&lt;p&gt;So, what makes them "liquid"?&lt;/p&gt;

&lt;p&gt;The adaptation happens within the &lt;strong&gt;hidden state&lt;/strong&gt; (
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;h&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
) and the &lt;strong&gt;effective time constant&lt;/strong&gt; (
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;τ&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
) of each individual neuron, which change continuously with time. Unlike LLMs, which are governed by the discrete-time architecture of the Transformer, LNNs are fundamentally built on &lt;strong&gt;Ordinary Differential Equations (ODEs)&lt;/strong&gt;.&lt;/p&gt;

&lt;p&gt;Inside an LNN, the equation describing a neuron's state looks like this:&lt;/p&gt;


&lt;div class="katex-element"&gt;
  &lt;span class="katex-display"&gt;&lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord"&gt;&lt;span class="mtable"&gt;&lt;span class="col-align-r"&gt;&lt;span class="vlist-t vlist-t2"&gt;&lt;span class="vlist-r"&gt;&lt;span class="vlist"&gt;&lt;span&gt;&lt;span class="pstrut"&gt;&lt;/span&gt;&lt;span class="mord"&gt;&lt;span class="mord"&gt;&lt;span class="mopen nulldelimiter"&gt;&lt;/span&gt;&lt;span class="mfrac"&gt;&lt;span class="vlist-t vlist-t2"&gt;&lt;span class="vlist-r"&gt;&lt;span class="vlist"&gt;&lt;span&gt;&lt;span class="pstrut"&gt;&lt;/span&gt;&lt;span class="mord"&gt;&lt;span class="mord mathnormal"&gt;d&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;span&gt;&lt;span class="pstrut"&gt;&lt;/span&gt;&lt;span class="frac-line"&gt;&lt;/span&gt;&lt;/span&gt;&lt;span&gt;&lt;span class="pstrut"&gt;&lt;/span&gt;&lt;span class="mord"&gt;&lt;span class="mord mathnormal"&gt;d&lt;/span&gt;&lt;span class="mord mathnormal"&gt;h&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;span class="vlist-s"&gt;​&lt;/span&gt;&lt;/span&gt;&lt;span class="vlist-r"&gt;&lt;span class="vlist"&gt;&lt;span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;span class="mclose nulldelimiter"&gt;&lt;/span&gt;&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mrel"&gt;=&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mord"&gt;−&lt;/span&gt;&lt;span class="mord"&gt;&lt;span class="mopen nulldelimiter"&gt;&lt;/span&gt;&lt;span class="mfrac"&gt;&lt;span class="vlist-t vlist-t2"&gt;&lt;span class="vlist-r"&gt;&lt;span class="vlist"&gt;&lt;span&gt;&lt;span class="pstrut"&gt;&lt;/span&gt;&lt;span class="mord"&gt;&lt;span class="mord mathnormal"&gt;τ&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;x&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;span class="mpunct"&gt;,&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;h&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;))&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;span&gt;&lt;span class="pstrut"&gt;&lt;/span&gt;&lt;span class="frac-line"&gt;&lt;/span&gt;&lt;/span&gt;&lt;span&gt;&lt;span class="pstrut"&gt;&lt;/span&gt;&lt;span class="mord"&gt;&lt;span class="mord mathnormal"&gt;h&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;span class="vlist-s"&gt;​&lt;/span&gt;&lt;/span&gt;&lt;span class="vlist-r"&gt;&lt;span class="vlist"&gt;&lt;span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;span class="mclose nulldelimiter"&gt;&lt;/span&gt;&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mbin"&gt;+&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;f&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;W&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mbin"&gt;⋅&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;h&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mbin"&gt;+&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;U&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mbin"&gt;⋅&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;x&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mbin"&gt;+&lt;/span&gt;&lt;span class="mspace"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;b&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;span class="vlist-s"&gt;​&lt;/span&gt;&lt;/span&gt;&lt;span class="vlist-r"&gt;&lt;span class="vlist"&gt;&lt;span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/div&gt;


&lt;p&gt;Where 
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;f&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
 is a standard non-linear activation function (like 
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mop"&gt;tanh&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
 or sigmoid).&lt;/p&gt;

&lt;p&gt;Notice the time constant 
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;τ&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
. In a traditional continuous-time neural network, 
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;τ&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
 is a fixed number. In an LNN, &lt;strong&gt;
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;τ&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
 is a dynamic function of the current input 
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;x&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
 and the current hidden state 
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;h&lt;/span&gt;&lt;span class="mopen"&gt;(&lt;/span&gt;&lt;span class="mord mathnormal"&gt;t&lt;/span&gt;&lt;span class="mclose"&gt;)&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
.&lt;/strong&gt;&lt;/p&gt;

&lt;p&gt;In code, the right-hand side of that ODE is just this:&lt;br&gt;
&lt;/p&gt;

&lt;div class="highlight js-code-highlight"&gt;
&lt;pre class="highlight python"&gt;&lt;code&gt;&lt;span class="kn"&gt;import&lt;/span&gt; &lt;span class="n"&gt;numpy&lt;/span&gt; &lt;span class="k"&gt;as&lt;/span&gt; &lt;span class="n"&gt;np&lt;/span&gt;

&lt;span class="k"&gt;def&lt;/span&gt; &lt;span class="nf"&gt;dh_dt&lt;/span&gt;&lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="n"&gt;h&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;W&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;U&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;b&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;tau_fn&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;f&lt;/span&gt;&lt;span class="o"&gt;=&lt;/span&gt;&lt;span class="n"&gt;np&lt;/span&gt;&lt;span class="p"&gt;.&lt;/span&gt;&lt;span class="n"&gt;tanh&lt;/span&gt;&lt;span class="p"&gt;):&lt;/span&gt;
    &lt;span class="sh"&gt;"""&lt;/span&gt;&lt;span class="s"&gt;
    LNN ODE: dh/dt = -h / tau(x, h) + f(W·h + U·x + b)
    tau_fn(x, h) -&amp;gt; the dynamic time constant function.
    &lt;/span&gt;&lt;span class="sh"&gt;"""&lt;/span&gt;
    &lt;span class="c1"&gt;# 1. Compute the dynamic, liquid time constant
&lt;/span&gt;    &lt;span class="n"&gt;tau&lt;/span&gt; &lt;span class="o"&gt;=&lt;/span&gt; &lt;span class="nf"&gt;tau_fn&lt;/span&gt;&lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;h&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt;

    &lt;span class="c1"&gt;# 2. Compute the continuous rate of change
&lt;/span&gt;    &lt;span class="k"&gt;return&lt;/span&gt; &lt;span class="o"&gt;-&lt;/span&gt;&lt;span class="n"&gt;h&lt;/span&gt; &lt;span class="o"&gt;/&lt;/span&gt; &lt;span class="n"&gt;tau&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="nf"&gt;f&lt;/span&gt;&lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="n"&gt;W&lt;/span&gt; &lt;span class="o"&gt;@&lt;/span&gt; &lt;span class="n"&gt;h&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;U&lt;/span&gt; &lt;span class="o"&gt;@&lt;/span&gt; &lt;span class="n"&gt;x&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;b&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt;

&lt;span class="c1"&gt;# To update the hidden state over a discrete time step (dt):
&lt;/span&gt;&lt;span class="k"&gt;def&lt;/span&gt; &lt;span class="nf"&gt;euler_step&lt;/span&gt;&lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="n"&gt;h&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;W&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;U&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;b&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;tau_fn&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;dt&lt;/span&gt;&lt;span class="p"&gt;):&lt;/span&gt;
    &lt;span class="k"&gt;return&lt;/span&gt; &lt;span class="n"&gt;h&lt;/span&gt; &lt;span class="o"&gt;+&lt;/span&gt; &lt;span class="n"&gt;dt&lt;/span&gt; &lt;span class="o"&gt;*&lt;/span&gt; &lt;span class="nf"&gt;dh_dt&lt;/span&gt;&lt;span class="p"&gt;(&lt;/span&gt;&lt;span class="n"&gt;h&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;x&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;W&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;U&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;b&lt;/span&gt;&lt;span class="p"&gt;,&lt;/span&gt; &lt;span class="n"&gt;tau_fn&lt;/span&gt;&lt;span class="p"&gt;)&lt;/span&gt;
&lt;/code&gt;&lt;/pre&gt;

&lt;/div&gt;



&lt;p&gt;Therefore, "adaptation" in an LNN doesn't mean the network is rewriting its own code (altering weights) on the fly. Instead, it means &lt;strong&gt;the speed and trajectory of the neuron's state change based entirely on the input.&lt;/strong&gt; If the incoming data suddenly becomes noisy, erratic, or unpredictable, the neuron can mathematically "slow down" its integration to filter out the noise — all without needing new weights.&lt;/p&gt;

&lt;h2&gt;
  
  
  The Architecture of the Continuous World: Where LNNs Excel
&lt;/h2&gt;

&lt;p&gt;Because of this unique, continuous-time framework, LNNs are uniquely suited for &lt;strong&gt;physical, real-world, streaming environments&lt;/strong&gt;.&lt;/p&gt;

&lt;ul&gt;
&lt;li&gt;
&lt;strong&gt;Ultra-Low Parameter Footprint &amp;amp; Edge Deployment:&lt;/strong&gt; While an LLM needs massive server farms, LNNs achieve staggering complexity with very few parameters. Because they require orders of magnitude less memory and power, LNNs can run directly on edge devices, microcontrollers, and drones.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Handling Noisy, Irregular Time-Series Data:&lt;/strong&gt; Transformers process time discreetly, expecting data in neat, sequential chunks. If you are analyzing continuous motion or gait datasets to detect subtle biomechanical anomalies, LLMs struggle. LNNs thrive here because their underlying ODEs treat time as a continuous flow.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Out-of-Distribution Survival:&lt;/strong&gt; When an LLM encounters a scenario completely foreign to its training data, it hallucinates. If a drone powered by an LNN is trained in clear skies but suddenly encounters a heavy rainstorm, its liquid time constants automatically adjust to compensate for the erratic sensor noise, allowing it to maintain control.&lt;/li&gt;
&lt;/ul&gt;

&lt;h2&gt;
  
  
  The Architecture of the Discrete World: Where LLMs Excel
&lt;/h2&gt;

&lt;p&gt;Despite the elegance of LNNs, they are not designed to process semantic knowledge. LLMs remain the undisputed engines of &lt;strong&gt;reasoning and massive knowledge retrieval&lt;/strong&gt;.&lt;/p&gt;

&lt;ul&gt;
&lt;li&gt;
&lt;strong&gt;Discrete Logic and Language:&lt;/strong&gt; Language is not a continuous, flowing physical signal; it is composed of discrete symbols. The self-attention mechanism of LLMs allows them to weigh the relationship between every word in a document simultaneously. If you are building a system to parse and classify legal contract clauses, you reach for an LLM. LNNs, as sequential ODE solvers, are fundamentally ill-suited for parsing complex grammar or writing code.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Encyclopedic World Knowledge:&lt;/strong&gt; Because LLMs boast billions of parameters, they double as incredibly vast, compressed databases of human knowledge. An LNN's hyper-efficient, compact footprint means it possesses virtually no capacity to retain world facts.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Long-Range Static Context:&lt;/strong&gt; If you feed a massive codebase into a long-context LLM, it can instantly connect a variable on page 1 with a function on page 99. Because LNNs process data sequentially over time, they are susceptible to "forgetting" older static information when forced to read massive blocks of text.&lt;/li&gt;
&lt;/ul&gt;

&lt;h2&gt;
  
  
  Summary: Different Engines for Different Realities
&lt;/h2&gt;

&lt;div class="table-wrapper-paragraph"&gt;&lt;table&gt;
&lt;thead&gt;
&lt;tr&gt;
&lt;th&gt;Feature&lt;/th&gt;
&lt;th&gt;Liquid Neural Networks (LNNs)&lt;/th&gt;
&lt;th&gt;Large Language Models (LLMs)&lt;/th&gt;
&lt;/tr&gt;
&lt;/thead&gt;
&lt;tbody&gt;
&lt;tr&gt;
&lt;td&gt;&lt;strong&gt;Primary Domain&lt;/strong&gt;&lt;/td&gt;
&lt;td&gt;Robotics, autonomous systems, continuous sensors&lt;/td&gt;
&lt;td&gt;Text, code, semantic reasoning&lt;/td&gt;
&lt;/tr&gt;
&lt;tr&gt;
&lt;td&gt;&lt;strong&gt;Time Paradigm&lt;/strong&gt;&lt;/td&gt;
&lt;td&gt;Continuous (ODEs)&lt;/td&gt;
&lt;td&gt;Discrete (Tokens + Attention)&lt;/td&gt;
&lt;/tr&gt;
&lt;tr&gt;
&lt;td&gt;&lt;strong&gt;Scale&lt;/strong&gt;&lt;/td&gt;
&lt;td&gt;Extremely small (thousands of parameters)&lt;/td&gt;
&lt;td&gt;Massive (billions of parameters)&lt;/td&gt;
&lt;/tr&gt;
&lt;tr&gt;
&lt;td&gt;&lt;strong&gt;Adaptation Mechanism&lt;/strong&gt;&lt;/td&gt;
&lt;td&gt;Dynamic, input-dependent time constants (
&lt;span class="katex-element"&gt;
  &lt;span class="katex"&gt;&lt;span class="katex-mathml"&gt;&lt;/span&gt;&lt;span class="katex-html"&gt;&lt;span class="base"&gt;&lt;span class="strut"&gt;&lt;/span&gt;&lt;span class="mord mathnormal"&gt;τ&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;&lt;/span&gt;
&lt;/span&gt;
)&lt;/td&gt;
&lt;td&gt;Contextual attention weights&lt;/td&gt;
&lt;/tr&gt;
&lt;/tbody&gt;
&lt;/table&gt;&lt;/div&gt;

&lt;p&gt;Ultimately, we are looking at two brilliant but entirely distinct philosophies of computation. LLMs are built to store, parse, and reason through the accumulated, discrete data of human history. LNNs are built like agile nervous systems to react, filter noise, and adapt to the chaotic, continuous flow of the physical world.&lt;/p&gt;

</description>
      <category>ai</category>
      <category>llm</category>
      <category>lnn</category>
    </item>
    <item>
      <title>The Data Bottleneck and the Myth of "Recursive Learning" in LLMs</title>
      <dc:creator>Sujal Suyash</dc:creator>
      <pubDate>Mon, 20 Jul 2026 10:51:30 +0000</pubDate>
      <link>https://dev.to/suzuskiee/the-data-bottleneck-and-the-myth-of-recursive-learning-in-llms-597j</link>
      <guid>https://dev.to/suzuskiee/the-data-bottleneck-and-the-myth-of-recursive-learning-in-llms-597j</guid>
      <description>&lt;p&gt;&lt;a href="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fimages.unsplash.com%2Fphoto-1620712943543-bcc4688e7485%3Fq%3D80%26w%3D1000%26auto%3Dformat%26fit%3Dcrop" class="article-body-image-wrapper"&gt;&lt;img src="https://media2.dev.to/dynamic/image/width=800%2Cheight=%2Cfit=scale-down%2Cgravity=auto%2Cformat=auto/https%3A%2F%2Fimages.unsplash.com%2Fphoto-1620712943543-bcc4688e7485%3Fq%3D80%26w%3D1000%26auto%3Dformat%26fit%3Dcrop" alt="LLM Training &amp;amp; Data Bottleneck" width="1000" height="1250"&gt;&lt;/a&gt;&lt;br&gt;
&lt;/p&gt;
The ultimate competitive moat isn't just algorithmic magic; it's data hoarding and compute infrastructure.
 &lt;br&gt;&lt;br&gt;&lt;br&gt;
Trying to train a Large Language Model (LLM) from scratch requires a massive amount of computational power, resources, and, most importantly, data. Nowadays, even if we hypothetically consider that we can create a very powerful machine capable of performing large and complex computations, the biggest issue that lies in front of us is data. It is the most vital component responsible for generating a model.

&lt;p&gt;Big MNCs like Google, Anthropic, and OpenAI were able to build their own LLMs from scratch because they had huge funding, years of collected user data, and very advanced, complex machines for training. These resources are generally not available to the general public. It has become a massive business model: they create huge LLMs, and the general public is forced to purchase API keys or subscriptions to get full access to them. The ultimate competitive moat isn't just algorithmic magic; it's data hoarding and compute infrastructure.&lt;/p&gt;

&lt;p&gt;But while the economics of these models are clear, there is a common misconception about how they actually operate.&lt;/p&gt;

&lt;h2&gt;
  
  
  The Myth of Live Recursive Learning
&lt;/h2&gt;

&lt;p&gt;A popular theory is that models like ChatGPT, Claude, and Gemini are based on "Recursive Learning", that they are self-learning machines constantly updating themselves based on the data you provide in a chat, recursively minimizing their loss with every prompt you enter.&lt;/p&gt;

&lt;p&gt;In reality, current LLMs do not use a live, continuous recursive learning loop. When you interact with an LLM, its core neural network weights are completely frozen. The "learning" happens in massive, centralized, and highly structured phases long before you ever type a prompt.&lt;/p&gt;

&lt;h2&gt;
  
  
  Then how LLMs actually learn ?
&lt;/h2&gt;

&lt;p&gt;Instead of live recursive updates, models are trained in distinct and compute heavy stages:&lt;/p&gt;

&lt;ol&gt;
&lt;li&gt;
&lt;strong&gt;Pre-training:&lt;/strong&gt; The model is fed trillions of tokens from diverse sources (the web, books, code). Its only objective is to predict the next word (Self-Supervised Learning). This is where it learns grammar, facts, and reasoning patterns. The compute required here is massive, taking thousands of GPUs months to complete.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Supervised Fine-Tuning (SFT):&lt;/strong&gt; The model is trained on 10k–100k high-quality, human-curated instruction-response pairs. This teaches the model to follow instructions, adopt an "Assistant" persona, and format responses correctly.&lt;/li&gt;
&lt;li&gt;
&lt;strong&gt;Reward Modeling &amp;amp; RLHF (Reinforcement Learning from Human Feedback):&lt;/strong&gt; Humans rank different model outputs. The AI uses these rankings to train a "Judge" model, which then guides the main model to generate responses that are safer and more aligned with human preferences.&lt;/li&gt;
&lt;/ol&gt;

&lt;h2&gt;
  
  
  The Danger of &lt;strong&gt;True Recursion: Model Collapse&lt;/strong&gt;
&lt;/h2&gt;

&lt;p&gt;What would happen if an AI actually did train recursively on its own generated outputs? It leads to a critical failure known as &lt;strong&gt;Model Collapse&lt;/strong&gt;.&lt;/p&gt;

&lt;p&gt;If an LLM trains iteratively on synthetic data (data generated by itself or other LLMs) without being grounded by fresh, human-generated data, its performance degrades over successive generations. The model begins to forget rare, nuanced information and over-indexes on the most probable, generic tokens. Eventually, this amateurish recursive loop causes the model to generate repetitive, homogenized, and nonsensical outputs.&lt;/p&gt;

&lt;h2&gt;
  
  
  So where does recursion actually happens ?
&lt;/h2&gt;

&lt;p&gt;While live learning isn't happening in your chat window, &lt;strong&gt;Recursive Self-Improvement (RSI)&lt;/strong&gt; is a massive focus in AI research. AI is increasingly being used to build the next generation of AI, but as a carefully supervised engineering tool rather than an automatic black box. &lt;/p&gt;

&lt;p&gt;For instance:&lt;/p&gt;

&lt;ul&gt;
&lt;li&gt;Companies use highly capable models to generate rigorously filtered datasets to train smaller models.&lt;/li&gt;
&lt;li&gt;AI agents are used to write the code that optimizes training infrastructure.&lt;/li&gt;
&lt;/ul&gt;

&lt;p&gt;Ultimately, the biggest challenge in the field remains exactly what you might suspect: the &lt;strong&gt;DATA BOTTLENECK&lt;/strong&gt;. We are running out of high-quality, human-generated text on the internet to train the next generation of models. That finite supply of human data, combined with the extreme cost of compute, is exactly why proprietary infrastructure remains the ultimate, heavily-guarded moat.&lt;/p&gt;

</description>
      <category>ai</category>
      <category>recursivelearning</category>
      <category>llm</category>
    </item>
  </channel>
</rss>
