wai-quantum 0.6.0

//! The fabric's CPU vector-unit members: the gate's fixed-point arithmetic in
//! the processor's own vector extension, one kernel per instruction set:
//!
//! - `avx2` (x86-64): 256-bit vectors, four 64-bit lanes ([`AVX2_TEXT`]);
//! - `avx512` (x86-64, the foundation instructions): 512-bit vectors, eight
//!   64-bit lanes, with the unit's own 64-bit arithmetic shift
//!   ([`AVX512_TEXT`]);
//! - `sve` (aarch64, the scalable vector extension): 128 to 2048-bit vectors,
//!   the length the processor's ([`SVE_TEXT`]);
//! - `rvv` (riscv64, the vector extension 1.0): any vector length
//!   ([`RVV_TEXT`]).
//!
//! **Why it is bit-exact.** Each kernel is the same transcription of
//! [`lanes::apply_words`] as the other members': each operand sign-extended to
//! 64 bits, each `fx(a, b) = (a·b + 2^29) >> 30` an exact 64-bit multiply, an
//! add and an arithmetic shift, each row the sum of four. The host gathers the
//! pairs a gate selects, the kernel forms their four rows a vector at a time,
//! and the host checks each row's range and keeps its low 32 bits, as
//! [`lanes::apply_words`] does. The scalable kernels set their vector length
//! per pass, so one text runs at every length; CI runs them at several.
//!
//! **Where it runs.** Discovery reads the processor's feature flags (on Linux,
//! the kernel's `AT_HWCAP`, which a user-mode emulator supplies for its guest).
//! Under an emulator the runner says so (`WAI_QUANTUM_VECTOR_EMULATOR`) and every
//! certificate carries it. Three mutants per instruction set are compiled in and
//! must each be caught.

use crate::quantum::{Amp, Circuit, Gateset, MAX_QUBITS, ONE, QuantumError, StateVector};
use crate::quantum_fabric::lanes::{self, GateWords, Lane};
use crate::quantum_fabric::plan::{self, Config, LaneRun, Program, Step};
use crate::quantum_fabric::{Certificate, Certification, Corpus, FabricClass, KernelEntry, Member, MemberError, Presence, Provenance, RunFuture, RunMeter, Runner, apply_kernel};

#[cfg(target_arch = "x86_64")]
#[path = "quantum_fabric/vector_avx2.rs"]
mod avx2;
#[cfg(target_arch = "x86_64")]
#[path = "quantum_fabric/vector_avx512.rs"]
mod avx512;

/// One probe case: [`lanes::mul_wide`] and [`lanes::fxmul_wide`] as
/// `(high, low)` words.
pub type ProbeOutput = ((u32, u32), (u32, u32));

/// Why a vector unit is not a member, or a run has no state.
#[derive(Debug, PartialEq, Eq)]
#[non_exhaustive]
pub enum VectorError {
    Unavailable(String),
    Circuit(QuantumError),
    TooWide { n_qubits: u8, max: u8 },
    OutOfRange { gate: u32 },
    Internal(&'static str),
    Probe { a: i32, b: i32 },
}

impl std::fmt::Display for VectorError {
    fn fmt(&self, f: &mut std::fmt::Formatter<'_>) -> std::fmt::Result {
        match self {
            VectorError::Unavailable(why) => write!(f, "no vector unit: {why}"),
            VectorError::Circuit(e) => write!(f, "{e}"),
            VectorError::TooWide { n_qubits, max } => write!(f, "{n_qubits} qubits; this member holds at most {max}"),
            VectorError::OutOfRange { gate } => write!(f, "the result of op {gate} leaves 32 bits"),
            VectorError::Internal(why) => write!(f, "host defect: {why}"),
            VectorError::Probe { a, b } => write!(f, "the kernel's arithmetic differs from the CPU's at ({a}, {b})"),
        }
    }
}

impl std::error::Error for VectorError {}

impl From<VectorError> for MemberError {
    fn from(e: VectorError) -> MemberError {
        match e {
            VectorError::TooWide { n_qubits, max } => MemberError::TooWide { n_qubits, max },
            VectorError::OutOfRange { gate } => MemberError::OutOfRange { gate },
            VectorError::Circuit(e) => MemberError::Circuit(e),
            other => MemberError::Other(other.to_string()),
        }
    }
}

/// The instruction sets a member runs.
#[derive(Clone, Copy, Debug, PartialEq, Eq, Hash)]
#[non_exhaustive]
pub enum Isa {
    Avx2,
    Avx512,
    Sve,
    Rvv,
}

impl Isa {
    pub const ALL: [Isa; 4] = [Isa::Avx2, Isa::Avx512, Isa::Sve, Isa::Rvv];

    /// The toolchain's name for the extension.
    pub fn label(self) -> &'static str {
        match self {
            Isa::Avx2 => "avx2",
            Isa::Avx512 => "avx512",
            Isa::Sve => "sve",
            Isa::Rvv => "rvv",
        }
    }

    /// The key a runner declares.
    pub fn key(self) -> &'static str {
        match self {
            Isa::Avx2 => "cpu-vector-avx2",
            Isa::Avx512 => "cpu-vector-avx512",
            Isa::Sve => "cpu-vector-sve",
            Isa::Rvv => "cpu-vector-rvv",
        }
    }

    /// The processor architecture the extension belongs to.
    pub fn arch(self) -> &'static str {
        match self {
            Isa::Avx2 | Isa::Avx512 => "x86_64",
            Isa::Sve => "aarch64",
            Isa::Rvv => "riscv64",
        }
    }
}

/// The kernels a member runs: the conforming one, or a mutant.
#[derive(Clone, Copy, Debug, PartialEq, Eq, Hash)]
pub enum Variant {
    Conforming,
    /// The rounding constant is half what it should be.
    M1,
    /// The product is shifted by one bit too few.
    M2,
    /// Row 0's real part adds its last term instead of subtracting it.
    M3,
}

impl Variant {
    pub const MUTANTS: [Variant; 3] = [Variant::M1, Variant::M2, Variant::M3];

    /// The kernel's constants: the rounding constant, the shift, and whether
    /// row 0's last real term is added.
    pub fn constants(self) -> (i64, i32, bool) {
        match self {
            Variant::Conforming => (1 << 29, 30, false),
            Variant::M1 => (1 << 28, 30, false),
            Variant::M2 => (1 << 29, 29, false),
            Variant::M3 => (1 << 29, 30, true),
        }
    }
}

