Skip to main content

rlx_optim/
lib.rs

1// RLX — versatile ML compiler + runtime.
2// Copyright (C) 2026 Eugene Hauptmann, Nataliya Kosmyna.
3// SPDX-License-Identifier: MIT OR Apache-2.0
4
5//! RLX training-step optimizers.
6//!
7//! Host-side `f32` step functions for the families surveyed in
8//! "A Systematic Review of Optimization Algorithms for Modern Deep
9//! Learning" (arXiv:2509.02046v1). Each algorithm exposes a small
10//! state struct keyed by parameter *name* (so the same struct holds
11//! moments for every tensor in a model) and a `step` method that
12//! consumes `(name, shape, &mut params, &grads)`.
13//!
14//! The API is deliberately minimal: it operates on flat `&mut [f32]`
15//! / `&[f32]` slices plus a `&[usize]` shape — matching the
16//! [`rlx_umap::adam`](../rlx_umap/adam/index.html) pattern. Backends
17//! that already ship a fused step kernel (see e.g.
18//! `rlx_metal::splat_adam`) are free to bypass this crate for their
19//! hot path; this crate is the portable reference / CPU fallback / the
20//! one used when there is no backend fused kernel for the requested
21//! algorithm.
22//!
23//! # Algorithms
24//!
25//! | Family          | Type                          |
26//! |-----------------|-------------------------------|
27//! | [`Sgd`]         | SGD ± momentum / Nesterov     |
28//! | [`Adam`]        | Adam                          |
29//! | [`AdamW`]       | AdamW (decoupled decay)       |
30//! | [`NAdamW`]      | Nesterov AdamW                |
31//! | [`RAdam`]       | Rectified Adam                |
32//! | [`QHAdamW`]     | Quasi-hyperbolic AdamW        |
33//! | [`Lamb`]        | LAMB (layer-wise adaptive)    |
34//! | [`Adafactor`]   | Adafactor (factored 2nd mom.) |
35//! | [`Lion`]        | Lion (sign of EMA)            |
36//! | [`Soap`]        | SOAP (Shampoo-in-Adam-basis)  |
37//! | [`KronPsgd`]    | Kron / PSGD                   |
38//! | [`Muon`]        | Muon (Newton–Schulz orth.)    |
39//! | [`Sophia`]      | Sophia-H                      |
40//! | [`Mars`]        | MARS (variance-reduced)       |
41//! | [`Stiefel`]     | Riemannian SGD on St(m,n)     |
42
43// Pure-safe by default. Relaxed from `forbid` to `deny` so the one
44// performance-critical exception — the Accelerate/AMX `cblas_sgemm` shim behind
45// Muon's Newton–Schulz on macOS — can opt in with a scoped `#[allow(unsafe_code)]`
46// (see `muon::accel`). Every other module stays unsafe-free.
47#![deny(unsafe_code)]
48
49mod common;
50
51mod adafactor;
52mod adam;
53mod adamw;
54mod kron_psgd;
55mod lamb;
56mod lion;
57mod mars;
58mod muon;
59mod nadamw;
60mod qhadamw;
61mod radam;
62mod sgd;
63mod soap;
64mod sophia;
65mod stiefel;
66
67use std::collections::HashMap;
68
69pub use adafactor::Adafactor;
70pub use adam::Adam;
71pub use adamw::AdamW;
72pub use kron_psgd::KronPsgd;
73pub use lamb::Lamb;
74pub use lion::Lion;
75pub use mars::Mars;
76pub use muon::{Muon, newton_schulz_orth};
77pub use nadamw::NAdamW;
78pub use qhadamw::QHAdamW;
79pub use radam::RAdam;
80pub use sgd::Sgd;
81pub use soap::Soap;
82pub use sophia::Sophia;
83pub use stiefel::Stiefel;
84
85pub use common::{global_grad_clip_scale, l2_norm};
86
87/// Common parameter-update interface.
88///
89/// `name` keys the per-parameter state (moments, preconditioners),
90/// `shape` is the parameter's logical shape (used by matrix-aware
91/// algorithms like Adafactor / SOAP / Muon — ignored by elementwise
92/// ones), `param` is updated in place from `grad`. `grad` is treated
93/// as read-only; callers that need gradient clipping should pre-scale
94/// it (see [`global_grad_clip_scale`]).
95///
96/// # Implementing for a backend
97///
98/// Every algorithm in this crate provides a CPU reference impl. A
99/// backend (e.g. `rlx-metal`, `rlx-cuda`) is free to write its own
100/// fused step kernel and impl `Optimizer` for a wrapper struct that
101/// owns device buffers — the trait places no requirement on where
102/// the state lives, only on the entry-point signature. The
103/// `rlx-metal::splat_adam` kernel is the canonical example of a
104/// backend that bypasses this crate entirely; you can wrap it with a
105/// 5-line `impl Optimizer` if you want a uniform interface from a
106/// generic trainer.
107///
108/// # Per-tensor learning rate
109///
110/// For optimizers that don't need per-tensor LR variation (most
111/// transformer pre-training), set `lr_scale` to
112/// return `1.0` (the default). For domain-specific use cases — e.g.
113/// 3D Gaussian splatting, where different attributes need wildly
114/// different step sizes — override `lr_scale` to
115/// multiply the base `lr` by a per-name factor. The provided method
116/// on the trait does NOT scale automatically; algorithms are free to
117/// consult it via [`Optimizer::lr_scale`] inside their `step`.
118/// One parameter's data for a batched optimizer step ([`Optimizer::step_batch`]):
119/// its name, static shape, mutable data slice (owned by the caller), and gradient.
120pub struct OptItem<'a> {
121    pub name: &'a str,
122    pub shape: &'a [usize],
123    pub param: &'a mut [f32],
124    pub grad: &'a [f32],
125}
126
127/// An optimizer's internal state, for checkpointing a training run.
128///
129/// `buffers` are the per-parameter accumulators keyed `"<slot>/<param>"` — Adam
130/// writes `"m/w1"` and `"v/w1"` — so a reload can match them back up by name
131/// even if the parameter order changed. `step` is the bias-correction counter.
132///
133/// Resuming without this is not resuming: Adam's moments are most of what it
134/// knows, and restarting them at zero produces a loss spike and a different
135/// trajectory from the one that was interrupted.
136#[derive(Debug, Clone, Default, PartialEq)]
137pub struct OptimizerState {
138    pub step: u64,
139    pub buffers: Vec<(String, Vec<f32>)>,
140}
141
142impl OptimizerState {
143    /// Collect one named slot out of a `HashMap<param, Vec<f32>>`.
144    pub fn extend_slot(&mut self, slot: &str, map: &HashMap<String, Vec<f32>>) {
145        let mut names: Vec<&String> = map.keys().collect();
146        // Sorted, so a checkpoint's byte layout does not depend on hash order.
147        names.sort();
148        for name in names {
149            self.buffers
150                .push((format!("{slot}/{name}"), map[name].clone()));
151        }
152    }
153
154    /// Restore one named slot into a `HashMap<param, Vec<f32>>`.
155    pub fn take_slot(&self, slot: &str, map: &mut HashMap<String, Vec<f32>>) {
156        let prefix = format!("{slot}/");
157        for (key, values) in &self.buffers {
158            if let Some(name) = key.strip_prefix(&prefix) {
159                map.insert(name.to_string(), values.clone());
160            }
161        }
162    }
163}
164
165pub trait Optimizer {
166    fn step(&mut self, name: &str, shape: &[usize], param: &mut [f32], grad: &[f32]);
167
168    /// Batched step over ALL parameters in one call. Default: sequential
169    /// `step` per item — bit-identical to the per-parameter loop. Optimizers
170    /// whose parameter groups are **independent** (e.g. Muon on the 2-D weight
171    /// matrices vs AdamW on the embeddings/biases/norms) can override this to
172    /// run the groups on separate threads; because the groups touch disjoint
173    /// parameters and disjoint optimizer state, the result is bit-for-bit the
174    /// same as the serial loop — only the wall-clock (the CPU-side optimizer
175    /// bubble) shrinks toward `max(group_times)` instead of their sum.
176    fn step_batch(&mut self, items: &mut [OptItem<'_>]) {
177        for it in items.iter_mut() {
178            self.step(it.name, it.shape, it.param, it.grad);
179        }
180    }
181
182    /// Advance the global step counter. Most algorithms increment per
183    /// call to `step`, so most implementations leave this a no-op.
184    fn end_iteration(&mut self) {}
185
186    /// Set the base learning rate (for LR schedules / warmup). Default is a
187    /// no-op for algorithms without a scalar `lr` (e.g. Adafactor's relative
188    /// step size); every algorithm in this crate that has an `lr` field
189    /// overrides this to update it.
190    fn set_lr(&mut self, _lr: f32) {}
191
192    /// Per-tensor multiplier on the effective learning rate. Default
193    /// is `1.0` for every name. Override when wrapping this crate to
194    /// support per-name LR schedules (e.g. embedding-vs-attention
195    /// splits, or the Gaussian-splat attribute-typed LR setup). The
196    /// CPU impls in this crate currently honor this only when the
197    /// caller passes a pre-scaled `lr` for the relevant call —
198    /// backends are encouraged to consult it inside their fused
199    /// kernel.
200    fn lr_scale(&self, _name: &str) -> f32 {
201        1.0
202    }
203
204    /// Snapshot the optimizer's state for a checkpoint.
205    ///
206    /// `None` means this algorithm has not opted in, so a run using it **cannot
207    /// be resumed exactly**. Callers should say so rather than silently
208    /// restarting the accumulators at zero.
209    fn state_dict(&self) -> Option<OptimizerState> {
210        None
211    }
212
213    /// Restore a snapshot. Returns false when unsupported or when the state does
214    /// not belong to this algorithm.
215    fn load_state_dict(&mut self, _state: &OptimizerState) -> bool {
216        false
217    }
218}
219
220#[cfg(test)]
221mod state_dict_tests {
222    use super::*;
223
224    /// A checkpointed optimizer must produce the *same next step* as one that
225    /// was never interrupted. Restarting the moments at zero passes a
226    /// "does it train" test and still changes the trajectory, so this compares
227    /// against an uninterrupted reference rather than against itself.
228    fn resume_matches_uninterrupted<O: Optimizer, F: Fn() -> O>(make: F) {
229        let grads: [[f32; 4]; 6] = [
230            [0.5, -0.25, 0.125, 1.0],
231            [0.4, -0.20, 0.100, 0.9],
232            [0.3, -0.15, 0.075, 0.8],
233            [0.2, -0.10, 0.050, 0.7],
234            [0.1, -0.05, 0.025, 0.6],
235            [0.05, -0.02, 0.01, 0.5],
236        ];
237        let shape = [2usize, 2];
238
239        // Reference: six steps, no interruption.
240        let mut reference = make();
241        let mut p_ref = [1.0f32, 1.0, 1.0, 1.0];
242        for g in &grads {
243            reference.step("w", &shape, &mut p_ref, g);
244            reference.end_iteration();
245        }
246
247        // Interrupted: three steps, snapshot, rebuild, restore, three more.
248        let mut first = make();
249        let mut p = [1.0f32, 1.0, 1.0, 1.0];
250        for g in &grads[..3] {
251            first.step("w", &shape, &mut p, g);
252            first.end_iteration();
253        }
254        let state = first.state_dict().expect("this optimizer opted in");
255        drop(first);
256
257        let mut second = make();
258        assert!(second.load_state_dict(&state), "load_state_dict refused");
259        for g in &grads[3..] {
260            second.step("w", &shape, &mut p, g);
261            second.end_iteration();
262        }
263
264        for (i, (a, b)) in p_ref.iter().zip(p.iter()).enumerate() {
265            assert_eq!(
266                a.to_bits(),
267                b.to_bits(),
268                "element {i}: resumed {b} != uninterrupted {a}"
269            );
270        }
271    }
272
273    #[test]
274    fn adam_resumes_bit_exactly() {
275        resume_matches_uninterrupted(|| Adam::new(0.1));
276    }
277
278    #[test]
279    fn adamw_resumes_bit_exactly() {
280        resume_matches_uninterrupted(|| AdamW::new(0.1));
281    }
282
283    #[test]
284    fn sgd_with_momentum_resumes_bit_exactly() {
285        resume_matches_uninterrupted(|| {
286            let mut o = Sgd::new(0.1);
287            o.momentum = 0.9;
288            o
289        });
290    }
291
292    #[test]
293    fn lion_resumes_bit_exactly() {
294        resume_matches_uninterrupted(|| Lion::new(0.1));
295    }
296
297    #[test]
298    fn an_optimizer_without_state_says_so() {
299        // Honest `None` beats a snapshot that silently drops the accumulators:
300        // the caller can warn that the run is not exactly resumable.
301        assert!(Soap::new(0.1).state_dict().is_none());
302        assert!(!Soap::new(0.1).load_state_dict(&OptimizerState::default()));
303    }
304}
305
306#[cfg(test)]
307mod hot_loop_tests {
308    use super::*;
309
310    fn spread(n: usize, seed: u32) -> Vec<f32> {
311        let mut s = seed;
312        (0..n)
313            .map(|_| {
314                s = s.wrapping_mul(1_664_525).wrapping_add(1_013_904_223);
315                ((s >> 8) as f32 / 8_388_608.0 - 1.0) * 0.1
316            })
317            .collect()
318    }
319
320    fn assert_bit_equal(label: &str, got: &[f32], want: &[f32]) {
321        assert_eq!(got.len(), want.len(), "{label}: length");
322        for (i, (a, b)) in got.iter().zip(want).enumerate() {
323            assert_eq!(
324                a.to_bits(),
325                b.to_bits(),
326                "{label}[{i}]: {a} != {b} (bit-exactness is the contract here — the loops \
327                 were restructured for speed, not for different arithmetic)"
328            );
329        }
330    }
331
332    /// SGD's hot loop was rewritten into one specialized loop per configuration
333    /// (it used to test `mu == 0.0` and read `self.nesterov` per element, which
334    /// blocked vectorization and cost 7x). The arithmetic must be untouched, so
335    /// this compares against the original formulas written out by hand.
336    #[test]
337    fn sgd_matches_its_scalar_reference_bit_exactly() {
338        const N: usize = 257; // deliberately not a multiple of any vector width
339        let grad = spread(N, 11);
340        for (momentum, nesterov, weight_decay) in [
341            (0.0f32, false, 0.0f32),
342            (0.0, false, 0.05),
343            (0.9, false, 0.0),
344            (0.9, false, 0.05),
345            (0.9, true, 0.0),
346            (0.9, true, 0.05),
347        ] {
348            let lr = 0.03f32;
349            let mut opt = Sgd::new(lr);
350            opt.momentum = momentum;
351            opt.nesterov = nesterov;
352            opt.weight_decay = weight_decay;
353            let mut param = spread(N, 7);
354
355            // The reference: exactly the expressions the original loop used.
356            let mut ref_param = spread(N, 7);
357            let mut ref_v = vec![0.0f32; N];
358
359            for _ in 0..4 {
360                opt.step("w", &[N], &mut param, &grad);
361                opt.end_iteration();
362                for i in 0..N {
363                    let g = grad[i] + weight_decay * ref_param[i];
364                    if momentum == 0.0 {
365                        ref_param[i] -= lr * g;
366                    } else {
367                        ref_v[i] = momentum * ref_v[i] + g;
368                        let update = if nesterov {
369                            g + momentum * ref_v[i]
370                        } else {
371                            ref_v[i]
372                        };
373                        ref_param[i] -= lr * update;
374                    }
375                }
376            }
377            assert_bit_equal(
378                &format!("sgd(mu={momentum}, nesterov={nesterov}, wd={weight_decay})"),
379                &param,
380                &ref_param,
381            );
382        }
383    }
384
385    /// Adam gained an `f32_math` option (AdamW already had one). The default
386    /// must still be the f64-intermediate path, bit for bit, and the new path
387    /// must be the f32 arithmetic it claims.
388    #[test]
389    fn adam_f32_and_f64_paths_each_match_their_reference() {
390        const N: usize = 129;
391        let grad = spread(N, 11);
392        let (lr, b1, b2, eps, wd) = (0.03f32, 0.9f32, 0.999f32, 1e-8f32, 0.01f32);
393
394        for f32_math in [false, true] {
395            let mut opt = Adam::new(lr).with_f32_math(f32_math);
396            opt.beta1 = b1;
397            opt.beta2 = b2;
398            opt.eps = eps;
399            opt.weight_decay = wd;
400            let mut param = spread(N, 7);
401
402            let mut ref_param = spread(N, 7);
403            let mut ref_m = vec![0.0f32; N];
404            let mut ref_v = vec![0.0f32; N];
405
406            for step in 0..4 {
407                opt.step("w", &[N], &mut param, &grad);
408                opt.end_iteration();
409                if f32_math {
410                    let t = step + 1;
411                    let bc1 = 1.0 - b1.powi(t);
412                    let bc2 = 1.0 - b2.powi(t);
413                    for i in 0..N {
414                        let g = grad[i] + wd * ref_param[i];
415                        ref_m[i] = b1 * ref_m[i] + (1.0 - b1) * g;
416                        ref_v[i] = b2 * ref_v[i] + (1.0 - b2) * (g * g);
417                        let m_hat = ref_m[i] / bc1;
418                        let v_hat = ref_v[i] / bc2;
419                        ref_param[i] -= lr * (m_hat / (v_hat.sqrt() + eps));
420                    }
421                } else {
422                    let t = (step + 1) as f64;
423                    let (b1, b2) = (b1 as f64, b2 as f64);
424                    let bc1 = 1.0 - b1.powf(t);
425                    let bc2 = 1.0 - b2.powf(t);
426                    let (eps, lr) = (eps as f64, lr as f64);
427                    for i in 0..N {
428                        let g = (grad[i] + wd * ref_param[i]) as f64;
429                        let new_m = b1 * ref_m[i] as f64 + (1.0 - b1) * g;
430                        let new_v = b2 * ref_v[i] as f64 + (1.0 - b2) * g * g;
431                        ref_m[i] = new_m as f32;
432                        ref_v[i] = new_v as f32;
433                        let m_hat = new_m / bc1;
434                        let v_hat = new_v / bc2;
435                        ref_param[i] -= (lr * m_hat / (v_hat.sqrt() + eps)) as f32;
436                    }
437                }
438            }
439            assert_bit_equal(&format!("adam(f32_math={f32_math})"), &param, &ref_param);
440        }
441    }
442
443    /// The two paths must *differ* — otherwise the option is a no-op and the
444    /// 2.1x is coming from somewhere unexplained.
445    #[test]
446    fn adams_two_precision_paths_are_not_the_same_computation() {
447        const N: usize = 64;
448        let grad = spread(N, 11);
449        let run = |f32_math: bool| {
450            let mut opt = Adam::new(0.03).with_f32_math(f32_math);
451            let mut param = spread(N, 7);
452            for _ in 0..8 {
453                opt.step("w", &[N], &mut param, &grad);
454                opt.end_iteration();
455            }
456            param
457        };
458        let f64_path = run(false);
459        let f32_path = run(true);
460        assert!(
461            f64_path
462                .iter()
463                .zip(&f32_path)
464                .any(|(a, b)| a.to_bits() != b.to_bits()),
465            "f32_math made no difference, so it is not selecting a different path"
466        );
467        // …but they must agree to f32 tolerance, or one of them is wrong.
468        for (i, (a, b)) in f64_path.iter().zip(&f32_path).enumerate() {
469            assert!(
470                (a - b).abs() <= 1e-5 * a.abs().max(1.0),
471                "element {i}: f64 path {a} vs f32 path {b}"
472            );
473        }
474    }
475
476    /// Nothing above should depend on the length being a nice multiple, since a
477    /// vectorized loop has a scalar tail.
478    #[test]
479    fn odd_lengths_are_handled_by_the_tail() {
480        for n in [1usize, 3, 7, 15, 16, 17, 31, 33, 63, 65] {
481            let grad = spread(n, 11);
482            let mut param = spread(n, 7);
483            let before = param.clone();
484            let mut opt = Sgd::new(0.1);
485            opt.momentum = 0.9;
486            opt.step("w", &[n], &mut param, &grad);
487            assert!(
488                param.iter().zip(&before).any(|(a, b)| a != b),
489                "n={n}: nothing moved"
490            );
491            assert!(param.iter().all(|x| x.is_finite()), "n={n}: not finite");
492        }
493    }
494}