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 ¶m,
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})"), ¶m, &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}