// kernel-begin
/// The scalable vector extension's kernels: the probe (each `a·b` and
/// `fx(a, b)`) and the rows of a gate. `$rnd` is the rounding constant, `$sh`
/// the shift, `$s` the sign of row 0's last real term.
macro_rules! sve_rows_text {
    ($rnd:literal, $sh:literal, $s:literal) => {
        concat!(
            ".arch_extension sve\n",
            "ptrue p0.d\n",
            "ldr x9, [{m}, #0]\nmov z16.d, x9\n",
            "ldr x9, [{m}, #8]\nmov z17.d, x9\n",
            "ldr x9, [{m}, #16]\nmov z18.d, x9\n",
            "ldr x9, [{m}, #24]\nmov z19.d, x9\n",
            "ldr x9, [{m}, #32]\nmov z20.d, x9\n",
            "ldr x9, [{m}, #40]\nmov z21.d, x9\n",
            "ldr x9, [{m}, #48]\nmov z22.d, x9\n",
            "ldr x9, [{m}, #56]\nmov z23.d, x9\n",
            "mov x9, #", $rnd, "\nmov z31.d, x9\n",
            "mov x10, #0\n",
            "2:\n",
            "whilelt p1.d, x10, {n}\n",
            "b.none 3f\n",
            "ld1sw {{z0.d}}, p1/z, [{a0r}, x10, lsl #2]\n",
            "ld1sw {{z1.d}}, p1/z, [{a0i}, x10, lsl #2]\n",
            "ld1sw {{z2.d}}, p1/z, [{a1r}, x10, lsl #2]\n",
            "ld1sw {{z3.d}}, p1/z, [{a1i}, x10, lsl #2]\n",
            // Row 0, real.
            "movprfx z4, z0\nmul z4.d, p0/m, z4.d, z16.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nmov z8.d, z4.d\n",
            "movprfx z4, z1\nmul z4.d, p0/m, z4.d, z17.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nsub z8.d, z8.d, z4.d\n",
            "movprfx z4, z2\nmul z4.d, p0/m, z4.d, z18.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nadd z8.d, z8.d, z4.d\n",
            "movprfx z4, z3\nmul z4.d, p0/m, z4.d, z19.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\n", $s, " z8.d, z8.d, z4.d\n",
            // Row 0, imaginary.
            "movprfx z4, z1\nmul z4.d, p0/m, z4.d, z16.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nmov z9.d, z4.d\n",
            "movprfx z4, z0\nmul z4.d, p0/m, z4.d, z17.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nadd z9.d, z9.d, z4.d\n",
            "movprfx z4, z3\nmul z4.d, p0/m, z4.d, z18.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nadd z9.d, z9.d, z4.d\n",
            "movprfx z4, z2\nmul z4.d, p0/m, z4.d, z19.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nadd z9.d, z9.d, z4.d\n",
            // Row 1, real.
            "movprfx z4, z0\nmul z4.d, p0/m, z4.d, z20.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nmov z10.d, z4.d\n",
            "movprfx z4, z1\nmul z4.d, p0/m, z4.d, z21.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nsub z10.d, z10.d, z4.d\n",
            "movprfx z4, z2\nmul z4.d, p0/m, z4.d, z22.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nadd z10.d, z10.d, z4.d\n",
            "movprfx z4, z3\nmul z4.d, p0/m, z4.d, z23.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nsub z10.d, z10.d, z4.d\n",
            // Row 1, imaginary.
            "movprfx z4, z1\nmul z4.d, p0/m, z4.d, z20.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nmov z11.d, z4.d\n",
            "movprfx z4, z0\nmul z4.d, p0/m, z4.d, z21.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nadd z11.d, z11.d, z4.d\n",
            "movprfx z4, z3\nmul z4.d, p0/m, z4.d, z22.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nadd z11.d, z11.d, z4.d\n",
            "movprfx z4, z2\nmul z4.d, p0/m, z4.d, z23.d\nadd z4.d, z4.d, z31.d\nasr z4.d, z4.d, #", $sh, "\nadd z11.d, z11.d, z4.d\n",
            "st1d {{z8.d}}, p1, [{o0r}, x10, lsl #3]\n",
            "st1d {{z9.d}}, p1, [{o0i}, x10, lsl #3]\n",
            "st1d {{z10.d}}, p1, [{o1r}, x10, lsl #3]\n",
            "st1d {{z11.d}}, p1, [{o1i}, x10, lsl #3]\n",
            "incd x10\n",
            "b 2b\n",
            "3:\n",
        )
    };
}

macro_rules! sve_probe_text {
    ($rnd:literal, $sh:literal) => {
        concat!(
            ".arch_extension sve\n",
            "ptrue p0.d\n",
            "mov x9, #", $rnd, "\nmov z31.d, x9\n",
            "mov x10, #0\n",
            "2:\n",
            "whilelt p1.d, x10, {n}\n",
            "b.none 3f\n",
            "ld1sw {{z0.d}}, p1/z, [{a}, x10, lsl #2]\n",
            "ld1sw {{z1.d}}, p1/z, [{b}, x10, lsl #2]\n",
            "movprfx z4, z0\nmul z4.d, p0/m, z4.d, z1.d\n",
            "add z5.d, z4.d, z31.d\nasr z5.d, z5.d, #", $sh, "\n",
            "st1d {{z4.d}}, p1, [{prod}, x10, lsl #3]\n",
            "st1d {{z5.d}}, p1, [{fx}, x10, lsl #3]\n",
            "incd x10\n",
            "b 2b\n",
            "3:\n",
        )
    };
}

/// One term of a riscv64 row: `fx(m[$k], $a)` into v12, then `$op` into `$acc`
/// (`vmv.v.v` starts the row).
macro_rules! rvv_term {
    ($k:literal, $a:literal, $sh:literal, $op:literal, $acc:literal) => {
        concat!(
            "ld t1, ", $k, "({m})\n",
            "vmul.vx v12, ", $a, ", t1\n",
            "vadd.vx v12, v12, t2\n",
            "vsra.vi v12, v12, ", $sh, "\n",
            $op, " ", $acc, ", ", $acc, ", v12\n",
        )
    };
}

/// The riscv64 vector extension's kernels. Each pass sets the vector length
/// for what remains (`vsetvli`), loads 32-bit operands and sign-extends them to
/// 64 bits.
macro_rules! rvv_rows_text {
    ($rnd:literal, $sh:literal, $s:literal) => {
        concat!(
            ".option push\n",
            ".option arch, +v\n",
            "li t2, ", $rnd, "\n",
            "2:\n",
            "vsetvli t0, {n}, e32, mf2, ta, ma\n",
            "beqz t0, 3f\n",
            "vle32.v v1, ({a0r})\n",
            "vle32.v v2, ({a0i})\n",
            "vle32.v v3, ({a1r})\n",
            "vle32.v v4, ({a1i})\n",
            "vsetvli zero, t0, e64, m1, ta, ma\n",
            "vsext.vf2 v8, v1\n",
            "vsext.vf2 v9, v2\n",
            "vsext.vf2 v10, v3\n",
            "vsext.vf2 v11, v4\n",
            "vmv.v.i v16, 0\nvmv.v.i v17, 0\nvmv.v.i v18, 0\nvmv.v.i v19, 0\n",
            // Row 0, real.
            rvv_term!("0", "v8", $sh, "vadd.vv", "v16"),
            rvv_term!("8", "v9", $sh, "vsub.vv", "v16"),
            rvv_term!("16", "v10", $sh, "vadd.vv", "v16"),
            rvv_term!("24", "v11", $sh, $s, "v16"),
            // Row 0, imaginary.
            rvv_term!("0", "v9", $sh, "vadd.vv", "v17"),
            rvv_term!("8", "v8", $sh, "vadd.vv", "v17"),
            rvv_term!("16", "v11", $sh, "vadd.vv", "v17"),
            rvv_term!("24", "v10", $sh, "vadd.vv", "v17"),
            // Row 1, real.
            rvv_term!("32", "v8", $sh, "vadd.vv", "v18"),
            rvv_term!("40", "v9", $sh, "vsub.vv", "v18"),
            rvv_term!("48", "v10", $sh, "vadd.vv", "v18"),
            rvv_term!("56", "v11", $sh, "vsub.vv", "v18"),
            // Row 1, imaginary.
            rvv_term!("32", "v9", $sh, "vadd.vv", "v19"),
            rvv_term!("40", "v8", $sh, "vadd.vv", "v19"),
            rvv_term!("48", "v11", $sh, "vadd.vv", "v19"),
            rvv_term!("56", "v10", $sh, "vadd.vv", "v19"),
            "vse64.v v16, ({o0r})\n",
            "vse64.v v17, ({o0i})\n",
            "vse64.v v18, ({o1r})\n",
            "vse64.v v19, ({o1i})\n",
            "slli t3, t0, 2\n",
            "add {a0r}, {a0r}, t3\nadd {a0i}, {a0i}, t3\nadd {a1r}, {a1r}, t3\nadd {a1i}, {a1i}, t3\n",
            "slli t3, t0, 3\n",
            "add {o0r}, {o0r}, t3\nadd {o0i}, {o0i}, t3\nadd {o1r}, {o1r}, t3\nadd {o1i}, {o1i}, t3\n",
            "sub {n}, {n}, t0\n",
            "j 2b\n",
            "3:\n",
            ".option pop\n",
        )
    };
}

