Function enabled 

pub fn enabled() -> bool

Examples found in repository

examples/qwen_image_denoiser_speed_fixture.rs (line 1549)

fn main() -> Result<(), Box<dyn Error>> {
    let mut args = std::env::args_os().skip(1);
    let cmf_path = PathBuf::from(
        args.next()
            .unwrap_or_else(|| "/tmp/qwen-image-speed-q4tp.cmf".into()),
    );
    let b = args
        .next()
        .map(|v| v.to_string_lossy().parse::<usize>())
        .transpose()?
        .unwrap_or(5120);
    let hidden = args
        .next()
        .map(|v| v.to_string_lossy().parse::<usize>())
        .transpose()?
        .unwrap_or(3072);
    let inter = args
        .next()
        .map(|v| v.to_string_lossy().parse::<usize>())
        .transpose()?
        .unwrap_or(12288);
    let mode = args
        .next()
        .unwrap_or_default()
        .to_string_lossy()
        .into_owned();
    let qkv = mode == "qkv";
    let qwen = mode == "qwen";
    let qwen_block = mode == "qwen-block";
    let qwen_chain = mode == "qwen-chain";
    if !mode.is_empty() && !qkv && !qwen && !qwen_block && !qwen_chain {
        return Err("optional mode must be qkv, qwen, qwen-block, or qwen-chain".into());
    }
    if b < 32 || hidden == 0 || inter == 0 || hidden % 32 != 0 || inter % 32 != 0 {
        return Err("geometry requires b>=32 and hidden/inter multiples of 32".into());
    }

    write_cmf(
        &cmf_path,
        hidden,
        inter,
        qkv,
        qwen || qwen_block || qwen_chain,
        qwen_block || qwen_chain,
    )?;
    let model = Arc::new(CmfModel::open(&cmf_path)?);
    if !model.verify().is_empty() {
        return Err(format!("fixture CMF failed verification: {:?}", model.verify()).into());
    }
    unsafe {
        std::env::set_var("CMF_GPU", "wgpu");
        std::env::set_var("CMF_GPU_PROBE", "0");
    }
    if !cortiq_engine::gpu::enabled() {
        return Err(
            "WGPU backend did not initialize; run with a real Vulkan/Metal WGPU adapter".into(),
        );
    }

    let x = input_values(b, hidden);
    let bias_in = bias_values(inter, 13);
    let bias_out = bias_values(hidden, 29);
    let x_mb = (x.len() * 4) as f64 / (1024.0 * 1024.0);
    let mid_mb = (b * inter * 4) as f64 / (1024.0 * 1024.0);
    println!(
        "speed fixture: backend=wgpu b={b} hidden={hidden} inter={inter} x_f32_mib={x_mb:.1} intermediate_f32_mib={mid_mb:.1} qkv={qkv} qwen={qwen}"
    );

    if qwen_chain {
        let text_tokens = (b / 4).max(1);
        qwen_chain_ab(&model, b, text_tokens, hidden, inter)?;
        return Ok(());
    }
    if qwen_block {
        let text_tokens = (b / 4).max(1);
        qwen_block_ab(&model, b, text_tokens, hidden, inter)?;
        return Ok(());
    }

    if qwen {
        let text_tokens = (b / 4).max(1);
        // The full image geometry is reserved for the resident attention
        // timing; a bounded 128-row panel proves the second Qwen sub-block
        // against its decomposed host reference without another large
        // activation allocation.
        qwen_mlp_ab(&model, b.min(128).max(32), hidden, inter)?;
        qwen_attention_ab(&model, b, text_tokens, hidden)?;
        return Ok(());
    }

    let (fallback, fallback_ms) = fallback_mlp(&model, &x, b, hidden, inter, &bias_in, &bias_out)?;
    let (fused, fused_ms) = fused_mlp(&model, &x, b, hidden, inter, &bias_in, &bias_out)?;
    let mlp_abs = max_abs(&fused, &fallback);
    let mlp_rel = rel_rms(&fused, &fallback);
    println!(
        "mlp A/B: fallback_scalar_ms={fallback_ms:.3} fused_resident_ms={fused_ms:.3} max_abs={mlp_abs:.8e} rel_rms={mlp_rel:.8e}"
    );
    if !fused.iter().all(|v| v.is_finite()) || mlp_abs > 0.25 || mlp_rel > 5.0e-3 {
        return Err(format!(
            "fused MLP parity failed: max_abs={mlp_abs:.8e} rel_rms={mlp_rel:.8e}"
        )
        .into());
    }

    if qkv {
        let (qkv_fallback, fallback_ms) = fallback_qkv(&model, &x, b, hidden)?;
        let (qkv_fused, fused_ms) = fused_qkv(&model, &x, b, hidden)?;
        let abs = max_abs(&qkv_fused, &qkv_fallback);
        let rel = rel_rms(&qkv_fused, &qkv_fallback);
        println!(
            "qkv A/B: fallback_3x_scalar_ms={fallback_ms:.3} fused_one_submit_ms={fused_ms:.3} max_abs={abs:.8e} rel_rms={rel:.8e}"
        );
        if !qkv_fused.iter().all(|v| v.is_finite()) || abs > 0.25 || rel > 5.0e-3 {
            return Err(
                format!("fused QKV parity failed: max_abs={abs:.8e} rel_rms={rel:.8e}").into(),
            );
        }
    }
    Ok(())
}

