diff --git a/issues/page-global-is-a-decision-nobody-has-made.md b/issues/page-global-is-a-decision-nobody-has-made.md
deleted file mode 100644
index 3e753abec1f..00000000000
--- a/issues/page-global-is-a-decision-nobody-has-made.md
+++ /dev/null
@@ -1,44 +0,0 @@
----
-status: open
-kind: track
-opened: 2026-08-19
----
-
-# There is no `PAGE_GLOBAL` anywhere, and the owner decided: turn it on
-
-**The owner ruled on 2026-08-20: `PGE` goes on**, sequenced into the
-interrupt-distribution/scheduler performance era
-(`issues/every-interrupt-lands-on-the-boot-cpu.md`'s wave), never as a
-drive-by — because the work below is real: every `Owed::discharge` in
-`mm/paging.rs` must answer for global translations, `guard_4k`'s local full
-flush too, the bit joins `control_regs.rs`'s one declaration (never a
-read-modify-write at a use site), and the gain is measured before/after on
-the KVM instrument so the improvement is a number against a number. TCG
-cannot price it (tests/CLAUDE.md: the local guest has no such cache model),
-which is one more reason it rides the measured era.
-
-`CR4.PGE` is absent from `kernel/src/arch/x86_64/control_regs.rs`'s declaration and no
-`PAGE_GLOBAL` exists in `kernel/src/mm/`. With PCID active, every address space
-therefore caches a private copy of the kernel's direct map, and every
-`flush_tlb_all` — which is `INVPCID` all-context — throws away the kernel's own
-translations on every CPU along with the process's.
-
-This is not an oversight to fix silently. 2020+ hardware is not Meltdown-
-affected and this kernel has no KPTI, so global kernel pages are available.
-`control_regs.rs`'s own doctrine — every bit of both registers is decided in one
-place, and the bits left out are as much of the declaration as the bits in —
-means `PGE`'s absence should be *stated* whichever way it goes.
-
-**What the answer would have to revisit.** `kernel/src/arch/x86_64/paging.rs` now derives
-every invalidation from the entry a write replaced and discharges it with
-`INVPCID` type 0 or `INVLPG`. Neither touches a global translation, which is
-sound only because no entry in this kernel is global; the module header says so
-in as many words, and a `PGE` that goes in has to answer for every `discharge`
-there and for `guard_4k`'s local full flush.
-
-This file's earlier heading said TLB invalidation was chosen, not derived; that
-question's other half — an
-invalidation chosen by each caller, an unconditional `invlpg` over a not-present
-entry, a duplicate one on the demand-paging path, and one aimed at the parent's
-PCID while writing a child's tables — is resolved. `git log --follow` on this
-file carries the evidence; `Owed` in `mm/paging.rs` is what it became.
diff --git a/kernel/loom/Cargo.toml b/kernel/loom/Cargo.toml
index 757e1e1b9a9..70bc4a60a1d 100644
--- a/kernel/loom/Cargo.toml
+++ b/kernel/loom/Cargo.toml
@@ -322,4 +322,4 @@ test = false
 # library's tests and the simulator run. A `cfg` name and not a feature, since a
 # feature here is a model control.
 [lints.rust]
-unexpected_cfgs = { level = "warn", check-cfg = ['cfg(feature, values("boot-actuators", "mask-windows", "sched-check", "sched-tripwire", "protocol-port", "placement-ignores-staleness"))'] }
+unexpected_cfgs = { level = "warn", check-cfg = ['cfg(feature, values("boot-actuators", "test-actuators", "mask-windows", "sched-check", "sched-tripwire", "protocol-port", "placement-ignores-staleness"))'] }
diff --git a/kernel/loom/tests/counters_round.rs b/kernel/loom/tests/counters_round.rs
index 866407c82bb..5ef886da49b 100644
--- a/kernel/loom/tests/counters_round.rs
+++ b/kernel/loom/tests/counters_round.rs
@@ -20,7 +20,7 @@
 use std::sync::atomic::{AtomicBool, Ordering::SeqCst};
 
 use kernel_loom::seqlock::Published;
-use kernel_loom::shootdown::Shootdown;
+use kernel_loom::shootdown::{Reach, Shootdown};
 use loom::sync::Arc;
 
 /// What the CPU's block held before the round, and what it answers the round with.