macro_rules! rvv_probe_text {
    ($rnd:literal, $sh:literal) => {
        concat!(
            ".option push\n",
            ".option arch, +v\n",
            "li t2, ", $rnd, "\n",
            "2:\n",
            "vsetvli t0, {n}, e32, mf2, ta, ma\n",
            "beqz t0, 3f\n",
            "vle32.v v1, ({a})\n",
            "vle32.v v2, ({b})\n",
            "vsetvli zero, t0, e64, m1, ta, ma\n",
            "vsext.vf2 v8, v1\n",
            "vsext.vf2 v9, v2\n",
            "vmul.vv v12, v8, v9\n",
            "vse64.v v12, ({prod})\n",
            "vadd.vx v13, v12, t2\n",
            "vsra.vi v13, v13, ", $sh, "\n",
            "vse64.v v13, ({fx})\n",
            "slli t3, t0, 2\n",
            "add {a}, {a}, t3\nadd {b}, {b}, t3\n",
            "slli t3, t0, 3\n",
            "add {prod}, {prod}, t3\nadd {fx}, {fx}, t3\n",
            "sub {n}, {n}, t0\n",
            "j 2b\n",
            "3:\n",
            ".option pop\n",
        )
    };
}
// kernel-end

/// One scalable kernel: the rows of `n` pairs.
macro_rules! sve_rows_fn {
    ($name:ident, $rnd:literal, $sh:literal, $s:literal) => {
        #[cfg(target_arch = "aarch64")]
        #[allow(clippy::too_many_arguments)]
        /// # Safety
        ///
        /// The processor has the scalable vector extension; every slice holds
        /// `n` values.
        unsafe fn $name(m: &[i64; 8], n: usize, a0r: &[i32], a0i: &[i32], a1r: &[i32], a1i: &[i32], o0r: &mut [i64], o0i: &mut [i64], o1r: &mut [i64], o1i: &mut [i64]) {
            debug_assert!([a0r.len(), a0i.len(), a1r.len(), a1i.len(), o0r.len(), o0i.len(), o1r.len(), o1i.len()].iter().all(|&l| l >= n));
            // SAFETY: the caller's contract; every vector and predicate
            // register the kernel writes is declared clobbered.
            unsafe {
                std::arch::asm!(
                    sve_rows_text!($rnd, $sh, $s),
                    m = in(reg) m.as_ptr(),
                    n = in(reg) n,
                    a0r = in(reg) a0r.as_ptr(),
                    a0i = in(reg) a0i.as_ptr(),
                    a1r = in(reg) a1r.as_ptr(),
                    a1i = in(reg) a1i.as_ptr(),
                    o0r = in(reg) o0r.as_mut_ptr(),
                    o0i = in(reg) o0i.as_mut_ptr(),
                    o1r = in(reg) o1r.as_mut_ptr(),
                    o1i = in(reg) o1i.as_mut_ptr(),
                    out("x9") _, out("x10") _, out("p0") _, out("p1") _,
                    out("v0") _, out("v1") _, out("v2") _, out("v3") _, out("v4") _, out("v5") _, out("v6") _, out("v7") _,
                    out("v8") _, out("v9") _, out("v10") _, out("v11") _, out("v12") _, out("v13") _, out("v14") _, out("v15") _,
                    out("v16") _, out("v17") _, out("v18") _, out("v19") _, out("v20") _, out("v21") _, out("v22") _, out("v23") _,
                    out("v24") _, out("v25") _, out("v26") _, out("v27") _, out("v28") _, out("v29") _, out("v30") _, out("v31") _,
                    options(nostack),
                );
            }
        }
    };
}

macro_rules! sve_probe_fn {
    ($name:ident, $rnd:literal, $sh:literal) => {
        #[cfg(target_arch = "aarch64")]
        /// # Safety
        ///
        /// The processor has the scalable vector extension; every slice holds
        /// `n` values.
        unsafe fn $name(n: usize, a: &[i32], b: &[i32], prod: &mut [i64], fx: &mut [i64]) {
            debug_assert!(a.len() >= n && b.len() >= n && prod.len() >= n && fx.len() >= n);
            // SAFETY: as for the rows.
            unsafe {
                std::arch::asm!(
                    sve_probe_text!($rnd, $sh),
                    n = in(reg) n,
                    a = in(reg) a.as_ptr(),
                    b = in(reg) b.as_ptr(),
                    prod = in(reg) prod.as_mut_ptr(),
                    fx = in(reg) fx.as_mut_ptr(),
                    out("x9") _, out("x10") _, out("p0") _, out("p1") _,
                    out("v0") _, out("v1") _, out("v2") _, out("v3") _, out("v4") _, out("v5") _, out("v6") _, out("v7") _,
                    out("v8") _, out("v9") _, out("v10") _, out("v11") _, out("v12") _, out("v13") _, out("v14") _, out("v15") _,
                    out("v16") _, out("v17") _, out("v18") _, out("v19") _, out("v20") _, out("v21") _, out("v22") _, out("v23") _,
                    out("v24") _, out("v25") _, out("v26") _, out("v27") _, out("v28") _, out("v29") _, out("v30") _, out("v31") _,
                    options(nostack),
                );
            }
        }
    };
}

