diff --git a/compiler/rustc_mir_build/src/builder/mod.rs b/compiler/rustc_mir_build/src/builder/mod.rs
index ff2b59ee07d20..f9fa750e750b3 100644
--- a/compiler/rustc_mir_build/src/builder/mod.rs
+++ b/compiler/rustc_mir_build/src/builder/mod.rs
@@ -612,7 +612,8 @@ fn construct_error(tcx: TyCtxt<'_>, def_id: LocalDefId, guar: ErrorGuaranteed) -
         | DefKind::AssocConst
         | DefKind::AnonConst
         | DefKind::InlineConst
-        | DefKind::Static { .. } => (vec![], tcx.type_of(def_id).instantiate_identity(), None),
+        | DefKind::Static { .. }
+        | DefKind::GlobalAsm => (vec![], tcx.type_of(def_id).instantiate_identity(), None),
         DefKind::Ctor(..) | DefKind::Fn | DefKind::AssocFn => {
             let sig = tcx.liberate_late_bound_regions(
                 def_id.to_def_id(),
diff --git a/compiler/rustc_mir_transform/src/lib.rs b/compiler/rustc_mir_transform/src/lib.rs
index 5df12ac4d8bc2..8aa14d15644b0 100644
--- a/compiler/rustc_mir_transform/src/lib.rs
+++ b/compiler/rustc_mir_transform/src/lib.rs
@@ -316,6 +316,10 @@ fn mir_keys(tcx: TyCtxt<'_>, (): ()) -> FxIndexSet<LocalDefId> {
     // All body-owners have MIR associated with them.
     let mut set: FxIndexSet<_> = tcx.hir_body_owners().collect();
 
+    // Remove the fake bodies for `global_asm!`, since they're not useful
+    // to be emitted (`--emit=mir`) or encoded (in metadata).
+    set.retain(|&def_id| !matches!(tcx.def_kind(def_id), DefKind::GlobalAsm));
+
     // Coroutine-closures (e.g. async closures) have an additional by-move MIR
     // body that isn't in the HIR.
     for body_owner in tcx.hir_body_owners() {
diff --git a/tests/codegen/target-feature-inline-closure.rs b/tests/codegen/target-feature-inline-closure.rs
index 73bdbc0e1a841..5d54444f994ac 100644
--- a/tests/codegen/target-feature-inline-closure.rs
+++ b/tests/codegen/target-feature-inline-closure.rs
@@ -12,7 +12,7 @@ use std::arch::x86_64::*;
 #[cfg(target_arch = "x86_64")]
 #[target_feature(enable = "avx")]
 fn with_avx(x: __m256) -> __m256 {
-    // CHECK: fadd
+    // CHECK: fadd <8 x float>
     let add = {
         #[inline(always)]
         |x, y| unsafe { _mm256_add_ps(x, y) }
@@ -24,14 +24,10 @@ fn with_avx(x: __m256) -> __m256 {
 #[no_mangle]
 #[cfg(target_arch = "x86_64")]
 unsafe fn without_avx(x: __m256) -> __m256 {
-    // CHECK-NOT: fadd
+    // CHECK-NOT: fadd <8 x float>
     let add = {
         #[inline(always)]
         |x, y| unsafe { _mm256_add_ps(x, y) }
     };
     add(x, x)
 }
-
-// Don't allow the above CHECK-NOT to accidentally match a commit hash in the
-// compiler version.
-// CHECK-LABEL: rustc version
diff --git a/tests/ui/asm/global-asm-isnt-really-a-mir-body.rs b/tests/ui/asm/global-asm-isnt-really-a-mir-body.rs
new file mode 100644
index 0000000000000..b7636d116ec6c
--- /dev/null
+++ b/tests/ui/asm/global-asm-isnt-really-a-mir-body.rs
@@ -0,0 +1,26 @@
+//@ revisions: emit_mir instrument cfi
+
+// Make sure we don't try to emit MIR for it.
+//@[emit_mir] compile-flags: --emit=mir
+
+// Make sure we don't try to instrument it.
+//@[instrument] compile-flags: -Cinstrument-coverage -Zno-profiler-runtime
+//@[instrument] only-linux
+
+// Make sure we don't try to CFI encode it.
+//@[cfi] compile-flags: -Zsanitizer=cfi -Ccodegen-units=1 -Clto -Ctarget-feature=-crt-static -Clink-dead-code=true
+//@[cfi] needs-sanitizer-cfi
+//@[cfi] no-prefer-dynamic
+// FIXME(#122848) Remove only-linux once OSX CFI binaries work
+//@[cfi] only-linux
+
+//@ build-pass
+//@ needs-asm-support
+
+use std::arch::global_asm;
+
+fn foo() {}
+
+global_asm!("/* {} */", sym foo);
+
+fn main() {}
diff --git a/tests/ui/asm/global-asm-with-error.rs b/tests/ui/asm/global-asm-with-error.rs
new file mode 100644
index 0000000000000..c2253e3cb8747
--- /dev/null
+++ b/tests/ui/asm/global-asm-with-error.rs
@@ -0,0 +1,11 @@
+// Ensure that we don't ICE when constructing the fake MIR body for a global
+// asm when the body has errors. See #137470.
+
+//@ needs-asm-support
+
+use std::arch::global_asm;
+
+global_asm!("/* {} */", sym a);
+//~^ ERROR cannot find value `a` in this scope
+
+fn main() {}
diff --git a/tests/ui/asm/global-asm-with-error.stderr b/tests/ui/asm/global-asm-with-error.stderr
new file mode 100644
index 0000000000000..3b76bfd19485c
--- /dev/null
+++ b/tests/ui/asm/global-asm-with-error.stderr
@@ -0,0 +1,9 @@
+error[E0425]: cannot find value `a` in this scope
+  --> $DIR/global-asm-with-error.rs:8:29
+   |
+LL | global_asm!("/* {} */", sym a);
+   |                             ^ not found in this scope
+
+error: aborting due to 1 previous error
+
+For more information about this error, try `rustc --explain E0425`.