examples/qwen_image_denoiser_metal_fixture.rs (line 393)

fn main() -> Result<(), Box<dyn Error>> {
    if !cfg!(target_os = "macos") {
        return Err("this fixture requires the native macOS Metal backend".into());
    }
    let require_metal = !matches!(
        std::env::var("CMF_REQUIRE_METAL").as_deref(),
        Ok("0") | Ok("off")
    );
    if require_metal && std::env::var("CMF_GPU").as_deref() != Ok("1") {
        return Err("run with CMF_GPU=1 to exercise native Metal (or set CMF_REQUIRE_METAL=0 for an explicit CPU-only oracle run)".into());
    }
    let metal_active = gpu::enabled();
    if require_metal && !metal_active {
        #[cfg(target_os = "macos")]
        return Err(format!(
            "CMF_GPU=1 did not initialize a Metal backend: {:?}",
            cortiq_engine::gpu_metal::initialization_error()
        )
        .into());
        #[cfg(not(target_os = "macos"))]
        return Err("CMF_GPU=1 did not initialize a Metal backend".into());
    }
    // Enable the existing q4tp phase counters; the fixture reports counts but
    // never claims a performance result from this correctness run.
    unsafe { std::env::set_var("CMF_METAL_MMPROF", "1") };

    let mut args = std::env::args_os().skip(1);
    let fixture_path = PathBuf::from(args.next().ok_or("missing oracle JSON path")?);
    let f16_path = PathBuf::from(args.next().ok_or("missing F16 CMF path")?);
    let q4tp_path = PathBuf::from(args.next().ok_or("missing Q4TP CMF path")?);
    let fixture: Fixture = serde_json::from_slice(&std::fs::read(&fixture_path)?)?;
    let g = geometry(&fixture)?;
    let image_tokens = fixture.image.len() / g.in_channels;
    let combined = image_tokens
        .checked_add(fixture.text_len)
        .ok_or("combined sequence overflow")?;
    if image_tokens < 32 || combined < 128 || g.hidden < 64 || g.head_dim % 2 != 0 {
        return Err(format!(
            "fixture does not cross Metal gates: image_tokens={image_tokens} combined={combined} hidden={} head_dim={}",
            g.hidden, g.head_dim
        )
        .into());
    }
    write_cmf(&fixture, g, &f16_path, false)?;
    write_cmf(&fixture, g, &q4tp_path, true)?;

    let f16_model = Arc::new(CmfModel::open(&f16_path)?);
    if !f16_model.verify().is_empty() {
        return Err(format!("F16 CMF failed verification: {:?}", f16_model.verify()).into());
    }
    let q4tp_model = Arc::new(CmfModel::open(&q4tp_path)?);
    if !q4tp_model.verify().is_empty() {
        return Err(format!("Q4TP CMF failed verification: {:?}", q4tp_model.verify()).into());
    }
    let f16_transformer = QwenImageTransformer::from_cmf(&f16_model)?;
    let q4tp_transformer = QwenImageTransformer::from_cmf(&q4tp_model)?;

    #[cfg(target_os = "macos")]
    let submit_before =
        cortiq_engine::gpu_metal::METAL_SUBMITS.load(std::sync::atomic::Ordering::Relaxed);
    #[cfg(target_os = "macos")]
    let mm_before = cortiq_engine::gpu_metal::MM_N.load(std::sync::atomic::Ordering::Relaxed);

    let (f16_gpu, f16_gpu_time) = run(&f16_transformer, &fixture, false)?;
    let (f16_cpu, f16_cpu_time) = run(&f16_transformer, &fixture, true)?;
    let (q4tp_gpu, q4tp_gpu_time) = run(&q4tp_transformer, &fixture, false)?;
    let (q4tp_cpu, q4tp_cpu_time) = run(&q4tp_transformer, &fixture, true)?;

    #[cfg(target_os = "macos")]
    let submit_after =
        cortiq_engine::gpu_metal::METAL_SUBMITS.load(std::sync::atomic::Ordering::Relaxed);
    #[cfg(target_os = "macos")]
    let mm_after = cortiq_engine::gpu_metal::MM_N.load(std::sync::atomic::Ordering::Relaxed);

    if f16_gpu.len() != fixture.expected.len() || f16_cpu.len() != fixture.expected.len() {
        return Err(format!(
            "F16 output length GPU/CPU={}/{} != oracle {}",
            f16_gpu.len(),
            f16_cpu.len(),
            fixture.expected.len()
        )
        .into());
    }
    let f16_oracle_err = max_abs(&f16_cpu, &fixture.expected);
    let f16_cpu_gpu_err = max_abs(&f16_cpu, &f16_gpu);
    let q4tp_cpu_gpu_err = max_abs(&q4tp_cpu, &q4tp_gpu);
    let q4tp_vs_f16_err = max_abs(&q4tp_cpu, &f16_cpu);
    let q4tp_oracle = fixture
        .expected_q4tp
        .as_deref()
        .ok_or("oracle JSON is missing expected_q4tp")?;
    if q4tp_cpu.len() != q4tp_oracle.len() {
        return Err(format!(
            "Q4TP output length {} != dequantized oracle {}",
            q4tp_cpu.len(),
            q4tp_oracle.len()
        )
        .into());
    }
    let q4tp_oracle_err = max_abs(&q4tp_cpu, q4tp_oracle);
    if f16_gpu.iter().any(|v| !v.is_finite())
        || f16_cpu.iter().any(|v| !v.is_finite())
        || q4tp_gpu.iter().any(|v| !v.is_finite())
        || q4tp_cpu.iter().any(|v| !v.is_finite())
    {
        return Err("Metal or CPU full-forward output contained non-finite values".into());
    }
    // The native attention kernels use f32 arithmetic but reduction ordering
    // differs from the CPU path. Keep this a correctness gate, not a perf
    // claim; a multi-ulp spread is expected on a 4k sequence.
    if f16_cpu_gpu_err > 2e-3 {
        return Err(format!("F16 CPU/GPU max_abs={f16_cpu_gpu_err:.8e} > 2e-3").into());
    }
    if f16_oracle_err > 2e-3 {
        return Err(
            format!("F16 CPU/dequantized Torch max_abs={f16_oracle_err:.8e} > 2e-3").into(),
        );
    }
    if q4tp_cpu_gpu_err > 5e-3 {
        return Err(format!("Q4TP CPU/GPU max_abs={q4tp_cpu_gpu_err:.8e} > 5e-3").into());
    }
    if q4tp_oracle_err > 2e-3 {
        return Err(
            format!("Q4TP CPU/dequantized Torch max_abs={q4tp_oracle_err:.8e} > 2e-3").into(),
        );
    }

    #[cfg(target_os = "macos")]
    let (metal_submits, q4tp_matmat_calls) = (submit_after - submit_before, mm_after - mm_before);
    #[cfg(not(target_os = "macos"))]
    let (metal_submits, q4tp_matmat_calls) = (0, 0);
    if metal_active {
        if metal_submits == 0 {
            return Err("GPU forward made no Metal command submissions".into());
        }
        if q4tp_matmat_calls == 0 {
            return Err("Q4TP forward did not enter the Metal q4tp matmat path".into());
        }
    }

    println!(
        "qwen_image_denoiser_metal_fixture: backend={} image_tokens={image_tokens} combined={combined} hidden={} layers={} f16_oracle={f16_oracle_err:.8e} f16_cpu_gpu={f16_cpu_gpu_err:.8e} q4tp_oracle={q4tp_oracle_err:.8e} q4tp_cpu_gpu={q4tp_cpu_gpu_err:.8e} q4tp_vs_f16={q4tp_vs_f16_err:.8e} f16_gpu_ms={:.3} f16_cpu_ms={:.3} q4tp_gpu_ms={:.3} q4tp_cpu_ms={:.3} metal_submits={metal_submits} q4tp_matmat_calls={q4tp_matmat_calls} f16_peak={:.3e} q4tp_peak={:.3e}",
        if metal_active { "metal" } else { "cpu-only" },
        g.hidden,
        g.layers,
        f16_gpu_time.as_secs_f64() * 1e3,
        f16_cpu_time.as_secs_f64() * 1e3,
        q4tp_gpu_time.as_secs_f64() * 1e3,
        q4tp_cpu_time.as_secs_f64() * 1e3,
        max_abs_finite(&f16_cpu),
        max_abs_finite(&q4tp_cpu),
    );
    Ok(())
}