macro_rules! rvv_rows_fn {
    ($name:ident, $rnd:literal, $sh:literal, $s:literal) => {
        #[cfg(target_arch = "riscv64")]
        #[allow(clippy::too_many_arguments)]
        /// # Safety
        ///
        /// The processor has the vector extension; every slice holds `n`
        /// values.
        unsafe fn $name(m: &[i64; 8], n: usize, a0r: &[i32], a0i: &[i32], a1r: &[i32], a1i: &[i32], o0r: &mut [i64], o0i: &mut [i64], o1r: &mut [i64], o1i: &mut [i64]) {
            debug_assert!([a0r.len(), a0i.len(), a1r.len(), a1i.len(), o0r.len(), o0i.len(), o1r.len(), o1i.len()].iter().all(|&l| l >= n));
            // SAFETY: the caller's contract; the pointers and count are
            // advanced in registers the kernel owns, and every vector register
            // and temporary it writes is declared clobbered.
            unsafe {
                std::arch::asm!(
                    rvv_rows_text!($rnd, $sh, $s),
                    m = in(reg) m.as_ptr(),
                    n = inout(reg) n => _,
                    a0r = inout(reg) a0r.as_ptr() => _,
                    a0i = inout(reg) a0i.as_ptr() => _,
                    a1r = inout(reg) a1r.as_ptr() => _,
                    a1i = inout(reg) a1i.as_ptr() => _,
                    o0r = inout(reg) o0r.as_mut_ptr() => _,
                    o0i = inout(reg) o0i.as_mut_ptr() => _,
                    o1r = inout(reg) o1r.as_mut_ptr() => _,
                    o1i = inout(reg) o1i.as_mut_ptr() => _,
                    out("t0") _, out("t1") _, out("t2") _, out("t3") _,
                    out("v0") _, out("v1") _, out("v2") _, out("v3") _, out("v4") _, out("v5") _, out("v6") _, out("v7") _,
                    out("v8") _, out("v9") _, out("v10") _, out("v11") _, out("v12") _, out("v13") _, out("v14") _, out("v15") _,
                    out("v16") _, out("v17") _, out("v18") _, out("v19") _, out("v20") _, out("v21") _, out("v22") _, out("v23") _,
                    out("v24") _, out("v25") _, out("v26") _, out("v27") _, out("v28") _, out("v29") _, out("v30") _, out("v31") _,
                    options(nostack),
                );
            }
        }
    };
}

macro_rules! rvv_probe_fn {
    ($name:ident, $rnd:literal, $sh:literal) => {
        #[cfg(target_arch = "riscv64")]
        /// # Safety
        ///
        /// The processor has the vector extension; every slice holds `n`
        /// values.
        unsafe fn $name(n: usize, a: &[i32], b: &[i32], prod: &mut [i64], fx: &mut [i64]) {
            debug_assert!(a.len() >= n && b.len() >= n && prod.len() >= n && fx.len() >= n);
            // SAFETY: as for the rows.
            unsafe {
                std::arch::asm!(
                    rvv_probe_text!($rnd, $sh),
                    n = inout(reg) n => _,
                    a = inout(reg) a.as_ptr() => _,
                    b = inout(reg) b.as_ptr() => _,
                    prod = inout(reg) prod.as_mut_ptr() => _,
                    fx = inout(reg) fx.as_mut_ptr() => _,
                    out("t0") _, out("t1") _, out("t2") _, out("t3") _,
                    out("v0") _, out("v1") _, out("v2") _, out("v3") _, out("v4") _, out("v5") _, out("v6") _, out("v7") _,
                    out("v8") _, out("v9") _, out("v10") _, out("v11") _, out("v12") _, out("v13") _, out("v14") _, out("v15") _,
                    out("v16") _, out("v17") _, out("v18") _, out("v19") _, out("v20") _, out("v21") _, out("v22") _, out("v23") _,
                    out("v24") _, out("v25") _, out("v26") _, out("v27") _, out("v28") _, out("v29") _, out("v30") _, out("v31") _,
                    options(nostack),
                );
            }
        }
    };
}

sve_rows_fn!(sve_rows_conforming, "0x20000000", "30", "sub");
sve_rows_fn!(sve_rows_m1, "0x10000000", "30", "sub");
sve_rows_fn!(sve_rows_m2, "0x20000000", "29", "sub");
sve_rows_fn!(sve_rows_m3, "0x20000000", "30", "add");
sve_probe_fn!(sve_probe_conforming, "0x20000000", "30");
sve_probe_fn!(sve_probe_m1, "0x10000000", "30");
sve_probe_fn!(sve_probe_m2, "0x20000000", "29");
rvv_rows_fn!(rvv_rows_conforming, "0x20000000", "30", "vsub.vv");
rvv_rows_fn!(rvv_rows_m1, "0x10000000", "30", "vsub.vv");
rvv_rows_fn!(rvv_rows_m2, "0x20000000", "29", "vsub.vv");
rvv_rows_fn!(rvv_rows_m3, "0x20000000", "30", "vadd.vv");
rvv_probe_fn!(rvv_probe_conforming, "0x20000000", "30");
rvv_probe_fn!(rvv_probe_m1, "0x10000000", "30");
rvv_probe_fn!(rvv_probe_m2, "0x20000000", "29");

/// The text the conforming x86-64 kernels are compiled from.
pub const AVX2_TEXT: &str = include_str!("quantum_fabric/vector_avx2.rs");

/// The text the conforming 512-bit x86-64 kernels are compiled from.
pub const AVX512_TEXT: &str = include_str!("quantum_fabric/vector_avx512.rs");

/// The text the conforming scalable kernels are assembled from.
pub const SVE_TEXT: &str = concat!(sve_rows_text!("0x20000000", "30", "sub"), sve_probe_text!("0x20000000", "30"));

/// The text the conforming riscv64 kernels are assembled from.
pub const RVV_TEXT: &str = concat!(rvv_rows_text!("0x20000000", "30", "vsub.vv"), rvv_probe_text!("0x20000000", "30"));

/// A variant's kernel text for `isa`.
fn text_of(isa: Isa, v: Variant) -> &'static str {
    match (isa, v) {
        (Isa::Avx2, _) => AVX2_TEXT,
        (Isa::Avx512, _) => AVX512_TEXT,
        (Isa::Sve, Variant::Conforming) => SVE_TEXT,
        (Isa::Sve, Variant::M1) => concat!(sve_rows_text!("0x10000000", "30", "sub"), sve_probe_text!("0x10000000", "30")),
        (Isa::Sve, Variant::M2) => concat!(sve_rows_text!("0x20000000", "29", "sub"), sve_probe_text!("0x20000000", "29")),
        (Isa::Sve, Variant::M3) => concat!(sve_rows_text!("0x20000000", "30", "add"), sve_probe_text!("0x20000000", "30")),
        (Isa::Rvv, Variant::Conforming) => RVV_TEXT,
        (Isa::Rvv, Variant::M1) => concat!(rvv_rows_text!("0x10000000", "30", "vsub.vv"), rvv_probe_text!("0x10000000", "30")),
        (Isa::Rvv, Variant::M2) => concat!(rvv_rows_text!("0x20000000", "29", "vsub.vv"), rvv_probe_text!("0x20000000", "29")),
        (Isa::Rvv, Variant::M3) => concat!(rvv_rows_text!("0x20000000", "30", "vadd.vv"), rvv_probe_text!("0x20000000", "30")),
    }
}

