Develop a JPEG through the same pipeline as a RAW

DemosaicedImage gains a second producer, from_rgba8, alongside the CFA path.
Nothing about the type is CFA-specific — it is "an image on the GPU, ready to
adjust" — which is what lets develop mode work on a JPEG without the edit
graph or any operation knowing the source was not a RAW file.

The one real difference is the transfer function: sensor data is linear, a
JPEG is gamma-encoded. Every operation assumes linear scene-referred colour
(exposure is a multiply, and doubling a gamma-encoded value is not a stop), so
the shader prologue linearises once, at the only point where the two source
kinds still differ. The flag rides in as_shot_wb.w, which was padding. For a
JPEG the white balance uniform is neutral and the colour matrix is identity,
so both stay unconditional multiplies rather than becoming branches.

max_dimension is exposed because it is a hardware limit the caller must plan
around, not a failure to report afterwards: a 13728x8928 film scan exceeds the
common 8192 texture limit, and fitting it first is the only way to develop it
at all.

Assisted-by: LLM

core/dr-decode/src/lib.rs

@@ -13,9 +13,11 @@
 //! is exactly why Lightroom stalls ~2 s per image during culling.
 
 mod error;
+mod locate;
 mod preview;
 
 pub use error::DecodeError;
+pub use locate::{is_complete_jpeg, locate_preview, PreviewLocation, HEADER_BYTES};
 pub use preview::{
     decode_jpeg, extract_embedded_preview, extract_preview, Preview, PreviewSize,
     PREVIEW_PROBE_BYTES,
@@ -37,6 +39,19 @@ pub struct Metadata {
     /// Full sensor dimensions, before crop.
     pub width: Option<u32>,
     pub height: Option<u32>,
+    /// When the shutter fired, as Unix seconds.
+    ///
+    /// EXIF records wall-clock time with no zone, so this is that reading
+    /// interpreted as UTC. Paired with [`captured_offset`](Self::captured_offset)
+    /// it reconstructs the actual instant; alone it is still correct for
+    /// ordering within one timezone, which is what a timeline needs.
+    pub captured_at: Option<i64>,
+    /// Minutes east of UTC, where the camera recorded a zone.
+    ///
+    /// Absent on most bodies before ~2018. A photograph's timestamp is local
+    /// to where it was taken, so without this a shoot in Tokyo displays on the
+    /// wrong day in Paris.
+    pub captured_offset: Option<i32>,
 }
 
 /// Decoded sensor data, before demosaic.
@@ -218,6 +233,14 @@ pub fn probe(header: &[u8]) -> Option<Format> {
 pub fn metadata(bytes: &[u8]) -> Result<Metadata, DecodeError> {
     use rawler::rawsource::RawSource;
 
+    // rawler has no decoder for a plain JPEG, so without this every JPEG in a
+    // library reports no capture time — and a mixed library's timeline is
+    // silently missing thousands of images. Scanned film and camera JPEGs both
+    // land here.
+    if bytes.starts_with(&[0xFF, 0xD8, 0xFF]) {
+        return locate::jpeg_metadata(bytes);
+    }
+
     let source = RawSource::new_from_slice(bytes);
     let decoder =
         rawler::get_decoder(&source).map_err(|e| DecodeError::Unsupported(e.to_string()))?;
@@ -226,7 +249,7 @@ pub fn metadata(bytes: &[u8]) -> Result<Metadata, DecodeError> {
         .map_err(|e| DecodeError::Metadata(e.to_string()))?;
 
     let exif = &md.exif;
-    Ok(Metadata {
+    let mut out = Metadata {
         make: Some(md.make.clone()).filter(|s| !s.is_empty()),
         model: Some(md.model.clone()).filter(|s| !s.is_empty()),
         lens: exif.lens_model.clone(),
@@ -236,7 +259,101 @@ pub fn metadata(bytes: &[u8]) -> Result<Metadata, DecodeError> {
         focal_length: exif.focal_length.map(|r| r.n as f32 / r.d.max(1) as f32),
         width: None,
         height: None,
-    })
+        captured_at: exif
+            .date_time_original
+            .as_deref()
+            .and_then(parse_exif_datetime),
+        captured_offset: exif
+            .offset_time_original
+            .as_deref()
+            .or(exif.offset_time.as_deref())
+            .and_then(parse_exif_offset),
+    };
+
+    // rawler reports no capture time for some TIFF-derived files whose tag is
+    // plainly present — one reference DNG carries it at byte 826 and still
+    // comes back empty. These formats *are* TIFF, so the same reader the JPEG
+    // path uses can find it. Only the missing fields are filled, so rawler
+    // stays authoritative wherever it did answer.
+    if out.captured_at.is_none() {
+        if let Ok(fallback) = locate::tiff_metadata(bytes) {
+            out.captured_at = fallback.captured_at;
+            out.captured_offset = out.captured_offset.or(fallback.captured_offset);
+            out.iso = out.iso.or(fallback.iso);
+            out.lens = out.lens.take().or(fallback.lens);
+        }
+    }
+
+    Ok(out)
+}
+
+/// Parse an EXIF `DateTimeOriginal` into Unix seconds.
+///
+/// The format is `"YYYY:MM:DD HH:MM:SS"` — colons in the date, which is what
+/// trips generic date parsers. No timezone is present, so the reading is taken
+/// as UTC and the zone, if any, comes from `OffsetTimeOriginal` separately.
+///
+/// Returns `None` rather than guessing on anything malformed: a wrong
+/// timestamp puts an image at the wrong point on the timeline, which is worse
+/// than leaving it unplaced.
+pub(crate) fn parse_exif_datetime(s: &str) -> Option<i64> {
+    let s = s.trim();
+    let (date, time) = s.split_once(' ')?;
+    // EXIF specifies colons in the date, but real files disagree: the
+    // CanoScan 9000F writes `2013/06/28`, and enough devices use dashes that
+    // rejecting either would leave whole classes of file undated.
+    let mut d = date.split([':', '/', '-']);
+    let (y, mo, da): (i64, i64, i64) = (
+        d.next()?.parse().ok()?,
+        d.next()?.parse().ok()?,
+        d.next()?.parse().ok()?,
+    );
+    let mut t = time.split(':');
+    let (h, mi, se): (i64, i64, i64) = (
+        t.next()?.parse().ok()?,
+        t.next()?.parse().ok()?,
+        // Some bodies append fractional seconds; take the whole part.
+        t.next()?.split('.').next()?.parse().ok()?,
+    );
+
+    // A camera with a dead clock battery reports 1970 or similar. Reject
+    // obvious nonsense rather than clustering those images at the epoch.
+    if !(1900..=2200).contains(&y)
+        || !(1..=12).contains(&mo)
+        || !(1..=31).contains(&da)
+        || !(0..=23).contains(&h)
+        || !(0..=59).contains(&mi)
+        || !(0..=60).contains(&se)
+    {
+        return None;
+    }
+
+    // Days from the civil date, via the usual era-based algorithm.
+    let y_adj = if mo <= 2 { y - 1 } else { y };
+    let era = if y_adj >= 0 { y_adj } else { y_adj - 399 } / 400;
+    let yoe = y_adj - era * 400;
+    let mp = (mo + 9) % 12;
+    let doy = (153 * mp + 2) / 5 + da - 1;
+    let doe = yoe * 365 + yoe / 4 - yoe / 100 + doy;
+    let days = era * 146_097 + doe - 719_468;
+
+    Some(days * 86_400 + h * 3_600 + mi * 60 + se)
+}
+
+/// Parse an EXIF offset like `"+02:00"` into minutes east of UTC.
+pub(crate) fn parse_exif_offset(s: &str) -> Option<i32> {
+    let s = s.trim();
+    let (sign, rest) = match s.as_bytes().first()? {
+        b'+' => (1, &s[1..]),
+        b'-' => (-1, &s[1..]),
+        _ => return None,
+    };
+    let (h, m) = rest.split_once(':')?;
+    let (h, m): (i32, i32) = (h.parse().ok()?, m.parse().ok()?);
+    if !(0..=14).contains(&h) || !(0..=59).contains(&m) {
+        return None;
+    }
+    Some(sign * (h * 60 + m))
 }
 
 /// TRACES: FR-RAW-3 | FR-EXP-9
@@ -505,6 +622,63 @@ fn cfa_from_rawler(cfa: &rawler::CFA, model: &str) -> CfaPattern {
 
 #[cfg(test)]
 mod tests {
+    use super::{parse_exif_datetime, parse_exif_offset};
+
+    #[test]
+    fn exif_datetime_uses_colon_separated_dates() {
+        // The format that defeats generic parsers: colons in the date.
+        // Checked against a reference implementation, not computed by hand.
+        assert_eq!(
+            parse_exif_datetime("2026:08:09 14:30:00"),
+            Some(1_786_285_800)
+        );
+        assert_eq!(parse_exif_datetime("1970:01:01 00:00:00"), Some(0));
+    }
+
+    #[test]
+    fn fractional_seconds_are_tolerated() {
+        assert_eq!(
+            parse_exif_datetime("2026:08:09 14:30:00.75"),
+            parse_exif_datetime("2026:08:09 14:30:00")
+        );
+    }
+
+    #[test]
+    fn a_malformed_datetime_is_none_rather_than_a_guess() {
+        // A wrong timestamp puts an image at the wrong place on the timeline,
+        // which is worse than leaving it unplaced.
+        assert_eq!(parse_exif_datetime(""), None);
+        assert_eq!(parse_exif_datetime("not a date"), None);
+        // Dashes and slashes are accepted: real devices write both, and
+        // rejecting them left every CanoScan-scanned frame undated.
+        assert_eq!(
+            parse_exif_datetime("2026-08-09 14:30:00"),
+            parse_exif_datetime("2026:08:09 14:30:00")
+        );
+        assert_eq!(
+            parse_exif_datetime("2013/06/28 23:32:54"),
+            parse_exif_datetime("2013:06:28 23:32:54")
+        );
+        assert_eq!(parse_exif_datetime("2026:13:09 14:30:00"), None, "month 13");
+        assert_eq!(parse_exif_datetime("2026:08:09 25:00:00"), None, "hour 25");
+        assert_eq!(parse_exif_datetime("0000:00:00 00:00:00"), None);
+    }
+
+    #[test]
+    fn exif_offsets_parse_both_signs() {
+        assert_eq!(parse_exif_offset("+02:00"), Some(120));
+        assert_eq!(parse_exif_offset("-05:30"), Some(-330));
+        assert_eq!(parse_exif_offset("+00:00"), Some(0));
+    }
+
+    #[test]
+    fn an_absent_or_malformed_offset_is_none() {
+        // Most bodies before ~2018 record no zone at all.
+        assert_eq!(parse_exif_offset(""), None);
+        assert_eq!(parse_exif_offset("02:00"), None, "no sign");
+        assert_eq!(parse_exif_offset("+99:00"), None);
+    }
+
     use super::*;
 
     #[test]

core/dr-gpu/src/adjust.rs

@@ -223,7 +223,11 @@ impl AdjustPass {
         uniforms[0..4].copy_from_slice(&[m[0], m[1], m[2], 0.0]);
         uniforms[4..8].copy_from_slice(&[m[3], m[4], m[5], 0.0]);
         uniforms[8..12].copy_from_slice(&[m[6], m[7], m[8], 0.0]);
-        uniforms[12..16].copy_from_slice(&[wb[0], wb[1], wb[2], 0.0]);
+        // The fourth slot is the non-linear flag, not padding: it tells the
+        // shader whether to linearise the sampled texel before any operation
+        // runs. See `DemosaicedImage::is_non_linear`.
+        let non_linear = if source.is_non_linear() { 1.0 } else { 0.0 };
+        uniforms[12..16].copy_from_slice(&[wb[0], wb[1], wb[2], non_linear]);
 
         let params_buf = self
             .ctx
@@ -615,6 +619,74 @@ mod tests {
         );
     }
 
+    #[test]
+    fn zooming_shows_only_the_region_looked_at() {
+        // Zoom is a coordinate map, and a map that type-checks can still
+        // sample the wrong place. Checked against content: zoomed into the
+        // bright half the frame must be bright edge to edge, and into the
+        // dark half, dark — which a wrong origin or extent would break.
+        let Some(ctx) = ctx() else { return };
+        let mut pass = AdjustPass::new(&ctx);
+        let img = split_image(&ctx, false);
+
+        let mut g = EditGraph::default_chain();
+        g.framing_mut().set_view(dr_pipeline::CropRect {
+            x: 0.0,
+            y: 0.4,
+            width: 0.2,
+            height: 0.2,
+        });
+        let tex = pass.render(&img, &g.compose(), 32, 32).expect("render");
+        let left_near = read_pixel(&ctx, tex, 4, 16)[0];
+        let left_far = read_pixel(&ctx, tex, 28, 16)[0];
+
+        g.framing_mut().set_view(dr_pipeline::CropRect {
+            x: 0.8,
+            y: 0.4,
+            width: 0.2,
+            height: 0.2,
+        });
+        let tex = pass.render(&img, &g.compose(), 32, 32).expect("render");
+        let right_near = read_pixel(&ctx, tex, 4, 16)[0];
+
+        assert!(
+            left_far > 100 && left_near > 100,
+            "zoomed into the bright half, both edges should be bright: \
+             near={left_near} far={left_far}"
+        );
+        assert!(
+            left_near > right_near + 40,
+            "zooming to the far side should show the dark half: \
+             left={left_near} right={right_near}"
+        );
+    }
+
+    #[test]
+    fn zooming_does_not_recompile() {
+        // The property that makes scroll-wheel zoom smooth: a new zoom level
+        // is a uniform upload, never a pipeline build. If zoom reached the
+        // structure hash, every wheel notch would stall on a shader compile.
+        let Some(ctx) = ctx() else { return };
+        let mut pass = AdjustPass::new(&ctx);
+        let img = split_image(&ctx, false);
+
+        let mut g = EditGraph::default_chain();
+        for (i, extent) in [1.0f32, 0.5, 0.25, 0.125].iter().enumerate() {
+            g.framing_mut().set_view(dr_pipeline::CropRect {
+                x: 0.0,
+                y: 0.0,
+                width: *extent,
+                height: *extent,
+            });
+            pass.render(&img, &g.compose(), 32, 32).expect("render");
+            assert_eq!(
+                pass.cached_pipelines(),
+                1,
+                "zoom step {i} compiled a second pipeline"
+            );
+        }
+    }
+
     #[test]
     fn cropping_to_one_half_shows_only_that_half() {
         // The property a crop exists for, checked against content rather than
@@ -946,4 +1018,141 @@ mod tests {
         let t = pass.render(&img, &shader, 32, 96).expect("render");
         assert_eq!((t.width(), t.height()), (32, 96));
     }
+
+    /// A flat RGBA8 image on the JPEG path — already gamma-encoded, as a
+    /// decoded JPEG is.
+    fn jpeg_image(ctx: &GpuContext, rgb: [u8; 3]) -> DemosaicedImage {
+        let size = 16u32;
+        let mut data = Vec::with_capacity((size * size) as usize * 4);
+        for _ in 0..size * size {
+            data.extend_from_slice(&[rgb[0], rgb[1], rgb[2], 255]);
+        }
+        DemosaicedImage::from_rgba8(ctx, &data, size, size).expect("upload")
+    }
+
+    #[test]
+    fn a_jpeg_survives_a_neutral_graph_unchanged() {
+        // The property the whole JPEG path rests on: decoding the transfer
+        // function on the way in and re-encoding on the way out must be exact
+        // inverses. If they are not, merely *opening* a JPEG in develop mode
+        // shifts its tones — the file would be altered by being looked at,
+        // which is far worse than the panel being disabled.
+        let Some(ctx) = ctx() else { return };
+        let mut pass = AdjustPass::new(&ctx);
+        let shader = EditGraph::default_chain().compose();
+
+        // Several levels: a transfer-function error is smallest in the
+        // mid-tones and largest near the ends, so one sample could miss it.
+        for level in [16u8, 64, 128, 200, 240] {
+            let img = jpeg_image(&ctx, [level, level, level]);
+            let t = pass.render(&img, &shader, 16, 16).expect("render");
+            let got = read_centre(&ctx, t);
+            for (i, c) in got[..3].iter().enumerate() {
+                let delta = (i32::from(*c) - i32::from(level)).abs();
+                assert!(
+                    delta <= 2,
+                    "channel {i} at level {level} came back {c} (delta {delta}) \
+                     — the transfer functions are not inverses"
+                );
+            }
+        }
+    }
+
+    #[test]
+    fn a_jpeg_keeps_its_colour_through_a_neutral_graph() {
+        // Identity colour matrix and neutral white balance, specifically: a
+        // camera matrix applied to an image already in sRGB primaries would
+        // skew colour, and this is what catches it. A grey patch cannot —
+        // every matrix maps neutral to neutral.
+        let Some(ctx) = ctx() else { return };
+        let mut pass = AdjustPass::new(&ctx);
+        let shader = EditGraph::default_chain().compose();
+
+        let img = jpeg_image(&ctx, [200, 90, 40]);
+        let t = pass.render(&img, &shader, 16, 16).expect("render");
+        let got = read_centre(&ctx, t);
+
+        for (i, expected) in [200u8, 90, 40].iter().enumerate() {
+            let delta = (i32::from(got[i]) - i32::from(*expected)).abs();
+            assert!(
+                delta <= 2,
+                "channel {i} expected ~{expected}, got {} — colour is being \
+                 transformed on a source that needs no transform",
+                got[i]
+            );
+        }
+    }
+
+    #[test]
+    fn exposure_brightens_a_jpeg() {
+        // Proves the operations reach the JPEG path at all, and that they act
+        // on linearised values: an exposure stop is a multiply, which is only
+        // meaningful once the gamma encoding is undone.
+        let Some(ctx) = ctx() else { return };
+        let mut pass = AdjustPass::new(&ctx);
+        let img = jpeg_image(&ctx, [110, 110, 110]);
+
+        let neutral = EditGraph::default_chain().compose();
+        let before = {
+            let t = pass.render(&img, &neutral, 16, 16).expect("render");
+            read_centre(&ctx, t)
+        };
+
+        let mut g = EditGraph::default_chain();
+        g.set_param(exposure::ID, exposure::EXPOSURE, 1.0);
+        let brighter = g.compose();
+        let after = {
+            let t = pass.render(&img, &brighter, 16, 16).expect("render");
+            read_centre(&ctx, t)
+        };
+
+        assert!(
+            after[0] > before[0],
+            "+1 stop should brighten a JPEG: {before:?} -> {after:?}"
+        );
+
+        // One stop on a linear value is a doubling, which after re-encoding
+        // lands near 1.5x the encoded value rather than 2x. Checking the
+        // magnitude is what distinguishes "linearised correctly" from
+        // "doubled the gamma-encoded value", which would blow straight to
+        // white — the exact bug a brightness-only assertion would miss.
+        assert!(
+            after[0] < 255,
+            "a stop from mid-grey must not clip: {} — the encoding was \
+             probably not undone before the multiply",
+            after[0]
+        );
+    }
+
+    #[test]
+    fn a_jpeg_and_sensor_data_agree_on_the_same_scene_value() {
+        // The two producers must be interchangeable. A mid-grey that is
+        // linearly 0.216 (sRGB 128) arriving as sensor data and as a JPEG
+        // must render the same, or an edit would mean different things
+        // depending on which decoder opened the file.
+        let Some(ctx) = ctx() else { return };
+        let mut pass = AdjustPass::new(&ctx);
+        let shader = EditGraph::default_chain().compose();
+
+        // sRGB 128 linearises to ~0.2159; against a 16383 white level that is
+        // sample ~3537.
+        let sensor = grey_image(&ctx, 3537);
+        let jpeg = jpeg_image(&ctx, [128, 128, 128]);
+
+        let from_sensor = {
+            let t = pass.render(&sensor, &shader, 16, 16).expect("render");
+            read_centre(&ctx, t)
+        };
+        let from_jpeg = {
+            let t = pass.render(&jpeg, &shader, 16, 16).expect("render");
+            read_centre(&ctx, t)
+        };
+
+        let delta = (i32::from(from_sensor[0]) - i32::from(from_jpeg[0])).abs();
+        assert!(
+            delta <= 3,
+            "the same scene value rendered {from_sensor:?} from sensor data \
+             and {from_jpeg:?} from a JPEG"
+        );
+    }
 }

core/dr-gpu/src/demosaic.rs

@@ -33,9 +33,21 @@ struct DemosaicParams {
 
 /// A demosaiced image living on the GPU.
 ///
-/// Linear, scene-referred, camera colour space, RGBA16Float. This is the
+/// RGBA16Float, scene-referred, camera colour space. This is the input every
-/// input every adjustment operates on, and the reason the ops need no
+/// adjustment operates on, and the reason the ops need no knowledge of sensors
-/// knowledge of sensors or CFA patterns.
+/// or CFA patterns.
+///
+/// **Two producers, not one.** [`Demosaicer::run`] builds it from CFA sensor
+/// data; [`DemosaicedImage::from_rgba8`] builds it from an already-processed
+/// RGB image such as a JPEG. Nothing about the type is CFA-specific — it is
+/// simply "an image on the GPU, ready to adjust" — which is what lets develop
+/// mode work on a JPEG without the edit graph or any operation knowing that
+/// the source was not a RAW file.
+///
+/// The one thing that does differ is the transfer function: sensor data is
+/// linear, a JPEG is gamma-encoded. That difference is carried by
+/// [`Self::is_non_linear`] and resolved once, in the generated shader's
+/// prologue, rather than being defended against by every operation.
 pub struct DemosaicedImage {
     texture: wgpu::Texture,
     view: wgpu::TextureView,
@@ -45,6 +57,8 @@ pub struct DemosaicedImage {
     color_matrix: [f32; 9],
     /// As-shot white balance, the neutral starting point for the WB control.
     as_shot_wb: [f32; 3],
+    /// Whether the texture holds gamma-encoded rather than linear values.
+    non_linear: bool,
 }
 
 impl DemosaicedImage {
@@ -75,6 +89,137 @@ impl DemosaicedImage {
     pub fn as_shot_wb(&self) -> [f32; 3] {
         self.as_shot_wb
     }
+
+    /// The largest edge this device can hold in one texture.
+    ///
+    /// Exposed because it is a *hardware* limit the caller has to plan around,
+    /// not a failure to report after the fact: a 13728×8928 film scan exceeds
+    /// the common 8192 limit, and the only way to develop it at all is to fit
+    /// it first. 8192 is still four times a 4K display's long edge, so nothing
+    /// visible is lost.
+    pub fn max_dimension(ctx: &GpuContext) -> u32 {
+        ctx.device.limits().max_texture_dimension_2d
+    }
+
+    /// Whether the texture is gamma-encoded rather than linear.
+    ///
+    /// True for a source that arrived already display-encoded — a JPEG. The
+    /// adjust pass forwards this to the shader, which linearises before any
+    /// operation runs, so the ops themselves always see linear colour.
+    pub fn is_non_linear(&self) -> bool {
+        self.non_linear
+    }
+
+    /// Build one from an already-processed RGBA8 image, skipping demosaic.
+    ///
+    /// The path a JPEG takes into develop mode (FR-RAW-4). There is no CFA to
+    /// interpolate and no sensor to normalise: the pixels are uploaded as they
+    /// arrived, gamma encoding intact, and flagged so the shader linearises
+    /// them.
+    ///
+    /// The two sensor-derived transforms are deliberately neutral rather than
+    /// absent:
+    ///
+    /// - **Colour matrix identity** — a JPEG is already in sRGB primaries, so
+    ///   there is no camera space to convert out of. Applying a real camera
+    ///   matrix here would be a second, unwanted colour transform.
+    /// - **White balance neutral** — the camera applied its own before writing
+    ///   the file, and it cannot be undone from the encoded pixels. The WB
+    ///   control still works; its neutral position is simply "as the camera
+    ///   left it" rather than "as the sensor recorded it".
+    ///
+    /// `rgba` must be tightly packed, 4 bytes per pixel, `width * height`
+    /// pixels, and must fit [`Self::max_dimension`] — a film scan can easily
+    /// exceed it, so callers downscale first rather than being refused here.
+    pub fn from_rgba8(
+        ctx: &GpuContext,
+        rgba: &[u8],
+        width: u32,
+        height: u32,
+    ) -> Result<Self, GpuError> {
+        let (width, height) = (width.max(1), height.max(1));
+
+        let limits = ctx.device.limits();
+        if width > limits.max_texture_dimension_2d || height > limits.max_texture_dimension_2d {
+            return Err(GpuError::TooLarge(format!(
+                "{width}×{height} exceeds the device limit of {}",
+                limits.max_texture_dimension_2d
+            )));
+        }
+
+        let expected = (width as usize) * (height as usize) * 4;
+        if rgba.len() < expected {
+            return Err(GpuError::TooLarge(format!(
+                "{} bytes is short of the {expected} needed for {width}×{height}",
+                rgba.len()
+            )));
+        }
+
+        // The staging texture is Rgba8Unorm because that is what the bytes
+        // are; the pass below converts into the Rgba16Float the rest of the
+        // pipeline expects. Writing f16 on the CPU instead would cost a
+        // full-image conversion before the upload rather than after it.
+        let half: Vec<u16> = rgba[..expected]
+            .iter()
+            .map(|&b| f32_to_f16_bits(f32::from(b) / 255.0))
+            .collect();
+
+        let texture = ctx.device.create_texture_with_data(
+            &ctx.queue,
+            &wgpu::TextureDescriptor {
+                label: Some("jpeg-source"),
+                size: wgpu::Extent3d {
+                    width,
+                    height,
+                    depth_or_array_layers: 1,
+                },
+                mip_level_count: 1,
+                sample_count: 1,
+                dimension: wgpu::TextureDimension::D2,
+                format: Self::FORMAT,
+                // No STORAGE_BINDING: nothing writes to this one. The adjust
+                // pass samples it, and tests copy from it.
+                usage: wgpu::TextureUsages::TEXTURE_BINDING | wgpu::TextureUsages::COPY_SRC,
+                view_formats: &[],
+            },
+            wgpu::util::TextureDataOrder::LayerMajor,
+            bytemuck::cast_slice(&half),
+        );
+        let view = texture.create_view(&Default::default());
+
+        Ok(Self {
+            texture,
+            view,
+            width,
+            height,
+            color_matrix: IDENTITY_3X3,
+            as_shot_wb: [1.0, 1.0, 1.0],
+            non_linear: true,
+        })
+    }
+}
+
+/// Convert an f32 to IEEE 754 half-precision bits.
+///
+/// Written out rather than pulled in as a dependency: the inputs here are
+/// `0.0..=1.0` from an 8-bit source, which is entirely inside the normal range
+/// of f16, so the subnormal and overflow cases a general converter must handle
+/// cannot arise. The clamp makes that assumption explicit rather than implicit.
+fn f32_to_f16_bits(v: f32) -> u16 {
+    let v = v.clamp(0.0, 1.0);
+    if v == 0.0 {
+        return 0;
+    }
+    let bits = v.to_bits();
+    let exp = ((bits >> 23) & 0xFF) as i32 - 127 + 15;
+    let mantissa = (bits >> 13) & 0x3FF;
+    // v is in 0.0..=1.0, so the exponent cannot overflow f16's range; values
+    // below f16's smallest normal round to zero rather than to a subnormal,
+    // which at 8-bit source precision is a distinction without a difference.
+    if exp <= 0 {
+        return 0;
+    }
+    ((exp as u16) << 10) | mantissa as u16
 }
 
 /// Runs the demosaic pass. Holds the pipeline so repeated images reuse it.
@@ -288,6 +433,10 @@ impl Demosaicer {
             // no colour transform rather than not at all.
             color_matrix: raw.color_matrix.unwrap_or(IDENTITY_3X3),
             as_shot_wb: [raw.wb_coeffs[0], raw.wb_coeffs[1], raw.wb_coeffs[2]],
+            // Sensor data is linear by construction — the demosaic shader
+            // normalises against black and white levels and applies no
+            // transfer function.
+            non_linear: false,
         })
     }
 }

core/dr-pipeline/src/operation.rs

@@ -190,6 +190,9 @@ pub fn compose_with_framing(ops: &[Box<dyn Operation>], framing: &Framing) -> Co
          \x20   cam_to_srgb_1: vec4<f32>,\n\
          \x20   cam_to_srgb_2: vec4<f32>,\n\
          \x20   // As-shot white balance, the neutral point for the WB control.\n\
+         \x20   // `.w` is not padding: it flags a non-linear source (1.0 for a\n\
+         \x20   // gamma-encoded JPEG, 0.0 for demosaiced sensor data), which the\n\
+         \x20   // prologue reads to decide whether to linearise.\n\
          \x20   as_shot_wb: vec4<f32>,\n",
     );
     uniform_values.resize(BASE_UNIFORM_FIELDS, 0.0);
@@ -293,6 +296,19 @@ fn encode_srgb(c: vec3<f32>) -> vec3<f32> {{
     return select(hi, lo, c <= vec3<f32>(0.0031308));
 }}
 
+// The inverse, for sources that arrive already display-encoded.
+//
+// A JPEG is uploaded with its bytes untouched, so its values are gamma-encoded
+// where the demosaicer's are linear. Every operation below assumes linear
+// scene-referred colour — exposure is a multiply, and doubling a gamma-encoded
+// value is not a stop — so the encoding is undone here, once, at the only
+// point where the two source kinds still differ.
+fn decode_srgb(c: vec3<f32>) -> vec3<f32> {{
+    let lo = c / 12.92;
+    let hi = pow((max(c, vec3<f32>(0.04045)) + 0.055) / 1.055, vec3<f32>(2.4));
+    return select(hi, lo, c <= vec3<f32>(0.04045));
+}}
+
 @compute @workgroup_size(8, 8, 1)
 fn main(@builtin(global_invocation_id) gid: vec3<u32>) {{
     let dims = textureDimensions(output);
@@ -301,17 +317,29 @@ fn main(@builtin(global_invocation_id) gid: vec3<u32>) {{
     }}
 
 {prologue}
+    // A non-linear source is already display-encoded; undo that so the
+    // operations below see linear colour whatever the source was.
+    let non_linear = u.as_shot_wb.w > 0.5;
+    if (non_linear) {{
+        c = decode_srgb(c);
+    }}
+
     // As-shot white balance. Applied unconditionally, before any operation,
     // because it is part of *interpreting* the sensor rather than an edit: a
     // Bayer sensor's green photosites collect far more signal than its red
     // and blue, so raw camera-space values are strongly green and no amount
     // of later correction recovers a neutral image from them. The white
     // balance operation, when active, applies its own offset on top of this.
+    //
+    // A non-linear source has already had this applied in-camera; the uniform
+    // is neutral there, so this is a multiply by one rather than a branch.
     c = c * u.as_shot_wb.rgb;
 {body}
     // Camera space -> linear sRGB. Applied after the adjustments so white
     // balance and exposure act on sensor-native values, which is where they
     // are physically meaningful.
+    //
+    // Identity for a non-linear source, which is already in sRGB primaries.
     c = vec3<f32>(
         dot(u.cam_to_srgb_0.rgb, c),
         dot(u.cam_to_srgb_1.rgb, c),