@@ -44,11 +44,11 @@ fn a_cpu_read_as_answered_is_read_with_its_answer() {
         let rounds = Arc::new(Shootdown::new());
         let block = Arc::new(Published::new());
         block.publish(BEFORE);
-        let generation = rounds.issue();
+        let generation = rounds.issue(Reach::Narrow);
 
         let cpu1 = {
             let (rounds, block) = (rounds.clone(), block.clone());
-            loom::thread::spawn(move || rounds.serve(1, || block.publish(ANSWER)))
+            loom::thread::spawn(move || rounds.serve(1, |_| block.publish(ANSWER)))
         };
 
         for _ in 0..POLLS {
diff --git a/kernel/loom/tests/tlb_shootdown.rs b/kernel/loom/tests/tlb_shootdown.rs
index 939d6948a96..d956e94e9da 100644
--- a/kernel/loom/tests/tlb_shootdown.rs
+++ b/kernel/loom/tests/tlb_shootdown.rs
@@ -10,6 +10,8 @@
 //! to mean and the only one expressible here. So `UNMAPPED` after an
 //! acknowledged shootdown is "cpu 1 can no longer reach the page the initiator
 //! is about to free", and `MAPPED` is the use-after-free this stage removes.
+//! A kernel-half entry is a second cell, re-walked only by a wide answer, as a
+//! global translation is dropped only by the flush that drops global ones.
 //!
 //! The edge the property rests on is `serve`'s load of what it owes: it
 //! synchronizes with `issue`'s release, and that is what puts the initiator's
@@ -39,7 +41,7 @@
 
 use std::sync::atomic::{AtomicBool, Ordering::SeqCst};
 
-use kernel_loom::shootdown::Shootdown;
+use kernel_loom::shootdown::{Reach, Shootdown};
 use loom::sync::Arc;
 use loom::sync::atomic::{AtomicU64, Ordering::Relaxed};
 
@@ -85,10 +87,14 @@ static ACKED: AtomicBool = AtomicBool::new(false);
 
 struct Machine {
     shootdown: Shootdown,
-    /// The page-table entry the initiator unmaps.
+    /// The user-half page-table entry the initiator unmaps.
     pte: AtomicU64,
-    /// What cpu 1's last page walk found. A flush re-walks.
+    /// What cpu 1's last page walk of it found. Every flush re-walks.
     tlb: AtomicU64,
+    /// The kernel-half entry, whose translation is global.
+    kernel_pte: AtomicU64,
+    /// What cpu 1's last walk of it found. Only a wide flush re-walks.
+    kernel_tlb: AtomicU64,
 }
 
 impl Machine {
@@ -97,28 +103,47 @@ impl Machine {
             shootdown: Shootdown::new(),
             pte: AtomicU64::new(MAPPED),
             tlb: AtomicU64::new(MAPPED),
+            kernel_pte: AtomicU64::new(MAPPED),
+            kernel_tlb: AtomicU64::new(MAPPED),
         })
     }
 
+    /// cpu 1's flush, as far as `reach` asks.
+    fn flush(&self, reach: Reach) {
+        self.tlb.store(self.pte.load(Relaxed), Relaxed);
+        if reach == Reach::Wide {
+            self.kernel_tlb.store(self.kernel_pte.load(Relaxed), Relaxed);
+        }
+    }
+
+    /// One serve of vector 0xFE's body on cpu 1.
+    fn serve(&self) {
+        self.shootdown.serve(1, |reach| self.flush(reach));
+    }
+
     /// cpu 1's side: take the interrupt and run vector 0xFE's body.
     fn take_interrupts(&self) {
         for _ in 0..SERVES {
             loom::thread::yield_now();
-            self.shootdown
-                .serve(1, || self.tlb.store(self.pte.load(Relaxed), Relaxed));
+            self.serve();
         }
     }
 
-    /// The initiator's side, up to the point where it would free the pages.
-    /// `None` when the bounded poll ran out, which is a schedule this model
-    /// says nothing about.
-    fn unmap_and_wait(&self) -> Option<u64> {
-        self.pte.store(UNMAPPED, Relaxed);
-        let generation = self.shootdown.issue();
+    /// The initiator's side, up to the point where it would free the pages:
+    /// `pte` written for a user-half change, `kernel_pte` for a kernel-half
+    /// one, and what cpu 1 holds of it once it has answered. `None` when the
+    /// bounded poll ran out, which is a schedule this model says nothing about.
+    fn unmap_and_wait(&self, reach: Reach) -> Option<u64> {
+        let (pte, tlb) = match reach {
+            Reach::Narrow => (&self.pte, &self.tlb),
+            Reach::Wide => (&self.kernel_pte, &self.kernel_tlb),
+        };
+        pte.store(UNMAPPED, Relaxed);
+        let generation = self.shootdown.issue(reach);
         for _ in 0..POLLS {
             if self.shootdown.served(1, generation) {
                 ACKED.store(true, SeqCst);
-                return Some(self.tlb.load(Relaxed));
+                return Some(tlb.load(Relaxed));
             }
             loom::thread::yield_now();
         }
@@ -138,7 +163,7 @@ fn an_acknowledged_flush_postdates_the_page_table_write() {
             loom::thread::spawn(move || m.take_interrupts())
         };
 
-        if let Some(reachable) = m.unmap_and_wait() {
+        if let Some(reachable) = m.unmap_and_wait(Reach::Narrow) {
             assert_eq!(
                 reachable, UNMAPPED,
                 "cpu 1 acknowledged the shootdown while still holding a \
@@ -174,10 +199,10 @@ fn one_serve_answers_two_concurrent_shootdowns() {
 
         let second = {
             let m = m.clone();
-            loom::thread::spawn(move || m.unmap_and_wait())
+            loom::thread::spawn(move || m.unmap_and_wait(Reach::Narrow))
         };
 
-        if let Some(reachable) = m.unmap_and_wait() {
+        if let Some(reachable) = m.unmap_and_wait(Reach::Narrow) {
             assert_eq!(reachable, UNMAPPED, "the first initiator freed a page cpu 1 could still reach");
         }
         if let Some(reachable) = second.join().unwrap() {
@@ -213,15 +238,15 @@ fn an_initiator_answers_while_it_waits() {
 
         // cpu 0 unmaps and answers for itself; cpu 1 has not issued yet, so the
         // generation cpu 0 publishes is its own.
-        let g0 = s.issue();
-        s.serve(0, || {});
+        let g0 = s.issue(Reach::Narrow);
+        s.serve(0, |_| {});
 
         // cpu 1 unmaps while cpu 0 is between its own `serve` and its wait.
-        let g1 = s.issue();
+        let g1 = s.issue(Reach::Narrow);
 
         // cpu 0's wait is not satisfied — cpu 1 has not flushed — and that turn
         // is nevertheless where cpu 1's answer comes from.
-        assert!(!s.wait_turn(0, 1, g0, || {}));
+        assert!(!s.wait_turn(0, 1, g0, |_| {}));
         assert!(
             s.served(0, g1),
             "cpu 0 is waiting for cpu 1 and has stopped answering it, which is \
@@ -229,9 +254,9 @@ fn an_initiator_answers_while_it_waits() {
         );
 
         // And cpu 1's own wait then has an answer to find.
-        s.serve(1, || {});
-        assert!(s.wait_turn(1, 0, g1, || {}));
-        assert!(s.wait_turn(0, 1, g0, || {}));
+        s.serve(1, |_| {});
+        assert!(s.wait_turn(1, 0, g1, |_| {}));
+        assert!(s.wait_turn(0, 1, g0, |_| {}));
     });
 }
 
@@ -245,16 +270,16 @@ fn an_initiator_answers_while_it_waits() {
 fn a_nested_serve_is_not_undone_by_the_one_it_interrupted() {
     model(|| {
         let s = Arc::new(Shootdown::new());
-        let first = s.issue();
+        let first = s.issue(Reach::Narrow);
         let nested = {
             let s = Arc::clone(&s);
             loom::thread::spawn(move || {
-                let later = s.issue();
-                s.serve(1, || {});
+                let later = s.issue(Reach::Narrow);
+                s.serve(1, |_| {});
                 later
             })
         };
-        s.serve(1, || {});
+        s.serve(1, |_| {});
         let later = nested.join().unwrap();
         assert!(s.served(1, first));
         assert!(
@@ -264,3 +289,140 @@ fn a_nested_serve_is_not_undone_by_the_one_it_interrupted() {
         );
     });
 }
+
+/// A kernel-half shootdown acknowledged means cpu 1 has dropped its global
+/// translation since the write.
+///
+/// The edge is `issue`'s count of wide generations, raised before the
+/// generation exists: a serve that owes the generation reads the count, so its
+/// answer is wide. Reds when the count is raised after the generation is
+/// issued, where a serve can owe it and still answer narrow.
+#[test]
+fn an_acknowledged_kernel_half_shootdown_dropped_the_global_translation() {
+    ACKED.store(false, SeqCst);
+    model(|| {
+        let m = Machine::new();
+
+        let target = {
+            let m = m.clone();
+            loom::thread::spawn(move || m.take_interrupts())
+        };
+
+        if let Some(reachable) = m.unmap_and_wait(Reach::Wide) {
+            assert_eq!(
+                reachable, UNMAPPED,
+                "cpu 1 acknowledged a kernel-half shootdown while still holding the \
+                 global translation the initiator changed",
+            );
+        }
+
+        target.join().unwrap();
+    });
+    assert!(
+        ACKED.load(SeqCst),
+        "no interleaving completed the wait, so the assertion never ran",
+    );
+}
+
+/// One serve may answer a user-half and a kernel-half shootdown at once, and
+/// its answer is then the wide one: the user-half initiator's page is gone
+/// either way, and the kernel-half initiator's global translation with it.
+#[test]
+fn one_serve_answering_a_user_half_and_a_kernel_half_shootdown_answers_wide() {
+    ACKED.store(false, SeqCst);
+    model(|| {
+        let m = Machine::new();
+
+        let target = {
+            let m = m.clone();
+            loom::thread::spawn(move || m.take_interrupts())
+        };
+
+        let kernel = {
+            let m = m.clone();
+            loom::thread::spawn(move || m.unmap_and_wait(Reach::Wide))
+        };
+
+        if let Some(reachable) = m.unmap_and_wait(Reach::Narrow) {
+            assert_eq!(reachable, UNMAPPED, "the user-half initiator freed a page cpu 1 could still reach");
+        }
+        if let Some(reachable) = kernel.join().unwrap() {
+            assert_eq!(reachable, UNMAPPED, "the kernel-half initiator's global translation outlived cpu 1's answer");
+        }
+
+        target.join().unwrap();
+    });
+    assert!(
+        ACKED.load(SeqCst),
+        "no interleaving completed a wait, so the assertions never ran",
+    );
+}
+
+/// A serve nested inside a wide one that has not yet flushed answers wide
+/// itself: the record of a wide answer is raised only once its flush is done.
+///
+/// The two serves are two threads on cpu 1, so loom places the nested one at
+/// every point of the outer one. Reds when the record is raised before the
+/// flush, where the nested serve reads it, answers narrow, and publishes the
+/// generation while the global translation is still held.
+#[test]
+fn a_serve_nested_in_an_unfinished_wide_one_answers_wide() {
+    ACKED.store(false, SeqCst);
+    model(|| {
+        let m = Machine::new();
+
+        let outer = {
+            let m = m.clone();
+            loom::thread::spawn(move || m.serve())
+        };
+        let nested = {
+            let m = m.clone();
+            loom::thread::spawn(move || m.serve())
+        };
+
+        if let Some(reachable) = m.unmap_and_wait(Reach::Wide) {
+            assert_eq!(
+                reachable, UNMAPPED,
+                "a nested serve published the kernel-half shootdown while the wide \
+                 answer it nested in had not yet dropped the global translation",
+            );
+        }
+
+        outer.join().unwrap();
+        nested.join().unwrap();
+    });
+    assert!(
+        ACKED.load(SeqCst),
+        "no interleaving completed the wait, so the assertion never ran",
+    );
+}
+
+/// A serve owing only user-half shootdowns answers narrow, before and after a
+/// kernel-half one, which it answers wide once: a user-half change keeps the
+/// kernel's global translations.
+///
+/// **Not an interleaving question**, as `an_initiator_answers_while_it_waits`
+/// is not; reds when a serve answers wide with no wide generation owed.
+#[test]
+fn only_a_kernel_half_shootdown_asks_a_wide_answer() {
+    model(|| {
+        let s = Shootdown::new();
+        let reach_of = |s: &Shootdown| {
+            let mut answered = None;
+            s.serve(1, |reach| answered = Some(reach));
+            answered.expect("serve runs its flush")
+        };
+
+        s.issue(Reach::Narrow);
+        assert_eq!(reach_of(&s), Reach::Narrow, "a user-half shootdown was answered wide");
+        s.issue(Reach::Wide);
+        s.issue(Reach::Narrow);
+        assert_eq!(reach_of(&s), Reach::Wide, "a kernel-half shootdown was answered narrow");
+        s.issue(Reach::Narrow);
+        assert_eq!(
+            reach_of(&s),
+            Reach::Narrow,
+            "a user-half shootdown after an answered kernel-half one was answered wide",
+        );
+    });
+}
diff --git a/kernel/src/arch/aarch64/paging.rs b/kernel/src/arch/aarch64/paging.rs
index d1676cb8690..2117551d051 100644
--- a/kernel/src/arch/aarch64/paging.rs
+++ b/kernel/src/arch/aarch64/paging.rs
@@ -903,3 +903,10 @@ pub fn scanout_whole_pages(phys: u64, size: u64) -> Result<(), &'static str> {
     }
     Ok(())
 }
+
+/// x86-64's stages a global kernel leaf going stale, and this architecture's
+/// kernel leaves are not judged by it: refused.
+#[cfg(feature = "test-actuators")]
+pub fn debug_global_leaf(_how: u64) -> u64 {
+    toyos_abi::syscall::SyscallError::NotSupported.to_u64()
+}
diff --git a/kernel/src/arch/aarch64/tlb.rs b/kernel/src/arch/aarch64/tlb.rs
index b8a028933cc..8d302fc90d7 100644
--- a/kernel/src/arch/aarch64/tlb.rs
+++ b/kernel/src/arch/aarch64/tlb.rs
@@ -12,7 +12,7 @@
 
 use core::sync::atomic::{AtomicU64, Ordering};
 
-use crate::invalidation::Origin;
+use crate::invalidation::{Half, Origin};
 
 /// Issuer-side census, as x86-64's counts it; there is no receiver side.
 static ISSUED: [AtomicU64; Origin::COUNT] = [const { AtomicU64::new(0) }; Origin::COUNT];
@@ -61,7 +61,7 @@ pub(super) fn all(origin: Origin) {
 
 /// Nothing to drop: every caller follows an `AddressSpace` edit whose own
 /// `page` or `asid` already reached every CPU before it returned.
-pub fn shootdown(_origin: Origin) {}
+pub fn shootdown(_origin: Origin, _half: Half) {}
 
 /// Nothing to answer: no CPU waits on another's acknowledgement here.
 pub fn poll() {}
diff --git a/kernel/src/arch/x86_64/control_regs.rs b/kernel/src/arch/x86_64/control_regs.rs
index b8421b16a93..f228321d3ba 100644
--- a/kernel/src/arch/x86_64/control_regs.rs
+++ b/kernel/src/arch/x86_64/control_regs.rs
@@ -1,8 +1,8 @@
 //! What `CR0`, `CR4`, `IA32_EFER`, `XCR0`, `IA32_XSS`, the save form,
 //! `IA32_APIC_BASE` and the performance request hold on every CPU in this
-//! machine. One declaration,
-//! applied by the BSP and every AP and checked on each; nothing else may write
-//! any of them. Each register is written whole: `CR0` and `EFER` are
+//! machine. One declaration, applied by the BSP and every AP and checked on
+//! each; nothing else may write any of them but [`flush_global`], which puts
+//! back the `CR4` it read. Each register is written whole: `CR0` and `EFER` are
 //! constants, `CR4` is required bits plus whatever optional bits this CPU
 //! offers, `XCR0` and the save form are what `kernel::xsave` makes of this
 //! CPU's CPUID, `IA32_XSS` is zero wherever it exists (no supervisor component
@@ -64,6 +64,7 @@ mod cr4 {
     pub const DE: u64 = 1 << 3;
     pub const PAE: u64 = 1 << 5;
     pub const MCE: u64 = 1 << 6;
+    pub const PGE: u64 = 1 << 7;
     pub const OSFXSR: u64 = 1 << 9;
     pub const OSXMMEXCPT: u64 = 1 << 10;
     pub const OSXSAVE: u64 = 1 << 18;
@@ -139,9 +140,10 @@ const ENERGY_PERF_BIAS: u64 = 6;
 pub const CR0: u64 = cr0::PE | cr0::MP | cr0::ET | cr0::NE | cr0::WP | cr0::PG;
 
 /// `CR4` bits every CPU must have. `DE` is zero legacy, not need — this kernel
-/// touches no debug register. `FSGSBASE` is [`CR4_FORBIDDEN`], not here.
+/// touches no debug register. `PGE` makes a kernel leaf global
+/// (`paging::kernel_leaf`). `FSGSBASE` is [`CR4_FORBIDDEN`], not here.
 const CR4_REQUIRED: u64 =
-    cr4::DE | cr4::PAE | cr4::MCE | cr4::OSFXSR | cr4::OSXMMEXCPT | cr4::OSXSAVE;
+    cr4::DE | cr4::PAE | cr4::MCE | cr4::PGE | cr4::OSFXSR | cr4::OSXMMEXCPT | cr4::OSXSAVE;
 
 /// `CR4` bits this kernel takes when the CPU offers them (checked against CPUID first: an undefined bit is `#GP`).
 const CR4_OPTIONAL: u64 = cr4::SMEP | cr4::SMAP | cr4::PCIDE | cr4::UMIP;
@@ -243,10 +245,13 @@ pub fn init(cpu_id: u32) {
     if !skipped(cpu_id) {
         // SAFETY: `write_cr4` faults only on an undefined bit, on clearing `PAE`
         // in long mode, or on `PCIDE` with a nonzero PCID — `declaration` checked
-        // the first two and both callers use PCID 0; `wrmsr` writes [`EFER`], whose
-        // bits `declaration` has just confirmed this CPU defines and whose `LMA`
-        // is the one this CPU, in long mode, already holds.
+        // the first two for both values and both callers use PCID 0; `wrmsr`
+        // writes [`EFER`], whose bits `declaration` has just confirmed this CPU
+        // defines and whose `LMA` is the one this CPU, in long mode, already holds.
         unsafe {
+            // Through `PGE` clear: only a change of `PGE` drops the global
+            // translations a firmware that left it set may still hold.
+            cpu::write_cr4(declared & !cr4::PGE);
             cpu::write_cr4(declared);
             cpu::wrmsr(efer::MSR, EFER);
             // SAFETY: `CR4.OSXSAVE` is set by the write above, and
@@ -374,6 +379,28 @@ pub fn pcid_active() -> bool {
     DECLARED_CR4.load(Ordering::Acquire) & cr4::PCIDE != 0
 }
 
+/// Drops every translation this CPU holds, global ones included (SDM Vol. 3A
+/// §4.10.4.1, APM Vol. 2 §5.5): `PGE` cleared and put back where it is set;
+/// where it is not, nothing is global and a `CR3` reload drops the rest. For a
+/// CPU whose `PCIDE` is clear, which is every CPU before its declaration.
+///
+/// The live register and not the declaration, because it runs before any
+/// declaration exists. No interrupt guard: a nested call leaves `CR4` as it
+/// found it.
+pub fn flush_global() {
+    let live = cpu::read_cr4();
+    // SAFETY: `live` and `live` without `PGE` are both values this CPU holds or
+    // held, which fault nowhere; a `CR3` write of the `CR3` just read is valid.
+    unsafe {
+        if live & cr4::PGE != 0 {
+            cpu::write_cr4(live & !cr4::PGE);
+            cpu::write_cr4(live);
+        } else {
+            cpu::write_cr3(cpu::read_cr3());
+        }
+    }
+}
+
 /// Proof that this machine's declaration carries `PCIDE`, so `INVPCID` is
 /// not `#UD` here. Zero-sized: an `Option<PcidActive>` costs nothing extra.
 /// Never stales: `PCIDE`, once declared, is never cleared.
@@ -390,7 +417,7 @@ impl PcidActive {
 /// against what the BSP said. Recomputed per CPU rather than trusted from
 /// the BSP, so a divergent machine names the CPU instead of faulting blind.
 fn declaration(cpu_id: u32) -> u64 {
-    let have = supported();
+    let have = supported(cpu_id);
     let missing = CR4_REQUIRED & !have;
     assert!(
         missing == 0,
@@ -471,11 +498,12 @@ fn has_xss() -> bool {
 }
 
 /// The `CR4` bits this CPU will accept, as CPUID reports them.
-fn supported() -> u64 {
-    const CPUID_1_EDX: [(u32, u64); 5] = [
+fn supported(cpu_id: u32) -> u64 {
+    const CPUID_1_EDX: [(u32, u64); 6] = [
         (2, cr4::DE),
         (6, cr4::PAE),
         (7, cr4::MCE),
+        (13, cr4::PGE),
         (24, cr4::OSFXSR),
         (25, cr4::OSXMMEXCPT),
     ];
@@ -508,9 +536,15 @@ fn supported() -> u64 {
             have |= flag;
         }
     }
-    // PCID without INVPCID is not worth having: nothing could flush by ASID.
-    if ecx1 & (1 << 17) != 0 && ebx7 & (1 << 10) != 0 {
-        have |= cr4::PCIDE;
+    // PCID without INVPCID is refused: with `PCIDE` set every global discharge
+    // is `INVPCID` type 2 (`paging::Owed::discharge`), Intel's workaround for
+    // ADL063/RPL043's `INVLPG`, and nothing else could flush by PCID.
+    if ecx1 & (1 << 17) != 0 {
+        if ebx7 & (1 << 10) != 0 {
+            have |= cr4::PCIDE;
+        } else {
+            log!("control_regs: cpu{cpu_id} has PCID but no INVPCID: PCIDE refused");
+        }
     }
     have
 }
diff --git a/kernel/src/arch/x86_64/cpu.rs b/kernel/src/arch/x86_64/cpu.rs
index 842cad4caac..2f5dc664c7a 100644
--- a/kernel/src/arch/x86_64/cpu.rs
+++ b/kernel/src/arch/x86_64/cpu.rs
@@ -271,7 +271,7 @@ pub fn invlpg(addr: u64) {
     unsafe { asm!("invlpg [{}]", in(reg) addr, options(nostack)); }
 }
 
-/// Descriptor type for [`invpcid`] (SDM Vol. 3A §4.10.4.1); three variants, not four, because this kernel issues no global-page discard.
+/// Descriptor type for [`invpcid`] (SDM Vol. 3A §4.10.4.1).
 #[derive(Clone, Copy)]
 #[repr(u64)]
 pub enum Invpcid {
@@ -281,6 +281,8 @@ pub enum Invpcid {
     SinglePcid = 1,
     /// Every entry in every PCID, global pages included.
     AllIncludingGlobal = 2,
+    /// Every entry in every PCID, global pages excepted.
+    AllButGlobal = 3,
 }
 
 /// Discard TLB entries by descriptor type; both instruction faults are excluded by the signature ([`Invpcid`], [`PcidActive`]).
diff --git a/kernel/src/arch/x86_64/paging.rs b/kernel/src/arch/x86_64/paging.rs
index 29798570c88..a9e3e28ff8e 100644
--- a/kernel/src/arch/x86_64/paging.rs
+++ b/kernel/src/arch/x86_64/paging.rs
@@ -3,8 +3,9 @@
 // A live-structure write's TLB invalidation is derived from the entry it
 // replaced, never chosen by the caller, and discharges only on this CPU —
 // `arch::tlb::shootdown` is the caller's job for the rest of the machine.
-// No mapping here is global, which is what makes a single-address
-// invalidation (INVPCID or INVLPG) complete.
+// A leaf is global if and only if it is the same in every root that will ever
+// be loaded, which is the kernel half and only [`kernel_leaf`] writes it; a
+// mapping that differs between roots needs a constructor of its own.
 
 use crate::log;
 use alloc::boxed::Box;
@@ -31,6 +32,9 @@ const PAGE_CACHE_DISABLE: u64 = 1 << 4;
 const PAGE_ACCESSED: u64 = 1 << 5;
 const PAGE_DIRTY: u64 = 1 << 6;
 const PAGE_SIZE_BIT: u64 = 1 << 7;
+/// Cached across every `CR3` and every PCID, because `arch::control_regs`
+/// declares `CR4.PGE`; the same bit in a 2 MiB leaf and a 4 KiB one.
+const PAGE_GLOBAL: u64 = 1 << 8;
 /// In a 4 KiB PTE this bit is bit 7 instead, so flags can't cross a split unmoved.
 const PAGE_PAT_2M: u64 = 1 << 12;
 /// Not-executable only because `arch::control_regs` asserts `EFER.NXE` on every CPU.
@@ -42,7 +46,7 @@ const ADDR_MASK_2M: u64 = 0x000F_FFFF_FFE0_0000;
 const TABLE_FLAGS: u64 = PAGE_PRESENT | PAGE_WRITE | PAGE_USER;
 
 impl Prot {
-    /// The permission bits a leaf entry carries; address and cache policy stay the caller's.
+    /// The permission bits a user leaf carries; address and cache policy stay the caller's.
     fn leaf_bits(self) -> u64 {
         let common = PAGE_PRESENT | PAGE_USER;
         match self {
@@ -53,6 +57,13 @@ impl Prot {
     }
 }
 
+/// Every bit of a 2 MiB kernel-half leaf but its address and size: the one
+/// writer of [`PAGE_GLOBAL`]. Writable and executable, because the kernel's
+/// image is in the direct map.
+fn kernel_leaf(cache: CachePolicy) -> u64 {
+    PAGE_PRESENT | PAGE_WRITE | PAGE_GLOBAL | cache.pde_bits()
+}
+
 /// Which PAT entry a 2 MiB mapping selects: `Normal` is entry 0 (WB), whose
 /// type for a range is the MTRR's (SDM Vol. 3A Table 11-7); `Uncacheable` is
 /// [`pat::UC_ENTRY`](crate::arch::pat::UC_ENTRY), UC under every MTRR type;
@@ -144,7 +155,7 @@ impl PageTablePage {
     fn write_pde(&mut self, idx: usize, va: u64, value: u64) -> Owed {
         let prior = core::mem::replace(&mut self.0[idx], value);
         if prior & PAGE_PRESENT != 0 && prior & PAGE_SIZE_BIT == 0 {
-            Owed::Context
+            Owed::Context { va }
         } else {
             Owed::of(prior, va)
         }
@@ -152,10 +163,10 @@ impl PageTablePage {
 
     /// Owes `Context` only if the widen actually changed a bit — a stale
     /// narrower entry can raise a spurious fault.
-    fn widen(&mut self, idx: usize, flags: u64) -> Owed {
+    fn widen(&mut self, idx: usize, va: u64, flags: u64) -> Owed {
         let before = self.0[idx];
         self.0[idx] = before | flags;
-        if self.0[idx] == before { Owed::Nothing } else { Owed::Context }
+        if self.0[idx] == before { Owed::Nothing } else { Owed::Context { va } }
     }
 
     /// Not yet linked into a live structure, so nothing can be stale.
@@ -175,8 +186,8 @@ enum Owed {
     /// One linear address, whose leaf changed under it.
     Address { va: u64, prior: u64 },
     /// Everything under an upper-level entry that widened, or a PDE that
-    /// stopped naming its page table.
-    Context,
+    /// stopped naming its page table; `va` is an address under it.
+    Context { va: u64 },
 }
 
 impl Owed {
@@ -193,10 +204,19 @@ impl Owed {
     /// `INVPCID` names its tag directly (SDM Vol. 3A §4.10.4.1), so writing a
     /// child's tables invalidates only the child; without PCID, a CPU holds
     /// entries only for its loaded CR3, so nothing is owed unless `target`
-    /// is that CR3.
+    /// is that CR3. A global translation is the exception to both: it is held
+    /// under every CR3 and every PCID, and neither `INVPCID` type 0 or 1 nor
+    /// a CR3 write drops it.
     fn discharge(self, target: Cr3) {
         match self {
             Self::Nothing => {}
+            // With PCID, `INVPCID` type 2: Intel's ADL063 and RPL043 let an
+            // `INVLPG` keep the global translation in every PCID but the current
+            // one, and name type 2 "instead of INVLPG" as the software workaround.
+            Self::Address { va, prior } if prior & PAGE_GLOBAL != 0 => match pcid_active() {
+                Some(have) => crate::arch::cpu::invpcid(have, Invpcid::AllIncludingGlobal, 0, 0),
+                None => crate::arch::cpu::invlpg(va),
+            },
             Self::Address { va, .. } => {
                 if let Some(have) = pcid_active() {
                     crate::arch::cpu::invpcid(have, Invpcid::Address, target.pcid(), va);
@@ -204,11 +224,13 @@ impl Owed {
                     crate::arch::cpu::invlpg(va);
                 }
             }
-            Self::Context => {
+            // Every kernel-half leaf is global.
+            Self::Context { va } if crate::mm::is_kernel_addr(va) => flush_tlb_all(),
+            Self::Context { .. } => {
                 if let Some(have) = pcid_active() {
                     crate::arch::cpu::invpcid(have, Invpcid::SinglePcid, target.pcid(), 0);
                 } else if Cr3::current().phys() == target.phys() {
-                    flush_tlb_all();
+                    flush_tlb_user();
                 }
             }
         }
@@ -222,7 +244,7 @@ impl Owed {
             Self::Address { va, prior } => {
                 panic!("{what}: an install at {va:#x} found the present entry {prior:#x}")
             }
-            Self::Context => panic!("{what}: an install widened an upper-level entry"),
+            Self::Context { .. } => panic!("{what}: an install widened an upper-level entry"),
         }
     }
 
@@ -256,18 +278,25 @@ pub fn pcid_active() -> Option<PcidActive> {
     crate::arch::control_regs::PcidActive::ask()
 }
 
-/// Flush all TLB entries on this CPU, all PCIDs.
+/// Flush all TLB entries on this CPU, all PCIDs, global ones included: what a
+/// kernel-half change owes.
 pub fn flush_tlb_all() {
     if let Some(have) = pcid_active() {
         crate::arch::cpu::invpcid(have, Invpcid::AllIncludingGlobal, 0, 0);
     } else {
-        // SAFETY: the value is exactly what `read_cr3` just read on this
-        // CPU, so writing it back is a valid CR3 write that only flushes
-        // the TLB (there is no `PAGE_GLOBAL` in this kernel).
-        unsafe {
-            let cr3 = crate::arch::cpu::read_cr3();
-            crate::arch::cpu::write_cr3(cr3);
-        }
+        crate::arch::control_regs::flush_global();
+    }
+}
+
+/// Flush every TLB entry on this CPU but the global ones, all PCIDs: what a
+/// user-half change owes, since no user leaf is global.
+pub fn flush_tlb_user() {
+    if let Some(have) = pcid_active() {
+        crate::arch::cpu::invpcid(have, Invpcid::AllButGlobal, 0, 0);
+    } else {
+        // SAFETY: the value is exactly what `read_cr3` just read on this CPU,
+        // so writing it back is a valid CR3 write that only flushes the TLB.
+        unsafe { crate::arch::cpu::write_cr3(crate::arch::cpu::read_cr3()) };
     }
 }
 
@@ -351,7 +380,7 @@ fn alloc_pcid() -> Option<PcidGuard> {
         match pool.alloc() {
             Alloc::Ready(p) => return Some(PcidGuard(p)),
             Alloc::NeedsFlush => {
-                crate::arch::tlb::shootdown(crate::invalidation::Origin::Pcid);
+                crate::arch::tlb::shootdown(crate::invalidation::Origin::Pcid, crate::invalidation::Half::User);
                 pool.reclaim();
             }
             Alloc::Exhausted => return None,
@@ -667,7 +696,7 @@ impl AddressSpace {
         let end = (phys + size + PAGE_2M - 1) & !(PAGE_2M - 1);
         let mut cur = start;
         while cur < end {
-            self.map_2m(cur, PAGE_PRESENT | PAGE_WRITE | cache.pde_bits());
+            self.map_2m(cur, kernel_leaf(cache));
             cur += PAGE_2M;
         }
         crate::mm::Mmio::new(crate::mm::DirectMap::from_phys(phys), size)
@@ -720,6 +749,11 @@ impl AddressSpace {
         let pd = unsafe { PageTablePage::from_phys_mut(pd_phys) };
         let pde = pd[pd_idx];
         assert!(pde & PAGE_PRESENT != 0, "guard_4k: {phys:#x} is not in the direct map");
+        // Both writes discharge on this CPU alone: the guard belongs to one
+        // CPU, so only it dereferences the removed page; `alloc_idle_stack`
+        // runs on the BSP before an AP's TLB exists, and a sibling's stale
+        // 2 MiB entry stays correct for the other 511 pages it still maps.
+        let target = self.root();
 
         // Already split: an earlier guard in the same 2 MiB region.
         if pde & PAGE_SIZE_BIT != 0 {
@@ -735,9 +769,7 @@ impl AddressSpace {
             }
             let pt_phys = pt.phys();
             self.children.push(pt);
-            // Covered by the flush below, wider than either write could owe.
-            pd.write_pde(pd_idx, virt, pt_phys | PAGE_PRESENT | PAGE_WRITE)
-                .subsumed_by_flush();
+            pd.write_pde(pd_idx, virt, pt_phys | PAGE_PRESENT | PAGE_WRITE).discharge(target);
         }
 
         // SAFETY: `pd[pd_idx]` names a page table either way, permanently —
@@ -745,13 +777,7 @@ impl AddressSpace {
         let pt = unsafe { PageTablePage::from_phys_mut(pd[pd_idx] & ADDR_MASK) };
         let idx = ((phys >> 12) & 0x1FF) as usize;
         assert!(pt[idx] & PAGE_PRESENT != 0, "guard_4k: {phys:#x} is already unmapped");
-        pt.write(idx, virt, 0).subsumed_by_flush();
-
-        // Local only: the guard belongs to one CPU, so only it dereferences
-        // the removed page; `alloc_idle_stack` runs on the BSP before an
-        // AP's TLB exists, and a sibling's stale 2 MiB entry stays correct
-        // for the other 511 pages it still maps unchanged.
-        flush_tlb_all();
+        pt.write(idx, virt, 0).discharge(target);
     }
 
     /// Replaces whatever is there; a page `guard_4k` already split must not
@@ -789,7 +815,7 @@ impl AddressSpace {
             self.children.push(child);
         } else {
             // Upper-level entries: widen only (OR), never narrow.
-            self.root.widen(pml4_idx, flags).discharge(target);
+            self.root.widen(pml4_idx, va, flags).discharge(target);
         }
 
         // SAFETY: the branch above guaranteed PRESENT, so `& ADDR_MASK`
@@ -803,7 +829,7 @@ impl AddressSpace {
                 .expect_install("ensure_table: pdpt");
             self.children.push(child);
         } else {
-            pdpt.widen(pdpt_idx, flags).discharge(target);
+            pdpt.widen(pdpt_idx, va, flags).discharge(target);
         }
 
         // SAFETY: same argument as `pdpt` above, one level down.
@@ -891,7 +917,7 @@ pub fn map_mmio(phys: u64, size: u64, policy: MmioPolicy) -> crate::mm::Mmio {
             driven.push((phys, phys + size));
         }
     }
-    crate::arch::tlb::shootdown(crate::invalidation::Origin::Mmio);
+    crate::arch::tlb::shootdown(crate::invalidation::Origin::Mmio, crate::invalidation::Half::Kernel);
     // Read back off the table and logged beside firmware's MTRR verdict: the
     // boot's own evidence that no register window trusts firmware.
     let installed =
@@ -904,6 +930,95 @@ pub fn map_mmio(phys: u64, size: u64, policy: MmioPolicy) -> crate::mm::Mmio {
     mmio
 }
 
+/// Where [`debug_global_leaf`] maps: the kernel half's last 2 MiB, under a root
+/// slot [`seal_kernel_half`] installed, as far past the direct map as it goes.
+#[cfg(feature = "test-actuators")]
+const SCRATCH: u64 = 0u64.wrapping_sub(PAGE_2M);
+
+/// `SYS_DEBUG`'s `GLOBAL_LEAF_REWRITE`: a [`kernel_leaf`] at [`SCRATCH`] over one
+/// frame, read so this CPU caches it global; the leaf pointed at a second frame
+/// and read again with nothing invalidated, the control; invalidated as `how`
+/// asks; read a third time. Answers which frame the third read saw, or refuses:
+/// where the control already saw the second frame, or where a user-half
+/// shootdown's answer here reached wide because another CPU's kernel-half
+/// generation landed inside it. The kernel's lock keeps it on one CPU and
+/// interrupts stay closed, so no flush but `how`'s lands between the reads.
+#[cfg(feature = "test-actuators")]
+pub fn debug_global_leaf(how: u64) -> u64 {
+    use crate::invalidation::{Half, Origin};
+    use toyos_abi::syscall::debug_action as da;
+    use toyos_abi::syscall::SyscallError;
+
+    enum How {
+        Discharged,
+        ShotDown(Half),
+    }
+    let how = match how {
+        da::GLOBAL_LEAF_DISCHARGED => How::Discharged,
+        da::GLOBAL_LEAF_KERNEL_SHOT_DOWN => How::ShotDown(Half::Kernel),
+        da::GLOBAL_LEAF_USER_SHOT_DOWN => How::ShotDown(Half::User),
+        _ => return SyscallError::InvalidArgument.to_u64(),
+    };
+    let (Some(old), Some(new)) = (crate::mm::pmm::alloc_page(), crate::mm::pmm::alloc_page()) else {
+        return SyscallError::ResourceExhausted.to_u64();
+    };
+    // Each frame's first word is its own address, so a read names its frame.
+    for page in [&old, &new] {
+        let at = page.direct_map();
+        // SAFETY: a frame this call owns, through the direct map.
+        unsafe { core::ptr::write_volatile(at.as_mut_ptr::<u64>(), at.phys()) };
+    }
+    let (old_phys, new_phys) = (old.direct_map().phys(), new.direct_map().phys());
+    let leaf = kernel_leaf(CachePolicy::Normal) | PAGE_SIZE_BIT;
+    // SAFETY: every read below is made while `SCRATCH` maps one of the two
+    // frames this call owns, or a translation of one of them.
+    let read = || unsafe { core::ptr::read_volatile(SCRATCH as *const u64) };
+
+    let (held, widened, seen) = {
+        let mut space = kernel().lock();
+        let _closed = crate::arch::IrqGuard::close();
+        let target = space.root();
+        let pd = space.ensure_table(SCRATCH, leaf);
+        let idx = indices(SCRATCH).2;
+        if pd[idx] & PAGE_PRESENT != 0 {
+            return SyscallError::AlreadyExists.to_u64();
+        }
+        pd.write_pde(idx, SCRATCH, old_phys | leaf).expect_install("debug_global_leaf");
+        let first = read();
+        assert!(first == old_phys, "debug_global_leaf: a fresh leaf over {old_phys:#x} read {first:#x}");
+        let owed = pd.write_pde(idx, SCRATCH, new_phys | leaf);
+        let held = read();
+        let widened = match how {
+            How::Discharged => {
+                owed.discharge(target);
+                false
+            }
+            // The write's own invalidation is withheld: the arm asks what the shootdown drops.
+            How::ShotDown(half) => {
+                let _ = owed;
+                crate::arch::tlb::debug_shootdown_answered_wide(Origin::Staged, half) && matches!(half, Half::User)
+            }
+        };
+        let seen = read();
+        pd.write_pde(idx, SCRATCH, 0).discharge(target);
+        (held, widened, seen)
+    };
+    // No CPU keeps a translation of either frame past its free.
+    crate::arch::tlb::shootdown(Origin::Staged, Half::Kernel);
+    drop((old, new));
+    let old_frame = |at: u64| match at {
+        _ if at == old_phys => true,
+        _ if at == new_phys => false,
+        _ => panic!("debug_global_leaf: the leaf over {old_phys:#x} then {new_phys:#x} read {at:#x}"),
+    };
+    match (old_frame(held), widened, old_frame(seen)) {
+        (false, _, _) => da::GLOBAL_LEAF_UNHELD,
+        (true, true, _) => da::GLOBAL_LEAF_WIDENED,
+        (true, false, true) => da::GLOBAL_LEAF_SAW_OLD,
+        (true, false, false) => da::GLOBAL_LEAF_SAW_NEW,
+    }
+}
+
 /// Build kernel page tables: the direct map in the high half, in 2 MiB pages,
 /// as far as [`toyos_bootmap::x86_64::direct_map_end`] reaches, and answer
 /// that end. The scanout is not mapped for itself: in the low map the MTRRs
@@ -928,7 +1043,7 @@ pub(crate) fn init(memory_map: &[MemoryMapEntry], _scanout: Option<(u64, u64)>)
     }
     let mut addr: u64 = 0;
     while addr < end {
-        kernel.map_2m(addr, PAGE_PRESENT | PAGE_WRITE);
+        kernel.map_2m(addr, kernel_leaf(CachePolicy::Normal));
         addr += PAGE_2M;
     }
     crate::log!("paging: the direct map covers 0x0..{end:#x}");
diff --git a/kernel/src/arch/x86_64/pat.rs b/kernel/src/arch/x86_64/pat.rs
index 71201b75c88..ca0c1323352 100644
--- a/kernel/src/arch/x86_64/pat.rs
+++ b/kernel/src/arch/x86_64/pat.rs
@@ -1,6 +1,7 @@
 //! Programs `IA32_PAT` with one write-combining entry for the GOP scanout;
 //! SDM Vol. 3A Table 11-7 makes an MTRR range register unnecessary.
 
+use crate::arch::control_regs::flush_global;
 use crate::arch::cpu;
 
 const IA32_PAT: u32 = 0x277;
@@ -57,15 +58,14 @@ pub fn init() {
     unsafe {
         core::arch::asm!("pushfq", "pop {}", "cli", out(reg) flags);
         let cr0 = cpu::read_cr0();
-        let cr4 = cpu::read_cr4();
         // CD set with NW clear is no-fill mode: writes still hit, nothing new caches.
         cpu::write_cr0((cr0 | CR0_CD) & !CR0_NW);
         cpu::wbinvd();
-        flush_tlb(cr4);
+        flush_global();
 
         cpu::wrmsr(IA32_PAT, PAT_VALUE);
 
-        flush_tlb(cr4);
+        flush_global();
         cpu::wbinvd();
         cpu::write_cr0(cr0);
     }
@@ -88,21 +88,8 @@ pub fn check() {
 
 const CR0_NW: u64 = 1 << 29;
 const CR0_CD: u64 = 1 << 30;
-const CR4_PGE: u64 = 1 << 7;
 const RFLAGS_IF: u64 = 1 << 9;
 
-/// Flush every TLB entry, including global ones (SDM Vol. 3A §4.10.4.1).
-/// # Safety
-/// `cr4` must be this CPU's live `CR4`; both arms restore it verbatim.
-unsafe fn flush_tlb(cr4: u64) {
-    if cr4 & CR4_PGE != 0 {
-        cpu::write_cr4(cr4 & !CR4_PGE);
-        cpu::write_cr4(cr4);
-    } else {
-        cpu::write_cr3(cpu::read_cr3());
-    }
-}
-
 /// This CPU's live `IA32_PAT`.
 pub fn msr() -> u64 {
     cpu::rdmsr(IA32_PAT)
diff --git a/kernel/src/arch/x86_64/pio.rs b/kernel/src/arch/x86_64/pio.rs
index c27e1d7615b..320d23a3426 100644
--- a/kernel/src/arch/x86_64/pio.rs
+++ b/kernel/src/arch/x86_64/pio.rs
@@ -146,7 +146,7 @@ pub struct TakenBack(());
 /// shootdown returns only once every other CPU has answered it from Ring 0,
 /// so no thread that was in Ring 3 before is there still.
 pub fn take_back(_stopping: &crate::quiesce::Stopping) -> TakenBack {
-    super::tlb::shootdown(crate::invalidation::Origin::Stop);
+    super::tlb::shootdown(crate::invalidation::Origin::Stop, crate::invalidation::Half::User);
     TakenBack(())
 }
 
diff --git a/kernel/src/arch/x86_64/smi_cmd.rs b/kernel/src/arch/x86_64/smi_cmd.rs
index 3f17585892f..c513f05aa4a 100644
--- a/kernel/src/arch/x86_64/smi_cmd.rs
+++ b/kernel/src/arch/x86_64/smi_cmd.rs
@@ -54,7 +54,7 @@ use toyos_userbound::{KeptCommands, Mediated, Ports, Undeclared};
 
 use super::pio::{self, Declared, Slot, TakenBack};
 use super::{apic, cpu, percpu, IrqGuard};
-use crate::shootdown::Shootdown;
+use crate::shootdown::{Reach, Shootdown};
 use crate::sync::Lock;
 use crate::time::{Duration, DEAF_CPU};
 use kernel::bootwrite::{self, Turn};
@@ -143,7 +143,7 @@ pub fn write(value: u8) -> Option<Written> {
     let asked_from = percpu::cpu_id();
     ASKED.store(value, Relaxed);
     TAKEN_NS.store(UNTAKEN, Relaxed);
-    let generation = ROUND.issue();
+    let generation = ROUND.issue(Reach::Narrow);
     // The boot processor's own ask, unless a kick taken since the issue has answered it already.
     answer();
     if !ROUND.served(BOOT as usize, generation) {
@@ -196,7 +196,7 @@ fn answer() {
     if percpu::cpu_id() != BOOT {
         return;
     }
-    ROUND.serve_if_owed(BOOT as usize, || {
+    ROUND.serve_if_owed(BOOT as usize, |_| {
         let port = PORT.get().expect("smi_cmd: a write asked of a machine whose FADT names no SMI_CMD").port(0);
         let value = ASKED.load(Relaxed);
         let smi = || super::counters::read().smi.unwrap_or(UNREAD);
diff --git a/kernel/src/arch/x86_64/tlb.rs b/kernel/src/arch/x86_64/tlb.rs
index 902752dc763..4e7f9d007da 100644
--- a/kernel/src/arch/x86_64/tlb.rs
+++ b/kernel/src/arch/x86_64/tlb.rs
@@ -9,17 +9,22 @@
 //! **The target set is every other CPU the machine brought up**, not the CPUs
 //! sharing the initiator's address space: nothing a process does puts a CPU
 //! into it or takes one out.
+//!
+//! **A user-half change keeps global translations**: its flush is `INVPCID`
+//! type 3 or a `CR3` reload, and only a kernel-half change, whose leaves are
+//! global, asks the flush that drops them (SDM Vol. 3A §4.10.4.1, APM Vol. 2
+//! §5.5.2). An answer covering both is the wider.
 
 use core::sync::atomic::{AtomicU64, Ordering};
 
-use crate::shootdown::{Generation, Shootdown};
+use crate::shootdown::{Generation, Reach, Shootdown};
 
 use super::{apic, percpu};
 use crate::smp;
 
 static SHOOTDOWN: Shootdown = Shootdown::new();
 
-use crate::invalidation::Origin;
+use crate::invalidation::{Half, Origin};
 
 /// Issuer-side census; `irq_census`'s `tlb` column is the receiver side, and a
 /// delivery the two disagree on is an uncounted issuing path.
@@ -58,20 +63,25 @@ pub fn census(say: &mut impl FnMut(core::fmt::Arguments<'_>)) {
 const SPINS_PER_DEADLINE_CHECK: u32 = 1024;
 
 /// Write the page table, then call this, then free — it returns only once every
-/// CPU has flushed. The local-flush early return is uncounted: no IPI, no wait.
-pub fn shootdown(origin: Origin) {
+/// CPU has flushed what a change to `half` owes. The local-flush early return
+/// is uncounted: no IPI, no wait.
+pub fn shootdown(origin: Origin, half: Half) {
+    let reach = match half {
+        Half::User => Reach::Narrow,
+        Half::Kernel => Reach::Wide,
+    };
     let cpus = smp::cpu_count();
     if !smp::answering() || cpus <= 1 {
-        crate::mm::paging::flush_tlb_all();
+        flush(reach);
         return;
     }
     // Counted before the IPI, so a delivery can never precede its issue's count.
     ISSUED[origin as usize].fetch_add(1, Ordering::Relaxed);
     let began = crate::clock::nanos_since_boot();
     let me = percpu::cpu_id() as usize;
-    let generation = SHOOTDOWN.issue();
+    let generation = SHOOTDOWN.issue(reach);
     // This CPU answers itself locally instead of by self-IPI.
-    SHOOTDOWN.serve(me, crate::mm::paging::flush_tlb_all);
+    SHOOTDOWN.serve(me, flush);
     apic::tlb_ipi();
     for cpu in 0..cpus {
         if cpu as usize != me {
@@ -103,7 +113,7 @@ pub fn bench() {
     let mut took = [0u64; ROUNDS];
     for sample in took.iter_mut() {
         let began = crate::clock::nanos_since_boot();
-        shootdown(Origin::Bench);
+        shootdown(Origin::Bench, Half::User);
         *sample = crate::clock::nanos_since_boot().saturating_sub(began);
     }
     took.sort_unstable();
@@ -123,7 +133,7 @@ pub fn bench() {
 fn wait_for(me: usize, cpu: u32, generation: Generation) {
     let mut spins = 0u32;
     let mut deadline = None;
-    while !SHOOTDOWN.wait_turn(me, cpu as usize, generation, crate::mm::paging::flush_tlb_all) {
+    while !SHOOTDOWN.wait_turn(me, cpu as usize, generation, flush) {
         core::hint::spin_loop();
         spins += 1;
         if spins == SPINS_PER_DEADLINE_CHECK {
@@ -142,11 +152,20 @@ fn wait_for(me: usize, cpu: u32, generation: Generation) {
     }
 }
 
+/// What `reach` asks of this CPU: every translation, or every one but the
+/// global ones a user-half change leaves.
+fn flush(reach: Reach) {
+    match reach {
+        Reach::Wide => crate::mm::paging::flush_tlb_all(),
+        Reach::Narrow => crate::mm::paging::flush_tlb_user(),
+    }
+}
+
 /// Vector 0xFE's whole body: flush this CPU and say which generation it covers.
 pub fn serve_ipi() {
     let cpu = percpu::cpu_id() as usize;
-    SHOOTDOWN.serve(cpu, || {
-        crate::mm::paging::flush_tlb_all();
+    SHOOTDOWN.serve(cpu, |reach| {
+        flush(reach);
         stage_ack_delay();
     });
 }
@@ -161,17 +180,18 @@ pub fn poll() {
         return;
     }
     let cpu = percpu::cpu_id() as usize;
-    SHOOTDOWN.serve_if_owed(cpu, || {
-        crate::mm::paging::flush_tlb_all();
+    SHOOTDOWN.serve_if_owed(cpu, |reach| {
+        flush(reach);
         stage_ack_delay();
     });
 }
 
 /// Settle every shootdown issued before this CPU could answer one; called
-/// once after the machine is released.
+/// once after the machine is released. Wide whatever is owed: a kernel-half
+/// change made before the machine answered issued no generation to say so.
 pub fn join() {
     let cpu = percpu::cpu_id() as usize;
-    SHOOTDOWN.serve(cpu, crate::mm::paging::flush_tlb_all);
+    SHOOTDOWN.serve(cpu, |_| crate::mm::paging::flush_tlb_all());
 }
 
 #[cfg(not(feature = "test-actuators"))]
@@ -244,7 +264,7 @@ pub fn debug_arm_ack_delay(nanos: u64) -> u64 {
             delay::TARGET.store(cpu, Ordering::Relaxed);
             open_arm_window();
             let began = crate::clock::nanos_since_boot();
-            shootdown(Origin::Staged);
+            shootdown(Origin::Staged, Half::User);
             least = least.min(crate::clock::nanos_since_boot().saturating_sub(began));
         }
     });
@@ -259,6 +279,16 @@ pub fn debug_arm_ack_delay(nanos: u64) -> u64 {
     }
 }
 
+/// [`shootdown`], answering whether this CPU's own answers inside it reached
+/// wide: a kernel-half generation another CPU issued meanwhile widens them.
+#[cfg(feature = "test-actuators")]
+pub fn debug_shootdown_answered_wide(origin: Origin, half: Half) -> bool {
+    let me = percpu::cpu_id() as usize;
+    let before = SHOOTDOWN.widened(me);
+    shootdown(origin, half);
+    SHOOTDOWN.widened(me) != before
+}
+
 /// Give the machine its ordinary latency back before the window lapses.
 #[cfg(feature = "test-actuators")]
 pub fn debug_disarm_ack_delay() -> u64 {
diff --git a/kernel/src/counters.rs b/kernel/src/counters.rs
index 7aef37d90ec..6017190b963 100644
--- a/kernel/src/counters.rs
+++ b/kernel/src/counters.rs
@@ -25,7 +25,7 @@ use crate::arch::{irqchip, percpu, IrqGuard};
 use crate::irq_census::Source;
 use crate::scheduler::{Parkable, MAX_CPUS};
 use crate::seqlock::Published;
-use crate::shootdown::{Generation, Shootdown};
+use crate::shootdown::{Generation, Reach, Shootdown};
 use crate::sync::Lock;
 use crate::time::{Budget, Cadence, Deadline, Duration, Instant};
 use crate::user_ptr::UserBytesMut;
@@ -151,7 +151,7 @@ fn answer(kick: bool) {
         if deaf::DEAF.load(core::sync::atomic::Ordering::Relaxed) == me {
             return;
         }
-        ROUNDS.serve_if_owed(me, || BLOCKS[me].publish(sample(me)))
+        ROUNDS.serve_if_owed(me, |_| BLOCKS[me].publish(sample(me)))
     };
     if served {
         ANSWERED.post_in_place();
@@ -175,7 +175,7 @@ pub fn read(rights: Rights, out: &mut UserBytesMut) -> Result<usize, SyscallErro
         match *last {
             Some((generation, issued)) if now < issued + JOIN.duration() => (generation, issued, false),
             _ => {
-                let generation = ROUNDS.issue();
+                let generation = ROUNDS.issue(Reach::Narrow);
                 *last = Some((generation, now));
                 (generation, now, true)
             }
diff --git a/kernel/src/invalidation.rs b/kernel/src/invalidation.rs
index bf6511946c7..13bc01b8ab8 100644
--- a/kernel/src/invalidation.rs
+++ b/kernel/src/invalidation.rs
@@ -1,9 +1,10 @@
-//! Why a machine-wide translation invalidation was asked for: the census every
-//! architecture's shootdown (`arch::tlb`) counts its callers by.
+//! Why a machine-wide translation invalidation was asked for, the census every
+//! architecture's shootdown (`arch::tlb`) counts its callers by, and which half
+//! of the address space it answers for.
 
 /// Which path issued a shootdown, so the census names who pays: `Dlopen` (a
 /// `Shared` window or rollback unmap), `Pcid` (pool reclaim), `Mmio`, `Unmap`
-/// (`Unmapped::drop`), `Pipe`, `Staged` (the ack-delay actuator), `Bench`
+/// (`Unmapped::drop`), `Pipe`, `Staged` (a test actuator's), `Bench`
 /// (`arch::tlb::bench`'s own, so a measured shootdown is never counted as one
 /// a path in this kernel needed), `Stop` (the power-off's, for the answer from
 /// Ring 0 every other CPU owes it, not for a flush).
@@ -24,6 +25,17 @@ pub enum Origin {
     Stop,
 }
 
+/// Which half of every address space a shootdown's write changed: x86-64's
+/// kernel-half leaves are global, so only a change there asks a flush that
+/// drops global translations.
+#[derive(Clone, Copy, PartialEq, Eq)]
+pub enum Half {
+    User,
+    // The kernel half is global on x86-64 alone, so AArch64 builds a variant it never issues.
+    #[allow(dead_code)]
+    Kernel,
+}
+
 impl Origin {
     pub const COUNT: usize = 8;
     /// Order matches the variants; `tests/toyos.rs`'s `irq_census_conservation` reads the line back.
diff --git a/kernel/src/mm/unmapped.rs b/kernel/src/mm/unmapped.rs
index 5741e04619c..29b9c9655da 100644
--- a/kernel/src/mm/unmapped.rs
+++ b/kernel/src/mm/unmapped.rs
@@ -16,7 +16,7 @@ impl<T> Unmapped<T> {
 
 impl<T> Drop for Unmapped<T> {
     fn drop(&mut self) {
-        crate::arch::tlb::shootdown(crate::invalidation::Origin::Unmap);
+        crate::arch::tlb::shootdown(crate::invalidation::Origin::Unmap, crate::invalidation::Half::User);
         // SAFETY: the wrapped value is never taken before this drop.
         unsafe { ManuallyDrop::drop(&mut self.0) };
     }
diff --git a/kernel/src/process.rs b/kernel/src/process.rs
index 78bfe36fbe2..2318afee90c 100644
--- a/kernel/src/process.rs
+++ b/kernel/src/process.rs
@@ -661,7 +661,7 @@ pub fn revoke_pipe_maps(maps: &mut Vec<PipeMap>, pt: &PageTables, pipe: pipe::Pi
         });
     }
     // Outside the block: it waits, and a sibling can be spinning on this lock.
-    crate::arch::tlb::shootdown(crate::invalidation::Origin::Pipe);
+    crate::arch::tlb::shootdown(crate::invalidation::Origin::Pipe, crate::invalidation::Half::User);
 }
 
 /// One live `mmap` and its physical pages; the range's registration in the address space's `regions` is separate (placement search, `munmap`).
diff --git a/kernel/src/shootdown.rs b/kernel/src/shootdown.rs
index 068f41b8956..eb3c5f4b3ee 100644
--- a/kernel/src/shootdown.rs
+++ b/kernel/src/shootdown.rs
@@ -3,6 +3,8 @@
 //! Compiled a second time into `kernel/loom/` against loom's atomics, so this file must hold no `crate::` references.
 //! The read must happen before the answer, or a target could publish a generation its answer does not cover;
 //! and an initiator that sees a generation served sees everything that answer wrote.
+//! A generation is narrow or wide, and an answer reaches as far as the widest generation it covers: one
+//! answer covers every generation owed, so the reach is the target's to read and never the IPI's.
 
 #[cfg(not(feature = "loom"))]
 use core::sync::atomic::{AtomicU64, Ordering};
@@ -31,9 +33,20 @@ const SERVED: Ordering = Ordering::Relaxed;
 #[derive(Clone, Copy, PartialEq, Eq, PartialOrd, Ord, Debug)]
 pub struct Generation(u64);
 
+/// How far an answer must reach: what a wide generation asks of it, a narrow one does not.
+#[derive(Clone, Copy, PartialEq, Eq, Debug)]
+pub enum Reach {
+    Narrow,
+    Wide,
+}
+
 pub struct Shootdown {
     requested: AtomicU64,
+    /// Wide generations issued, each counted before it is issued.
+    wide: AtomicU64,
     flushed: [AtomicU64; MAX_CPUS],
+    /// Per CPU, `wide` as the last wide answer it finished read it.
+    widened: [AtomicU64; MAX_CPUS],
 }
 
 impl Shootdown {
@@ -42,7 +55,9 @@ impl Shootdown {
     pub const fn new() -> Self {
         Self {
             requested: AtomicU64::new(0),
+            wide: AtomicU64::new(0),
             flushed: [const { AtomicU64::new(0) }; MAX_CPUS],
+            widened: [const { AtomicU64::new(0) }; MAX_CPUS],
         }
     }
 
@@ -53,24 +68,43 @@ impl Shootdown {
     pub fn new() -> Self {
         Self {
             requested: AtomicU64::new(0),
+            wide: AtomicU64::new(0),
             flushed: core::array::from_fn(|_| AtomicU64::new(0)),
+            widened: core::array::from_fn(|_| AtomicU64::new(0)),
         }
     }
 
-    /// Names the generation every other CPU now owes a flush for.
-    pub fn issue(&self) -> Generation {
+    /// Names the generation every other CPU now owes a flush for, of `reach`.
+    pub fn issue(&self, reach: Reach) -> Generation {
+        if reach == Reach::Wide {
+            // Before the generation exists, and Release: a serve that owes the generation reads this
+            // count, and an answer that reads the count follows the write it was raised for.
+            self.wide.fetch_add(1, Ordering::Release);
+        }
         // AcqRel, not Release: the acquire half stops a later load being hoisted above this fetch_add.
         Generation(self.requested.fetch_add(1, Ordering::AcqRel) + 1)
     }
 
-    /// A target's whole side of the protocol.
-    pub fn serve(&self, cpu: usize, flush: impl FnOnce()) {
+    /// A target's whole side of the protocol: `flush` is told how far its answer must reach.
+    pub fn serve(&self, cpu: usize, flush: impl FnOnce(Reach)) {
         let owed = self.requested.load(OWED);
-        flush();
+        let wide = self.wide.load(Ordering::Acquire);
+        let reach = if wide > self.widened[cpu].load(Ordering::Acquire) { Reach::Wide } else { Reach::Narrow };
+        flush(reach);
+        if reach == Reach::Wide {
+            // After the flush: a serve nested inside this one, reading this record first, answers wide itself.
+            self.widened[cpu].fetch_max(wide, Ordering::Release);
+        }
         // A max, not a store: a serve nested inside this one may already have published a later generation.
         self.flushed[cpu].fetch_max(owed, Ordering::Release);
     }
 
+    /// `wide` as `cpu`'s last wide answer read it: a change across a span is a wide answer inside it.
+    #[cfg(feature = "test-actuators")]
+    pub fn widened(&self, cpu: usize) -> u64 {
+        self.widened[cpu].load(Ordering::Relaxed)
+    }
+
     /// Does `cpu` owe anyone a flush?
     pub fn owes(&self, cpu: usize) -> bool {
         // A hint: `serve` carries the real ordering, so a stale `false` costs nothing.
@@ -83,7 +117,7 @@ impl Shootdown {
     }
 
     /// [`serve`](Self::serve) for a CPU not taking the interrupt; whether it served.
-    pub fn serve_if_owed(&self, cpu: usize, flush: impl FnOnce()) -> bool {
+    pub fn serve_if_owed(&self, cpu: usize, flush: impl FnOnce(Reach)) -> bool {
         let owed = self.owes(cpu);
         if owed {
             self.serve(cpu, flush);
@@ -97,7 +131,7 @@ impl Shootdown {
         me: usize,
         cpu: usize,
         generation: Generation,
-        flush: impl FnOnce(),
+        flush: impl FnOnce(Reach),
     ) -> bool {
         // An initiator inside an `IrqGuard`, a handler or the tick's pass takes no IPI, so this CPU must serve itself or two initiators waiting on each other deadlock.
         self.serve_if_owed(me, flush);
diff --git a/kernel/src/syscall/dispatch.rs b/kernel/src/syscall/dispatch.rs
index dc119ec77ed..d5286ccfe98 100644
--- a/kernel/src/syscall/dispatch.rs
+++ b/kernel/src/syscall/dispatch.rs
@@ -652,6 +652,8 @@ pub(crate) fn syscall_dispatch(num: u64, a1: u64, a2: u64, a3: u64, a4: u64) ->
             // An interrupt inside a syscall's body, which nothing a guest does puts there on demand.
             DA::RING0_TIMER_IN_SYSCALL => ring0_timer_in_syscall(),
             DA::ACPI_FIRMWARE_LOCK => crate::arch::acpi_mode::debug_firmware_lock(a2),
+            // A translation outliving its entry, which only a TLB that keeps one shows.
+            DA::GLOBAL_LEAF_REWRITE => crate::mm::paging::debug_global_leaf(a2),
             _ => SyscallError::InvalidArgument.to_u64(),
         },
         SYS_SCHED_INFO => match ctx.copy_out(UserAddr::new(a1), &sys_sched_info()) {
diff --git a/kernel/src/syscall/vm.rs b/kernel/src/syscall/vm.rs
index 0b465c8c4b3..9a26e689e9a 100644
--- a/kernel/src/syscall/vm.rs
+++ b/kernel/src/syscall/vm.rs
@@ -293,7 +293,7 @@ pub(super) fn sys_dlopen(ctx: &crate::user_ptr::SyscallContext, path: &str, init
         // `map_window`'s shootdown reached only this CPU, so the rest of the
         // machine is told here.
         if matches!(lib.memory, crate::elf::LibMemory::Shared { .. }) {
-            crate::arch::tlb::shootdown(crate::invalidation::Origin::Dlopen);
+            crate::arch::tlb::shootdown(crate::invalidation::Origin::Dlopen, crate::invalidation::Half::User);
         }
         if vaddr != lib.user_base {
             lib.user_base = vaddr;
@@ -316,7 +316,7 @@ pub(super) fn sys_dlopen(ctx: &crate::user_ptr::SyscallContext, path: &str, init
         process::with_process_data(|_data| {
             pt.lock().free_and_unmap(base);
         });
-        crate::arch::tlb::shootdown(crate::invalidation::Origin::Dlopen);
+        crate::arch::tlb::shootdown(crate::invalidation::Origin::Dlopen, crate::invalidation::Half::User);
     });
 
     let lib_tls = lib.tls().and_then(toyos_elf::TlsSegment::occupied);
diff --git a/sdk/abi/src/syscall.rs b/sdk/abi/src/syscall.rs
index bc77b9c0f27..d2e9c5a3512 100644
--- a/sdk/abi/src/syscall.rs
+++ b/sdk/abi/src/syscall.rs
@@ -892,6 +892,28 @@ pub mod debug_action {
     pub const RING0_FIRE_REARMED: u64 = 0;
     pub const RING0_FIRE_NEVER: u64 = 1;
     pub const RING0_FIRE_OTHER_SPAN: u64 = 2;
+    /// Map a global 2 MiB kernel leaf over one frame and read through it, point
+    /// the leaf at a second frame and read again with nothing invalidated, the
+    /// control, invalidate as the argument says, and read a third time, all on
+    /// one CPU with interrupts closed. The argument is
+    /// [`GLOBAL_LEAF_DISCHARGED`], the write's own discharge,
+    /// [`GLOBAL_LEAF_KERNEL_SHOT_DOWN`], the machine-wide shootdown a
+    /// kernel-half change asks, or [`GLOBAL_LEAF_USER_SHOT_DOWN`], the one a
+    /// user-half change asks. Answers [`GLOBAL_LEAF_SAW_OLD`] or
+    /// [`GLOBAL_LEAF_SAW_NEW`], the frame the third read saw;
+    /// [`GLOBAL_LEAF_UNHELD`] where the control already saw the second frame;
+    /// [`GLOBAL_LEAF_WIDENED`] where the user-half shootdown's answer on this
+    /// CPU covered another CPU's kernel-half one and so dropped global
+    /// translations; `NotSupported` on AArch64, `AlreadyExists` where
+    /// something is already mapped where the leaf goes.
+    pub const GLOBAL_LEAF_REWRITE: u64 = 28;
+    pub const GLOBAL_LEAF_DISCHARGED: u64 = 0;
+    pub const GLOBAL_LEAF_KERNEL_SHOT_DOWN: u64 = 1;
+    pub const GLOBAL_LEAF_USER_SHOT_DOWN: u64 = 2;
+    pub const GLOBAL_LEAF_SAW_OLD: u64 = 0;
+    pub const GLOBAL_LEAF_SAW_NEW: u64 = 1;
+    pub const GLOBAL_LEAF_UNHELD: u64 = 2;
+    pub const GLOBAL_LEAF_WIDENED: u64 = 3;
 }
 
 /// Create a pipe. Returns the read and write ends.
diff --git a/tests/checks.rs b/tests/checks.rs
index c3d7c966979..be005f45cab 100644
--- a/tests/checks.rs
+++ b/tests/checks.rs
@@ -342,7 +342,7 @@ mod checks {
     fn control_regs_verdict() -> Result<(), String> {
         // `xcr0=0x0` is what the kernel prints for a CPU without `CR4.OSXSAVE`.
         const AP_BEFORE: (u64, u64, u64) = (0xe000_0011, 0x0031_0620, 0);
-        const DECLARED: (u64, u64, u64) = (0x8001_0033, 0x0034_0668, 0x7);
+        const DECLARED: (u64, u64, u64) = (0x8001_0033, 0x0034_06e8, 0x7);
         const AVX512: u64 = 0xe0;
 
         fn log(cpus: &[(u64, u64, u64)]) -> String {
@@ -396,15 +396,16 @@ mod checks {
         )?;
         // Two bits a machine could hold uniformly, each one line of kernel diff
         // away, and neither reachable by an actuator. `AM` is named clear above and
-        // answers by name; `PGE` is named nowhere, which is the case the whole
-        // never-named rule exists for — `TSD` and `PKE` are the same case.
+        // answers by name; `PKE` is named nowhere, which is the case the whole
+        // never-named rule exists for — `TSD` is the same case.
         refused("every CPU with AM set", &[(DECLARED.0 | (1 << 18), DECLARED.1, DECLARED.2); 4], "AM")?;
         refused(
-            "every CPU with PGE set",
-            &[(DECLARED.0, DECLARED.1 | (1 << 7), DECLARED.2); 4],
+            "every CPU with PKE set",
+            &[(DECLARED.0, DECLARED.1 | (1 << 22), DECLARED.2); 4],
             "never named",
         )?;
         refused("every CPU without SMEP", &[(DECLARED.0, DECLARED.1 & !(1 << 20), DECLARED.2); 4], "SMEP")?;
+        refused("every CPU without PGE", &[(DECLARED.0, DECLARED.1 & !(1 << 7), DECLARED.2); 4], "PGE")?;
         // A CPU that agrees about every named bit and differs in one the CPU is
         // allowed to withhold, so nothing above it can object.
         refused(
@@ -416,7 +417,7 @@ mod checks {
         // died before the check looks like.
         refused("three lines for four CPUs", &[DECLARED; 3], "{0, 1, 2, 3}")?;
 
-        eprintln!("  [control_regs] the verdict refuses 14 machines and accepts the two declared ones");
+        eprintln!("  [control_regs] the verdict refuses 15 machines and accepts the two declared ones");
         Ok(())
     }
 
diff --git a/tests/guest/rust-tests/src/bin/global_leaf_discharge.rs b/tests/guest/rust-tests/src/bin/global_leaf_discharge.rs
new file mode 100644
index 00000000000..10e241c3fa1
--- /dev/null
+++ b/tests/guest/rust-tests/src/bin/global_leaf_discharge.rs
@@ -0,0 +1,14 @@
+//! A global kernel leaf rewritten through its own discharge is read fresh by
+//! the CPU that wrote it: the discharge dropped the global translation, which
+//! `INVPCID` type 0 and a `CR3` write both leave.
+
+#[path = "../global_leaf.rs"]
+mod global_leaf;
+
+use toyos_abi::syscall::debug_action as da;
+
+fn main() {
+    let saw = global_leaf::rewritten(da::GLOBAL_LEAF_DISCHARGED);
+    assert_eq!(saw, da::GLOBAL_LEAF_SAW_NEW, "the write's discharge left the old frame's global translation");
+    println!("global_leaf_discharge: the discharge dropped the global translation");
+}
diff --git a/tests/guest/rust-tests/src/bin/global_leaf_flush.rs b/tests/guest/rust-tests/src/bin/global_leaf_flush.rs
new file mode 100644
index 00000000000..317b6025726
--- /dev/null
+++ b/tests/guest/rust-tests/src/bin/global_leaf_flush.rs
@@ -0,0 +1,14 @@
+//! A global kernel leaf rewritten under the machine-wide shootdown a kernel-half
+//! change asks is read fresh by the CPU that issued it: that flush drops global
+//! translations, which `INVPCID` type 3 and a `CR3` write both leave.
+
+#[path = "../global_leaf.rs"]
+mod global_leaf;
+
+use toyos_abi::syscall::debug_action as da;
+
+fn main() {
+    let saw = global_leaf::rewritten(da::GLOBAL_LEAF_KERNEL_SHOT_DOWN);
+    assert_eq!(saw, da::GLOBAL_LEAF_SAW_NEW, "the shootdown's flush left the old frame's global translation");
+    println!("global_leaf_flush: the shootdown's flush dropped the global translation");
+}
diff --git a/tests/guest/rust-tests/src/bin/global_leaf_user_flush.rs b/tests/guest/rust-tests/src/bin/global_leaf_user_flush.rs
new file mode 100644
index 00000000000..f4454f184c1
--- /dev/null
+++ b/tests/guest/rust-tests/src/bin/global_leaf_user_flush.rs
@@ -0,0 +1,15 @@
+//! A global kernel leaf rewritten under the machine-wide shootdown a user-half
+//! change asks is still read stale by the CPU that issued it: that flush is
+//! `INVPCID` type 3 or a `CR3` reload, which keep global translations, and only
+//! a kernel-half change asks the flush that drops them.
+
+#[path = "../global_leaf.rs"]
+mod global_leaf;
+
+use toyos_abi::syscall::debug_action as da;
+
+fn main() {
+    let saw = global_leaf::rewritten(da::GLOBAL_LEAF_USER_SHOT_DOWN);
+    assert_eq!(saw, da::GLOBAL_LEAF_SAW_OLD, "the user-half shootdown dropped a global kernel translation");
+    println!("global_leaf_user_flush: the user-half shootdown kept the global translation");
+}
diff --git a/tests/guest/rust-tests/src/global_leaf.rs b/tests/guest/rust-tests/src/global_leaf.rs
new file mode 100644
index 00000000000..2a05ebc3710
--- /dev/null
+++ b/tests/guest/rust-tests/src/global_leaf.rs
@@ -0,0 +1,33 @@
+//! `SYS_DEBUG`'s `GLOBAL_LEAF_REWRITE`, for the three metal rows that judge
+//! which flush drops a global kernel translation and which keeps it.
+//!
+//! The action maps a global 2 MiB kernel leaf, reads through it, points it at a
+//! second frame, reads again with nothing invalidated (the control),
+//! invalidates as asked and reads a third time, on one CPU with interrupts
+//! closed. **Metal rows and not guest tests**: a metal row is the cheaper tier,
+//! and this host's TCG guest drops every translation on a `CR3` write and has
+//! no `INVPCID`, so it reads fresh where a CPU that keeps a global translation
+//! reads stale.
+
+use toyos_abi::syscall::{self, debug_action as da};
+
+/// The frame the third read saw after `how`, refused where what it read after
+/// `how` would be no evidence of `how`: the control saw the second frame, so
+/// that CPU held no translation to go stale; or a user-half shootdown answered
+/// wide, covering a kernel-half one another CPU issued inside it.
+pub fn rewritten(how: u64) -> u64 {
+    let saw = syscall::debug_with(da::GLOBAL_LEAF_REWRITE, how);
+    assert_ne!(
+        saw,
+        da::GLOBAL_LEAF_UNHELD,
+        "the control: with nothing invalidated the second read did not see the first frame, so \
+         this CPU held no translation to go stale and the row judges nothing",
+    );
+    assert_ne!(
+        saw,
+        da::GLOBAL_LEAF_WIDENED,
+        "the user-half shootdown's answer covered another CPU's kernel-half one, which drops \
+         global translations, so the row judges nothing",
+    );
+    saw
+}
diff --git a/tests/toyos.rs b/tests/toyos.rs
index c5b2f432033..e3ec931e1e2 100644
--- a/tests/toyos.rs
+++ b/tests/toyos.rs
@@ -87,6 +87,14 @@ const RUST_SKIP: &[&str] = &[
     // is read off the clock around the syscall. The `tlb_shootdown_waits`
     // metal row runs it on the T14.
     "tlb_shootdown_waits",
+    // Each reads a global translation its kernel may leave stale, which a TCG
+    // guest's TLB drops on any `CR3` write and the T14's keeps: the
+    // `a_global_kernel_leafs_discharge_drops_it`,
+    // `a_kernel_half_shootdown_drops_a_global_kernel_translation` and
+    // `a_user_half_shootdown_keeps_a_global_kernel_translation` metal rows run them.
+    "global_leaf_discharge",
+    "global_leaf_flush",
+    "global_leaf_user_flush",
     // Its product is the windows a `mask-windows` kernel reports under it, and
     // its 256 threads' stacks would crowd every member after it in one boot.
     // The `mask_windows` metal row runs it.
@@ -706,6 +714,39 @@ const METAL: &[(&str, metal::Metal)] = &[
             judge: |b| b[0].job_passed("test_rs_tlb_shootdown_waits"),
         },
     ),
+    (
+        // A global kernel translation outlives its entry until something drops
+        // it, and the kernel that carries `SYS_DEBUG` stages one: the verdict
+        // is the guest's own exit, its control first.
+        "a_global_kernel_leafs_discharge_drops_it",
+        metal::Metal {
+            arms: &[metal::Arm {
+                features: ACTUATOR_KERNEL,
+                ..metal::once("shared-debug", "tests/images/testcases.toml", SELFTEST_PARAMS, &["test_rs_global_leaf_discharge"])
+            }],
+            judge: |b| b[0].job_passed("test_rs_global_leaf_discharge"),
+        },
+    ),
+    (
+        "a_kernel_half_shootdown_drops_a_global_kernel_translation",
+        metal::Metal {
+            arms: &[metal::Arm {
+                features: ACTUATOR_KERNEL,
+                ..metal::once("shared-debug", "tests/images/testcases.toml", SELFTEST_PARAMS, &["test_rs_global_leaf_flush"])
+            }],
+            judge: |b| b[0].job_passed("test_rs_global_leaf_flush"),
+        },
+    ),
+    (
+        "a_user_half_shootdown_keeps_a_global_kernel_translation",
+        metal::Metal {
+            arms: &[metal::Arm {
+                features: ACTUATOR_KERNEL,
+                ..metal::once("shared-debug", "tests/images/testcases.toml", SELFTEST_PARAMS, &["test_rs_global_leaf_user_flush"])
+            }],
+            judge: |b| b[0].job_passed("test_rs_global_leaf_user_flush"),
+        },
+    ),
     (
         // What one diary record costs its writer: a million written back to
         // back by `SYS_DEBUG`'s flood, so the kernel that carries it, timed
@@ -5525,6 +5566,8 @@ fn control_regs(log: &str, cpus: u32) -> Result<(), String> {
         (3, "DE", true),
         (5, "PAE", true),
         (6, "MCE", true),
+        // Every kernel leaf is global, and only with `PGE` set is one cached so.
+        (7, "PGE", true),
         (9, "OSFXSR", true),
         (10, "OSXMMEXCPT", true),
         (12, "LA57", false),
@@ -5582,7 +5625,7 @@ fn control_regs(log: &str, cpus: u32) -> Result<(), String> {
     }
 
     // A bit named nowhere above is as much of the declaration as a named one,
-    // and the kernel writes both registers whole — so `UMIP`, `PGE`, `TSD` or
+    // and the kernel writes both registers whole — so `UMIP`, `TSD` or
     // `PKE` would reach every CPU with nothing here to say so. Which is why
     // what follows is the set this gate has an opinion about, rather than a
     // second list of bits to forbid: a forbid-list fails open on the next bit.