/// Version of the host layout: the gather of a gate's pairs into four
/// columns (padded with zeros to whole vectors where the kernel needs it), the
/// 64-bit row outputs, and the host's range check. Part of the implementation
/// id.
pub const HOST_LAYOUT: u32 = 1;

/// `"<isa>-lanes64/<16 hex>"` over the instruction set, a variant's kernel
/// text and constants, and the layout.
pub fn implementation_id_of(isa: Isa, v: Variant) -> String {
    let text = text_of(isa, v);
    let (rnd, sh, last_add) = v.constants();
    let mut o = b"wai:quantum-fabric-vector\x01".to_vec();
    o.extend_from_slice(isa.label().as_bytes());
    o.push(0);
    o.extend_from_slice(&(text.len() as u64).to_be_bytes());
    o.extend_from_slice(text.as_bytes());
    o.extend_from_slice(&rnd.to_be_bytes());
    o.extend_from_slice(&sh.to_be_bytes());
    o.push(u8::from(last_add));
    o.extend_from_slice(&HOST_LAYOUT.to_be_bytes());
    let h: String = blake3::hash(&o).as_bytes()[..8].iter().map(|b| format!("{b:02x}")).collect();
    format!("{}-lanes64/{h}", isa.label())
}

/// The first instruction set written for this processor's architecture.
pub fn native_isa() -> Option<Isa> {
    native_isas().into_iter().next()
}

/// Every instruction set written for this processor's architecture, in
/// [`Isa::ALL`]'s order; which of them the processor has is [`has`]'s answer.
pub fn native_isas() -> Vec<Isa> {
    Isa::ALL.into_iter().filter(|i| i.arch() == std::env::consts::ARCH).collect()
}

/// The kernel's `AT_HWCAP` on Linux: the word whose bits name the base
/// extensions (an emulator supplies its guest's).
#[cfg(all(target_os = "linux", target_arch = "riscv64"))]
fn hwcap() -> Result<u64, String> {
    const AT_HWCAP: u64 = 16;
    let auxv = std::fs::read("/proc/self/auxv").map_err(|e| format!("/proc/self/auxv: {e}"))?;
    auxv.as_chunks::<16>()
        .0
        .iter()
        .find_map(|e| {
            let (k, v) = e.split_at(8);
            (u64::from_ne_bytes(k.try_into().expect("8 bytes")) == AT_HWCAP).then(|| u64::from_ne_bytes(v.try_into().expect("8 bytes")))
        })
        .ok_or_else(|| "the kernel reports no AT_HWCAP".to_string())
}

/// Whether this processor has `isa`'s vector unit, by its feature flags.
pub fn has(isa: Isa) -> Result<(), String> {
    if isa.arch() != std::env::consts::ARCH {
        return Err(format!("this processor is {}, not {}", std::env::consts::ARCH, isa.arch()));
    }
    match isa {
        Isa::Avx2 => {
            #[cfg(target_arch = "x86_64")]
            {
                if std::arch::is_x86_feature_detected!("avx2") { Ok(()) } else { Err("the processor reports no avx2".into()) }
            }
            #[cfg(not(target_arch = "x86_64"))]
            {
                unreachable!("the architecture was checked")
            }
        }
        Isa::Avx512 => {
            #[cfg(target_arch = "x86_64")]
            {
                if std::arch::is_x86_feature_detected!("avx512f") { Ok(()) } else { Err("the processor reports no avx512f".into()) }
            }
            #[cfg(not(target_arch = "x86_64"))]
            {
                unreachable!("the architecture was checked")
            }
        }
        Isa::Sve => {
            #[cfg(target_arch = "aarch64")]
            {
                if std::arch::is_aarch64_feature_detected!("sve") { Ok(()) } else { Err("the processor reports no sve".into()) }
            }
            #[cfg(not(target_arch = "aarch64"))]
            {
                unreachable!("the architecture was checked")
            }
        }
        Isa::Rvv => {
            #[cfg(all(target_os = "linux", target_arch = "riscv64"))]
            {
                // Bit `'V' − 'A'` of AT_HWCAP is the vector extension.
                const HWCAP_V: u64 = 1 << (b'V' - b'A');
                if hwcap()? & HWCAP_V != 0 { Ok(()) } else { Err("the kernel reports no vector extension (AT_HWCAP bit 21)".into()) }
            }
            #[cfg(all(target_arch = "riscv64", not(target_os = "linux")))]
            {
                Err("feature detection is written for Linux".into())
            }
            #[cfg(not(target_arch = "riscv64"))]
            {
                unreachable!("the architecture was checked")
            }
        }
    }
}

/// The vector length in bits, read from the unit itself.
fn vector_bits(isa: Isa) -> u32 {
    match isa {
        Isa::Avx2 => 256,
        Isa::Avx512 => 512,
        Isa::Sve => {
            #[cfg(target_arch = "aarch64")]
            {
                let bytes: u64;
                // SAFETY: called only once the processor reports the extension;
                // `cntb` reads the vector length and writes one register.
                unsafe { std::arch::asm!(".arch_extension sve", "cntb {b}", b = out(reg) bytes, options(nomem, nostack, pure)) };
                (bytes * 8) as u32
            }
            #[cfg(not(target_arch = "aarch64"))]
            {
                unreachable!("the architecture was checked")
            }
        }
        Isa::Rvv => {
            #[cfg(target_arch = "riscv64")]
            {
                let bytes: u64;
                // SAFETY: called only once the kernel reports the extension;
                // `vlenb` is a read-only register.
                unsafe { std::arch::asm!(".option push", ".option arch, +v", "csrr {b}, vlenb", ".option pop", b = out(reg) bytes, options(nomem, nostack)) };
                (bytes * 8) as u32
            }
            #[cfg(not(target_arch = "riscv64"))]
            {
                unreachable!("the architecture was checked")
            }
        }
    }
}

/// What [`discover`] found.
#[derive(Debug)]
pub struct Discovery {
    pub found: Vec<VectorMember>,
    pub refused: Vec<(String, VectorError)>,
    pub unavailable: Option<String>,
}

/// This processor's vector units, one member per instruction set it reports,
/// each probed in this process; or why there is none.
pub fn discover() -> Discovery {
    let mut d = Discovery { found: Vec::new(), refused: Vec::new(), unavailable: None };
    let isas = native_isas();
    if isas.is_empty() {
        d.unavailable = Some(format!("no vector kernel is written for {}", std::env::consts::ARCH));
        return d;
    }
    let mut missing = Vec::new();
    for isa in isas {
        if let Err(why) = has(isa) {
            missing.push(why);
            continue;
        }
        match VectorMember::open(isa, Variant::Conforming) {
            Ok(m) => d.found.push(m),
            Err(e) => d.refused.push((isa.key().into(), e)),
        }
    }
    if d.found.is_empty() && d.refused.is_empty() {
        d.unavailable = Some(missing.join("; "));
    }
    d
}

/// The most qubits a run holds.
pub const MAX_VECTOR_QUBITS: u8 = if 24 < MAX_QUBITS { 24 } else { MAX_QUBITS };

/// A processor's vector unit as a fabric member.
#[derive(Clone, Debug)]
pub struct VectorMember {
    isa: Isa,
    variant: Variant,
    bits: u32,
    id: String,
    implementation: String,
    detail: String,
    provenance: Provenance,
    max_qubits: u8,
}

/// `v` padded with zeros to a whole number of `lanes`.
#[cfg(target_arch = "x86_64")]
fn padded(v: &[i32], lanes: usize) -> Vec<i32> {
    let mut p = v.to_vec();
    p.resize(v.len().next_multiple_of(lanes), 0);
    p
}

impl VectorMember {
    fn open(isa: Isa, variant: Variant) -> Result<VectorMember, VectorError> {
        let emulator = std::env::var("WAI_QUANTUM_VECTOR_EMULATOR").ok().filter(|s| !s.trim().is_empty());
        let provenance = emulator.as_ref().map_or(Provenance::Silicon, |e| Provenance::Emulator(e.trim().to_string()));
        let bits = vector_bits(isa);
        let mut o = b"wai:quantum-fabric-vector-member\x01".to_vec();
        o.extend_from_slice(isa.label().as_bytes());
        o.extend_from_slice(&bits.to_be_bytes());
        o.extend_from_slice(std::env::consts::OS.as_bytes());
        o.extend_from_slice(provenance.label().as_bytes());
        let hex: String = blake3::hash(&o).as_bytes()[..6].iter().map(|b| format!("{b:02x}")).collect();
        // The CPU reference's id with the unit after it: it sorts after the
        // reference, so with no energy figures placement keeps the reference.
        let id = format!("local/cpu-{}-vector-{}-{hex}", isa.arch(), isa.label());
        let member = VectorMember {
            isa,
            variant,
            bits,
            id,
            implementation: implementation_id_of(isa, variant),
            detail: format!("cpu; the processor's vector extension ({}), {bits}-bit vectors, 64-bit lanes; {}; up to {MAX_VECTOR_QUBITS} qubits; unmetered", isa.label(), provenance.label()),
            provenance,
            max_qubits: MAX_VECTOR_QUBITS,
        };
        if variant == Variant::Conforming {
            let pairs = lanes::probe_pairs(4096, 0x5eed);
            let out = member.probe(&pairs)?;
            if let Some((a, b)) = lanes::first_mismatch(&pairs, &out) {
                return Err(VectorError::Probe { a, b });
            }
        }
        Ok(member)
    }

    /// The same processor running a mutant: for the mutation tests.
    #[doc(hidden)]
    pub fn with_variant(&self, v: Variant) -> VectorMember {
        VectorMember { variant: v, implementation: implementation_id_of(self.isa, v), ..self.clone() }
    }

    #[doc(hidden)]
    pub fn with_max_qubits(&self, max_qubits: u8) -> VectorMember {
        VectorMember { max_qubits: max_qubits.min(self.max_qubits), ..self.clone() }
    }

    pub fn isa(&self) -> Isa {
        self.isa
    }

    /// The kernel this member runs: the conforming one, or a mutant.
    pub fn variant(&self) -> Variant {
        self.variant
    }

    /// The vector length, in bits.
    pub fn vector_bits(&self) -> u32 {
        self.bits
    }

    pub fn id(&self) -> &str {
        &self.id
    }

    pub fn key(&self) -> &str {
        self.isa.key()
    }

    pub fn class(&self) -> FabricClass {
        FabricClass::Cpu
    }

    pub fn detail(&self) -> &str {
        &self.detail
    }

    pub fn implementation(&self) -> &str {
        &self.implementation
    }

    pub fn provenance(&self) -> &Provenance {
        &self.provenance
    }

    pub fn max_qubits(&self) -> u8 {
        self.max_qubits
    }

    pub fn config(&self) -> Config {
        Config::PER_GATE
    }

    /// The kernel's [`lanes::mul_wide`] and [`lanes::fxmul_wide`] of each pair.
    pub fn probe(&self, pairs: &[(i32, i32)]) -> Result<Vec<ProbeOutput>, VectorError> {
        let a: Vec<i32> = pairs.iter().map(|p| p.0).collect();
        let b: Vec<i32> = pairs.iter().map(|p| p.1).collect();
        let (prod, fx) = self.probe_kernel(&a, &b)?;
        let split = |v: i64| ((v as u64 >> 32) as u32, v as u64 as u32);
        Ok((0..pairs.len()).map(|k| (split(prod[k]), split(fx[k]))).collect())
    }

    #[allow(unreachable_code, unused_variables)]
    fn probe_kernel(&self, a: &[i32], b: &[i32]) -> Result<(Vec<i64>, Vec<i64>), VectorError> {
        let n = a.len();
        match self.isa {
            Isa::Avx2 => {
                #[cfg(target_arch = "x86_64")]
                {
                    let (a, b) = (padded(a, avx2::LANES), padded(b, avx2::LANES));
                    let (mut prod, mut fx) = (vec![0i64; a.len()], vec![0i64; a.len()]);
                    // SAFETY: the member exists only where the processor
                    // reports avx2; every slice holds whole vectors.
                    unsafe {
                        match self.variant {
                            Variant::M1 => avx2::probe::<{ 1 << 28 }, 30>(&a, &b, &mut prod, &mut fx),
                            Variant::M2 => avx2::probe::<{ 1 << 29 }, 29>(&a, &b, &mut prod, &mut fx),
                            Variant::Conforming | Variant::M3 => avx2::probe::<{ 1 << 29 }, 30>(&a, &b, &mut prod, &mut fx),
                        }
                    }
                    prod.truncate(n);
                    fx.truncate(n);
                    return Ok((prod, fx));
                }
            }
            Isa::Avx512 => {
                #[cfg(target_arch = "x86_64")]
                {
                    let (a, b) = (padded(a, avx512::LANES), padded(b, avx512::LANES));
                    let (mut prod, mut fx) = (vec![0i64; a.len()], vec![0i64; a.len()]);
                    // SAFETY: the member exists only where the processor
                    // reports avx512f; every slice holds whole vectors.
                    unsafe {
                        match self.variant {
                            Variant::M1 => avx512::probe::<{ 1 << 28 }, 30>(&a, &b, &mut prod, &mut fx),
                            Variant::M2 => avx512::probe::<{ 1 << 29 }, 29>(&a, &b, &mut prod, &mut fx),
                            Variant::Conforming | Variant::M3 => avx512::probe::<{ 1 << 29 }, 30>(&a, &b, &mut prod, &mut fx),
                        }
                    }
                    prod.truncate(n);
                    fx.truncate(n);
                    return Ok((prod, fx));
                }
            }
            Isa::Sve => {
                #[cfg(target_arch = "aarch64")]
                {
                    let (mut prod, mut fx) = (vec![0i64; n], vec![0i64; n]);
                    // SAFETY: the member exists only where the processor
                    // reports the extension; every slice holds n values.
                    unsafe {
                        match self.variant {
                            Variant::M1 => sve_probe_m1(n, a, b, &mut prod, &mut fx),
                            Variant::M2 => sve_probe_m2(n, a, b, &mut prod, &mut fx),
                            Variant::Conforming | Variant::M3 => sve_probe_conforming(n, a, b, &mut prod, &mut fx),
                        }
                    }
                    return Ok((prod, fx));
                }
            }
            Isa::Rvv => {
                #[cfg(target_arch = "riscv64")]
                {
                    let (mut prod, mut fx) = (vec![0i64; n], vec![0i64; n]);
                    // SAFETY: as above.
                    unsafe {
                        match self.variant {
                            Variant::M1 => rvv_probe_m1(n, a, b, &mut prod, &mut fx),
                            Variant::M2 => rvv_probe_m2(n, a, b, &mut prod, &mut fx),
                            Variant::Conforming | Variant::M3 => rvv_probe_conforming(n, a, b, &mut prod, &mut fx),
                        }
                    }
                    return Ok((prod, fx));
                }
            }
        }
        Err(VectorError::Unavailable(format!("no {} kernel in a build for {}", self.isa.label(), std::env::consts::ARCH)))
    }

    /// The four rows of each pair, from the kernel.
    #[allow(unreachable_code, unused_variables)]
    fn rows(&self, m: &[i32; 8], cols: &[Vec<i32>; 4]) -> Result<[Vec<i64>; 4], VectorError> {
        let n = cols[0].len();
        match self.isa {
            Isa::Avx2 => {
                #[cfg(target_arch = "x86_64")]
                {
                    let c = [padded(&cols[0], avx2::LANES), padded(&cols[1], avx2::LANES), padded(&cols[2], avx2::LANES), padded(&cols[3], avx2::LANES)];
                    let len = c[0].len();
                    let mut o = [vec![0i64; len], vec![0i64; len], vec![0i64; len], vec![0i64; len]];
                    {
                        let [o0, o1, o2, o3] = &mut o;
                        let (c, out) = ([&c[0][..], &c[1][..], &c[2][..], &c[3][..]], [&mut o0[..], &mut o1[..], &mut o2[..], &mut o3[..]]);
                        // SAFETY: the member exists only where the processor
                        // reports avx2; every slice holds whole vectors.
                        unsafe {
                            match self.variant {
                                Variant::Conforming => avx2::rows::<{ 1 << 29 }, 30, false>(m, c, out),
                                Variant::M1 => avx2::rows::<{ 1 << 28 }, 30, false>(m, c, out),
                                Variant::M2 => avx2::rows::<{ 1 << 29 }, 29, false>(m, c, out),
                                Variant::M3 => avx2::rows::<{ 1 << 29 }, 30, true>(m, c, out),
                            }
                        }
                    }
                    for v in &mut o {
                        v.truncate(n);
                    }
                    return Ok(o);
                }
            }
            Isa::Avx512 => {
                #[cfg(target_arch = "x86_64")]
                {
                    let c = [padded(&cols[0], avx512::LANES), padded(&cols[1], avx512::LANES), padded(&cols[2], avx512::LANES), padded(&cols[3], avx512::LANES)];
                    let len = c[0].len();
                    let mut o = [vec![0i64; len], vec![0i64; len], vec![0i64; len], vec![0i64; len]];
                    {
                        let [o0, o1, o2, o3] = &mut o;
                        let (c, out) = ([&c[0][..], &c[1][..], &c[2][..], &c[3][..]], [&mut o0[..], &mut o1[..], &mut o2[..], &mut o3[..]]);
                        // SAFETY: the member exists only where the processor
                        // reports avx512f; every slice holds whole vectors.
                        unsafe {
                            match self.variant {
                                Variant::Conforming => avx512::rows::<{ 1 << 29 }, 30, false>(m, c, out),
                                Variant::M1 => avx512::rows::<{ 1 << 28 }, 30, false>(m, c, out),
                                Variant::M2 => avx512::rows::<{ 1 << 29 }, 29, false>(m, c, out),
                                Variant::M3 => avx512::rows::<{ 1 << 29 }, 30, true>(m, c, out),
                            }
                        }
                    }
                    for v in &mut o {
                        v.truncate(n);
                    }
                    return Ok(o);
                }
            }
            Isa::Sve => {
                #[cfg(target_arch = "aarch64")]
                {
                    let m64 = m.map(i64::from);
                    let mut o = [vec![0i64; n], vec![0i64; n], vec![0i64; n], vec![0i64; n]];
                    let [o0r, o0i, o1r, o1i] = &mut o;
                    // SAFETY: the member exists only where the processor
                    // reports the extension; every slice holds n values.
                    unsafe {
                        let f = match self.variant {
                            Variant::Conforming => sve_rows_conforming,
                            Variant::M1 => sve_rows_m1,
                            Variant::M2 => sve_rows_m2,
                            Variant::M3 => sve_rows_m3,
                        };
                        f(&m64, n, &cols[0], &cols[1], &cols[2], &cols[3], o0r, o0i, o1r, o1i);
                    }
                    return Ok(o);
                }
            }
            Isa::Rvv => {
                #[cfg(target_arch = "riscv64")]
                {
                    let m64 = m.map(i64::from);
                    let mut o = [vec![0i64; n], vec![0i64; n], vec![0i64; n], vec![0i64; n]];
                    let [o0r, o0i, o1r, o1i] = &mut o;
                    // SAFETY: as above.
                    unsafe {
                        let f = match self.variant {
                            Variant::Conforming => rvv_rows_conforming,
                            Variant::M1 => rvv_rows_m1,
                            Variant::M2 => rvv_rows_m2,
                            Variant::M3 => rvv_rows_m3,
                        };
                        f(&m64, n, &cols[0], &cols[1], &cols[2], &cols[3], o0r, o0i, o1r, o1i);
                    }
                    return Ok(o);
                }
            }
        }
        Err(VectorError::Unavailable(format!("no {} kernel in a build for {}", self.isa.label(), std::env::consts::ARCH)))
    }

    /// Gates in order from `state`, as [`lanes::run_words`] runs them.
    pub fn run_lanes(&self, state: &[Lane], gates: &[GateWords]) -> Result<LaneRun, VectorError> {
        let len = state.len();
        if len < 2 || !len.is_power_of_two() {
            return Err(VectorError::Internal("a state that is not 2^n lanes, n at least 1"));
        }
        let n = len.trailing_zeros() as u8;
        if n > self.max_qubits {
            return Err(VectorError::TooWide { n_qubits: n, max: self.max_qubits });
        }
        let planned: Vec<(&GateWords, u32)> = gates.iter().zip(0u32..).collect();
        self.execute(state.to_vec(), &planned)
    }

    fn execute(&self, mut amps: Vec<Lane>, gates: &[(&GateWords, u32)]) -> Result<LaneRun, VectorError> {
        let len = amps.len();
        for (w, _) in gates {
            let (tbit, ctrl) = (w.tbit as usize, w.ctrl as usize);
            if tbit == 0 || !tbit.is_power_of_two() || tbit > len / 2 || ctrl & tbit != 0 || ctrl & !(len - 1) != 0 {
                return Err(VectorError::Internal("malformed gate words for this state"));
            }
        }
        let mut report: Option<u32> = None;
        for &(w, index) in gates {
            let m: [i32; 8] = w.m.map(|x| x as i32);
            let (tbit, ctrl) = (w.tbit as usize, w.ctrl as usize);
            let mut at = Vec::new();
            let mut cols: [Vec<i32>; 4] = Default::default();
            for p in 0..len / 2 {
                let low = p & (tbit - 1);
                let i = ((p - low) << 1) | low;
                if i & ctrl != ctrl {
                    continue;
                }
                at.push(i);
                let (a0, a1) = (amps[i], amps[i | tbit]);
                cols[0].push(a0[0]);
                cols[1].push(a0[1]);
                cols[2].push(a1[0]);
                cols[3].push(a1[1]);
            }
            let [o0r, o0i, o1r, o1i] = self.rows(&m, &cols)?;
            for (k, &i) in at.iter().enumerate() {
                let rows = [o0r[k], o0i[k], o1r[k], o1i[k]];
                if rows.iter().any(|&v| i32::try_from(v).is_err()) {
                    report = Some(report.map_or(index, |r| r.min(index)));
                }
                let low = |v: i64| v as u64 as u32 as i32;
                amps[i] = [low(rows[0]), low(rows[1])];
                amps[i | tbit] = [low(rows[2]), low(rows[3])];
            }
        }
        Ok(LaneRun { amps, out_of_range: report })
    }

    pub fn run_program(&self, p: &Program) -> Result<LaneRun, VectorError> {
        let n = p.n_qubits;
        if n == 0 {
            return Ok(LaneRun { amps: vec![[ONE as i32, 0]], out_of_range: None });
        }
        if n > self.max_qubits {
            return Err(VectorError::TooWide { n_qubits: n, max: self.max_qubits });
        }
        if p.start >= 1usize << n {
            return Err(VectorError::Internal("the start is outside the state"));
        }
        let mut gates = Vec::with_capacity(p.steps.len());
        for step in &p.steps {
            let Step::Gate { gate } = *step else { return Err(VectorError::Internal("a tile step; this member runs gate steps only")) };
            let g = p.gates.get(gate as usize).ok_or(VectorError::Internal("a step names a gate the program does not hold"))?;
            if !g.is_well_formed() {
                return Err(VectorError::Internal("malformed gate words for this state"));
            }
            gates.push((&g.words, g.index));
        }
        let mut amps = vec![[0i32, 0]; 1usize << n];
        amps[p.start] = [ONE as i32, 0];
        self.execute(amps, &gates)
    }

    pub fn simulate_from(&self, c: &Circuit, start: usize) -> Result<StateVector, VectorError> {
        let p = plan::compile_with(c, start, &self.config()).map_err(VectorError::Circuit)?;
        if p.n_qubits > self.max_qubits {
            return Err(VectorError::TooWide { n_qubits: p.n_qubits, max: self.max_qubits });
        }
        let run = self.run_program(&p)?;
        if let Some(gate) = run.out_of_range {
            return Err(VectorError::OutOfRange { gate });
        }
        Ok(StateVector { n_qubits: p.n_qubits, amps: run.amps.iter().map(|l| Amp { re: i64::from(l[0]), im: i64::from(l[1]) }).collect() })
    }

    pub fn simulate(&self, c: &Circuit) -> Result<StateVector, VectorError> {
        self.simulate_from(c, 0)
    }

    /// The certificate of [`apply_kernel`]`(gs)` on this member, with its
    /// provenance.
    pub fn certify(&self, gs: Gateset, corpus: Corpus) -> Result<Certificate, VectorError> {
        let mut cert = Certification::new();
        for c in corpus.circuits(gs) {
            let reference = lanes::reference_output(&c).map_err(VectorError::Circuit)?;
            let run = self.run_program(&plan::compile_with(&c, 0, &self.config()).map_err(VectorError::Circuit)?)?;
            if let Some(gate) = run.out_of_range {
                return Err(VectorError::OutOfRange { gate });
            }
            cert.add(&lanes::circuit_bytes(&c), &reference, &lanes::output(&run.amps));
        }
        Ok(cert.finish(apply_kernel(gs), &self.implementation, &self.id, &corpus.generator(gs)).with_provenance(self.provenance.clone()))
    }

    pub fn member(&self, corpus: Corpus) -> Result<Member, VectorError> {
        let mut kernels = Vec::new();
        for gs in Gateset::ALL {
            kernels.push(KernelEntry { kernel: apply_kernel(gs).into(), implementation: self.implementation.clone(), certificate: self.certify(gs, corpus)?, energy: None });
        }
        Ok(Member { id: self.id.clone(), class: FabricClass::Cpu, presence: Presence::Detected, kernels, detail: self.detail.clone() })
    }
}

impl Runner for VectorMember {
    fn member_id(&self) -> String {
        self.id.clone()
    }

    fn run<'r>(&'r self, implementation: &'r str, c: &'r Circuit, start: usize, _meter: &'r mut dyn RunMeter) -> RunFuture<'r> {
        Box::pin(async move {
            if implementation != self.implementation {
                return Err(MemberError::Other(format!("member {} runs implementation {}, not {implementation:?}", self.id, self.implementation)));
            }
            Ok((self.simulate_from(c, start)?, 0, None))
        })
    }
}

#[cfg(test)]
mod tests {
    use super::*;

    #[test]
    fn every_variant_of_every_isa_has_its_own_implementation_id() {
        let ids: std::collections::BTreeSet<String> = Isa::ALL
            .into_iter()
            .flat_map(|isa| [Variant::Conforming, Variant::M1, Variant::M2, Variant::M3].into_iter().map(move |v| implementation_id_of(isa, v)))
            .collect();
        assert_eq!(ids.len(), 16);
        for isa in Isa::ALL {
            assert!(implementation_id_of(isa, Variant::Conforming).starts_with(&format!("{}-lanes64/", isa.label())));
        }
        assert!(AVX2_TEXT.contains("_mm256_mul_epi32"));
        assert!(AVX512_TEXT.contains("_mm512_mul_epi32") && AVX512_TEXT.contains("_mm512_srai_epi64"));
        assert!(SVE_TEXT.contains("whilelt") && !SVE_TEXT.contains("smstart"));
        assert!(RVV_TEXT.contains("vsetvli") && RVV_TEXT.contains("vsext.vf2"));
    }

    #[test]
    fn the_mutants_differ_in_their_constants() {
        assert_eq!(Variant::Conforming.constants(), (1 << 29, 30, false));
        for v in Variant::MUTANTS {
            assert_ne!(v.constants(), Variant::Conforming.constants());
        }
    }

    #[test]
    fn the_member_id_sorts_after_the_cpu_reference() {
        for isa in Isa::ALL {
            let reference = format!("local/cpu-{}", isa.arch());
            let member = format!("local/cpu-{}-vector-{}-000000000000", isa.arch(), isa.label());
            assert!(member > reference, "{member} vs {reference}");
        }
    }

    #[test]
    fn a_processor_without_a_kernel_or_the_unit_says_why() {
        let d = discover();
        if d.found.is_empty() {
            assert!(d.refused.is_empty() && d.unavailable.is_some(), "{d:?}");
        }
        for isa in Isa::ALL.into_iter().filter(|i| i.arch() != std::env::consts::ARCH) {
            assert!(has(isa).unwrap_err().contains("this processor is"));
        }
    }
}