Skip to content

LLVM/MLIR 清单

本页中的所有地址、偏移和版本字符串均适用于 libtpu-0.0.40-cp314 wheel 中的 libtpu.so(发布版 libtpu_lts_20260413_b_RC00,build-id md5 89edbbe81c5b328a958fe628a9f2207d,磁盘大小 781,691,048 字节)。其他 wheel 会有所不同。

摘要

libtpu.so 不是一个轻量的运行时垫片,而是把整个编译器静态链接进了一个共享对象。粗略估计,其代码和 rodata 字节中约三分之一来自上游 LLVM 和 MLIR,这些代码取自单个 Google 内部 LLVM monorepo 快照,并连同树外 TPU LLVM 后端、上游 MLIR 方言集合,以及一组 TPU 专用 MLIR 方言(tpu, llo, sparse_core, mosaic_sc, xtile)一起被拉入。本页建立权威的 what's-compiled-in 清单:精确的 LLVM/MLIR 版本证据、存在的 LLVM core/CodeGen/MC 组件、链接的 MLIR 方言和基础设施,以及随工具链一起进入二进制的第三方库主要版本。

版本框架是核心复杂点。Google 的构建系统会把上游 LLVM_VERSION_MAJOR 改写为滚动哨兵值,因此二进制中没有 LLVM 23.0.0 横幅。它确实携带的内容,固定在 .rodata 的字节偏移处,包括两个 monorepo 提交 SHA(clangLLVM head)、一个 g3_____-trunk 修订标签、9999.0.0 哨兵,以及构建 epoch。由此可通过发布分支日历把上游主版本界定为 LLVM 23-dev(tip-of-trunk,约 2026 年 4 月),而不是直接读出。下方每个组件判断都锚定在二进制中恢复出的确定符号或 .rodata 字面量;版本窗口是唯一的推断,并已明确标注。

注意 — 这个 wheel 中的二进制未 strip.symtab 带有 1,233,710 个条目(sh_size 0x1c3cc50 ÷ 24readelf 在该 section 的 Inf 列打印的 1232970 是第一个全局符号索引,不是条目数),并保留完整 Itanium mangling。因此 FLIRT 模式匹配没有意义:下面的每个组件都由精确的 demangled 符号或 rodata 字符串确认,而不是模糊签名。生产环境 strip 会移除 .symtab/.strtab,但会保留这里识别出的所有代码字节。

对重新实现而言,契约是:

  • 版本 pin — 工具链跟踪哪个 LLVM/MLIR 提交,以及如何从被哨兵遮蔽的构建界定上游主版本。
  • LLVM 组件集 — Core IR、两条 ISel 路径、MachineCodeGen、MC 层、七个注册目标后端、分析/转换 pass pipeline,以及两个 AOT MLGO advisor 模型。
  • MLIR 组件集 — Core IR、三层方言注册表、pass 基础设施、bytecode reader/writer、转换框架,以及 LLVM-IR 翻译路径(但没有 ExecutionEngine/JIT)。
  • 嵌入库版本 — 与工具链一起编译的关键第三方 pin(Abseil、protobuf、Eigen、DNNL、ICU、libc++ 等),完整树推迟到 Embedded-Library Atlas
LLVM/MLIR 版本LLVM 23-dev (trunk), monorepo commit 8918319853fbdf9e6f6cb69e96848f913a22bc31
clang 版本字面量g3 clang version 9999.0.0 (a70419505471bd8240ef3451dcdd541f8676477c) @ 0xaf591d4
LLVM 版本字面量LLVM version g3_____-trunk 8918319853fbdf… @ 0xb1fa070
llvm-mc 字面量llvm-mc (based on LLVM g3_____-trunk 8918319853fbdf…) @ 0xa12d48e
构建 epochBuilt on Apr 13 2026 14:17:21 (1776115041) @ 0x84a1d90
代码/数据模型大代码模型(.lrodata 113.3 MB,存在 .lbss),x86-64,clang/lld
MLIR 来源与 LLVM 相同的 monorepo commit — 没有单独的 MLIR 版本
注册目标后端X86, AArch64, ARM, AMDGPU (+R600), PowerPC, NVPTX, TPU(七个,由 LLVMInitialize*TargetInfo 判断)
LLVM + MLIR 占用~84.1 MB LLVM + ~72.2 MB MLIR = ~156 MB 代码+rodata 符号字节(约占 T+R 的 ~37%)

LLVM/MLIR 版本 Pin

目的

固定精确的工具链版本,使取证页面上的每个 llvm::/mlir:: 地址都没有歧义,并解释上游主版本为什么不能直接从字符串读出。

版本证据

.rodata 中有三个版本字面量,每个都可用 strings -t x 恢复。它们是整个工具链的主要锚点:

text
0xa12d48e  llvm-mc (based on LLVM g3_____-trunk 8918319853fbdf9e6f6cb69e96848f913a22bc31)
0xaf591d4  …PIC LevelCode ModelLarge Data Threshold…g3      clang version 9999.0.0 (a70419505471bd8240ef3451dcdd541f8676477c)
0xb1fa070  LLVM version g3_____-trunk 8918319853fbdf9e6f6cb69e96848f913a22bc31
```text

clang 和 LLVM SHA 是**同一个 monorepo 的两个 head**(快照点上 `clang` 最后触及的提交与 LLVM core 最后触及的提交),不是两个仓库。MLIR 也从同一个 monorepo 的同一提交发出,因此没有可寻找的独立 MLIR 版本号。`clang` 字面量与 `Code Model` / `Large Data Threshold` driver 标志位于同一段字符串池,这本身就是该二进制使用**大代码模型**构建的证据(见 [ELF Anatomy](elf-anatomy.md) 中的 `.lrodata`/`.lbss` sections)。

| 字段 | 值 | 锚点 |
|---|---|---|
| LLVM trunk SHA | `8918319853fbdf9e6f6cb69e96848f913a22bc31` | `.rodata` `0xb1fa070` |
| clang trunk SHA | `a70419505471bd8240ef3451dcdd541f8676477c` | `.rodata` `0xaf591d4` |
| 修订标签 | `g3_____-trunk`(Google rolling-trunk sync) | 两个字面量 |
| clang 版本字符串 | `9999.0.0`(Google tip-of-trunk 哨兵) | clang 字面量 |
| 构建工具链 | `Bazel, release r4rca-2026.04.04-1 (mainline @894239244)` @ `0x84a12a0` | `.rodata` |
| 构建 epoch (UTC) | `2026-04-13 21:17:21` (epoch 1776115041) | `.rodata` `0x84a1d90` |

### 界定上游主版本

`9999.0.0` 哨兵和 `g3_____-trunk` 标签有意遮蔽 `LLVM_VERSION_MAJOR`;没有 `LLVM 23.0.0` 风格的字面量保留下来。主版本由**发布分支日历**确定,而不是由字符串确定:

```text
Build epoch          : 2026-04-13  (RC cut from piper @894239244; Bazel r4rca-2026.04.04-1)
google3 sync lag     : days-to-low-weeks behind upstream main
=> upstream main window: ~late-March to mid-April 2026

LLVM release cadence (6-month major; branch ~6-8 wk before .1.0):
  21.x branched ~Jul 2025      (main was 21.0.0git in 2025)
  22.x branched ~late-Jan/Feb 2026, 22.1.0 ~Mar 2026
  23.x branches  ~Jul/Aug 2026

Once 22.x branches (Jan/Feb 2026), main's LLVM_VERSION_MAJOR bumps to 23.
By April 2026 upstream main reports 23.0.0git.
=> embedded LLVM = LLVM 23-dev (trunk), post-22.x-branch, pre-23.x-branch.

陷阱 — 不要把 9999.0.0 当成真实版本,并假设它是穷尽稳定的发布 ABI。这是 tip-of-trunk:它可能包含 22 分支之后的 IR/pass 变更,这些变更不属于任何带标签的发布版,并且它早于 23.0.0 发布。面向 “LLVM 22” 或 “LLVM 23” 带标签源码的重新实现者会遇到 API 漂移;唯一精确参考是 monorepo commit 8918319853fbdf…

注意 — 主版本窗口(23-dev)置信度为 HIGH,由构建 epoch 和确定性的分支/版本递增机制界定。这是本页唯一的推断数据;其他所有内容都是直接符号/字符串命中。剩余缺口是 8918319853fbdf… 的精确上游提交日期,它没有嵌入(只有 SHA),因此把“23-dev 窗口”转换为“截至 YYYY-MM-DD 的 23-dev”需要外部 monorepo 查询。


LLVM Core

目的

枚举静态链接进来的 LLVM 组件:这是完整的代码生成工具链,而不是桩。核心结论是:完整 SelectionDAG + MachineCodeGen + MC 栈都存在,若干上游目标后端与树外 TPU 目标一起被链接,并且两个 MLGO advisor 模型以 AOT 形式烘入 rodata。

组件清单

每一行都是确定符号命中(nm -C libtpu.so | rg …)。存在具体类符号时,置信度为 CERTAIN

LLVM 组件存在主要证据(defined symbol)
Core IRYESllvm::Module, llvm::Function, llvm::BasicBlock, llvm::Instruction, llvm::LLVMContext
Bitcode reader/writerYESllvm::BitcodeReader, llvm::parseBitcodeFile, llvm::WriteBitcodeToFile
SelectionDAG ISelYESllvm::SelectionDAG, llvm::SelectionDAGISel, llvm::TargetLowering
GlobalISel 基础设施YES(已链接)llvm::InstructionSelect, llvm::LegalizerInfo, llvm::RegisterBankInfo
MachineCodeGenYESllvm::MachineFunction, llvm::MachineInstr, llvm::LiveIntervals
MC 层YESllvm::MCStreamer, llvm::MCInst, llvm::MCCodeEmitter (+ TPUMCCodeEmitter)
TPU 目标后端YESllvm::TPUTargetMachine + llvm::TPU* 家族(见下方)
分析 passesYESllvm::ScalarEvolution, llvm::PassBuilder (NewPM)
MLGO advisor 模型YES (2)RegAllocEvictModel, InlinerSizeModel(AOT,见下方)
嵌入式 LLVM bitcodeYESkEigenUnaryLlIr_constant_buffer_contents @ 0xaf58000
MCJIT / ORCYESllvm::MCJIT, llvm::orc::*(LLVM ExecutionEngine — 注意:不同于缺失的 mlir::ExecutionEngine);XLA CPU 后端会 JIT llvm::Module

怪癖 — 两套 ISel 基础设施都被链接。GlobalISel(InstructionSelect/LegalizerInfo/RegisterBankInfo)存在,但 TPU 路径的 MC emitter(TPUMCCodeEmitter::getBinaryCodeForInstr @ 0x13c74da0,一个覆盖 InstBits 的 5,667-case switch)位于 MachineInstr 下游,符合 SelectionDAG 为主的后端。仅凭符号表面无法判定 TPU 目标是否也对某些 opcode 具备 GlobalISel 路径;这需要反汇编 pass-pipeline 构造器。

已链接目标后端

LLVMInitialize*TargetInfo 注册清单证明该二进制注册了七个 LLVM 目标后端,而不仅是自定义 TPU 后端:

text
LLVMInitializeX86TargetInfo        ── host backend (the binary runs on x86-64)
LLVMInitializeAArch64TargetInfo
LLVMInitializeARMTargetInfo
LLVMInitializeAMDGPUTargetInfo     ── (+ R600MCCodeEmitter, the legacy AMDGPU sub-target)
LLVMInitializePowerPCTargetInfo
LLVMInitializeNVPTXTargetInfo
LLVMInitializeTPUTargetInfo        ── the out-of-tree Google backend (this page's headline)
```text

`LLVMInitialize*Target`(codegen)initializer 的集合也是同样七个,并且每个都带有实例化的 `*TargetMachine` 类(`llvm::X86TargetMachine`, `llvm::AArch64TargetMachine`, `llvm::ARMBaseTargetMachine`, `llvm::AMDGPUTargetMachine`/`GCNTargetMachine`/`R600TargetMachine`, `llvm::PPCTargetMachine`, `llvm::NVPTXTargetMachine`, `llvm::TPUTargetMachine`)。每个树内目标都带有自己的 TableGen `InstBits` encoder table,例如 `AMDGPUMCCodeEmitter::…::InstBits` @ `0x29d8910`、`AArch64MCCodeEmitter::…::InstBits` @ `0x397e980`、`PPCMCCodeEmitter::…::InstBits` @ `0x3c0d770`。它们是 `registerAllTargets()` 的后果(与下方拉入未使用 MLIR target dialect 的同一种构建系统过度链接)。

> **怪癖 —** Hexagon 源 TU 被*部分*链接(约 140 个 `_GLOBAL__sub_I_Hexagon*.cpp` 静态初始化 thunk 和少量 `llvm::HexagonSubtarget::*Mutation` 符号保留下来),但 Hexagon **不是**可用后端:没有 `LLVMInitializeHexagon*` initializer,没有 `llvm::HexagonTargetMachine` vtable 或构造器,也没有 `HexagonMCCodeEmitter`。这是过度链接留下的死静态初始化残留,不是注册目标;不要把它计入这七个。
>
> **Note:** TPU 目标是唯一的*自定义*后端,但它是**七个**注册后端之一。**X86、AArch64、ARM、AMDGPU+R600、PowerPC 和 NVPTX 全都被静态链接并注册**,与 TPU 并列。X86 是 host backend,由 `llvm::X86TargetMachine`/`llvm::X86Subtarget`/`llvm::X86InstrInfo` 和 `createX86MCCodeEmitter` 确认。(另外存在 Hexagon 静态初始化残留,但它**不是**注册后端;见上面的 QUIRK。)

### TPU 目标后端

树外 `TPU` 目标是最具辨识度的 LLVM 组件:一个名为 “TPU” 的完整 `llvm::Target`,上游不存在。它的 TableGen 表被定位到固定地址并测得大小:

| 表 / 函数 | 地址 | 作用 |
|---|---|---|
| `TPUMCCodeEmitter::…::InstBits` | `0x3366d90` | 每 opcode 指令编码位(TensorCore) |
| `…::InstBits_BarnaCorePxcHwMode` | `0x33931f0` | BarnaCore(Pufferfish PXC)HwMode 编码变体 |
| `llvm::TPUDescs` | `0x33bf650` | 每条指令的 `MCInstrDesc` 表 |
| `llvm::TPUInstrNameData` | `0x33f2be0` | 指令 mnemonic 字符串池 |
| `llvm::TPUFeatureKV` | `0x21934550` | `SubtargetFeatureKV` key/value(16 个 feature;1,152 B / 72 B stride) |
| `llvm::TPUSubTypeKV` | `0x21934ca0` | subtype/CPU key/value(9 个 CPU variant;1,008 B / 112 B stride) |
| `TPUMCCodeEmitter::getBinaryCodeForInstr` | `0x13c74da0` | 5,667-case encoder switch(按 HwMode 选择 `InstBits`) |

五个硅片 subtarget 类(另加一个生成的 `llvm::TPUGenMCSubtarget`)覆盖 HAL 家族中的 TensorCore 世代(缩写映射到 TPU 代号;SparseCore sequencer 拆分记录在 [tpu dialect](../compiler/tpu-dialect-and-ops.md) 和 lowering 页面中):

```text
llvm::TPUSubtarget      ── base (jellyfish/dragonfish/pufferfish TensorCore)
llvm::TPUBcSubtarget    ── BarnaCore PXC
llvm::TPUVfcSubtarget   ── vfc = viperfish (v5)
llvm::TPUGlcSubtarget   ── glc = ghostlite (gen 4; mktg "Trillium")
llvm::TPUGfcSubtarget   ── gfc = 6acc60406 / TPU7x (gen 5; mktg "Ironwood")

每代调度模型以九个 *SchedModelSchedClasses 表的形式存在,按 sequencer 类型(SCS / TAC / TEC)和世代(VF / GL / GF)拆分:

text
BarnaCorePFSchedModelSchedClasses                       ── Pufferfish BarnaCore
SparseCoreScs{GF,GL,VF}SchedModelSchedClasses           ── SCS sequencer (3 gens)
SparseCoreTac{GL,VF}SchedModelSchedClasses              ── TAC sequencer (2 gens — NO GF)
SparseCoreTec{GF,GL,VF}SchedModelSchedClasses           ── TEC sequencer (3 gens)
```text

> **怪癖 —** 没有 `SparseCoreTacGF` 表。GF 世代(`gfc` = 6acc60406 / TPU7x,mktg "Ironwood")提供 SCS 和 TEC sequencer,但**没有 TAC** sequencer,因此它的 TAC 调度模型按设计缺失,而不是提取错误。若重新实现者按 generations × sequencer-types 迭代并假设完整 3×3 网格,就会分配硬件并不存在的表。

### MLGO Advisor 模型

两个 ML-guided-optimization 模型以 AOT 形式(通过 `tfcompile`)直接编译为 native code 加 constant-buffer rodata;这里**没有** TF runtime 或 interpreter。它们是通过上游 `LLVM_RAEVICT_MODEL_PATH` / `LLVM_INLINER_MODEL_PATH` 构建机制烘入的 LLVM “release-mode” MLGO 模型:

| 模型 | 使用方 | 证据 |
|---|---|---|
| `RegAllocEvictModel` | LLVM greedy RA `MLRegAllocEvictAdvisor` | `…RegAllocEvictModel…` 符号家族 |
| `InlinerSizeModel` | LLVM inliner `MLInlineAdvisor` | `_llvm__InlinerSizeModel_*_fusion_` constant-buffer 符号 |

这些是 **LLVM-backend** MLGO advisor,不是 XLA learned cost model。`InlinerSizeModel` constant buffer 甚至按 fusion shape 命名(`dot_add_fusion`, `iota_reduce_fusion`, `compare_convert_fusion`, …),确认它们是 trained inliner-for-size weights。

### 嵌入式 LLVM Bitcode

二进制至少携带一个预编译 LLVM bitcode module 作为 rodata blob,这是 bitcode reader 仍在使用的直接证据:

```text
0x0af58000  kEigenUnaryLlIr_constant_buffer_contents   (LLVM bitcode Module; vectorised tanh + FMA/min/max intrinsics)
            llvm_ir::kEigenUnaryLlIr                    (guard/storage symbol @ .bss 0x224ee910)

它在编译期通过 llvm::parseBitcodeFile 链接进 JIT 的 CPU llvm::Module,因此 Eigen lowering 的 tanh 会解析为 vector loop,而不是 libm 调用。这个 blob 是 LLVM IR bitcode(LLVM-23-dev 格式),不同于下方的 MLIR bytecode 表面;两个 reader 都被编译进来了。


MLIR 方言

目的

枚举 MLIR 组件。MLIR 与 LLVM core 来自同一个 monorepo commit,因此它是 LLVM-23-dev 的树内 MLIR,并叠加了针对该 API 编写的 TPU 专用树外方言。

Core IR 和基础设施

MLIR 组件存在主要证据
Core IRYESmlir::MLIRContext, mlir::OpBuilder, mlir::Operation (+ Block/Region/Value/Type)
方言注册表YES67 个具体 *Dialect 类(唯一 vtable for …Dialect,不含 base mlir::Dialect
Pass 基础设施YESmlir::PassManager, mlir::Pass
Bytecode reader/writerYESmlir::BytecodeReader, mlir::writeBytecodeToFile
转换框架YESmlir::ConversionTarget, mlir::TypeConverter, mlir::RewritePatternSet
LLVM-IR 翻译YESmlir::translateModuleToLLVMIR
ExecutionEngine / JITNOmlir::ExecutionEngine 符号

注意 — bytecode reader/writer 和 LLVM-IR 翻译路径都存在,但 mlir::ExecutionEngine 缺失。MLIR 被纯粹用作 AOT translate path:SparseCore lowering 链运行 translateModuleToLLVMIR,并把产生的 llvm::Module 交给 LLVM TPU 后端(LowerToSparseCoreLlvm)。没有 MLIR JIT,这解答了关于是否链接 MLIR ExecutionEngine 的一个长期问题:没有链接。

注意 — 可直接验证的实例化方言类数量是 67:每个具体方言对应一个 vtable for …Dialect 符号,去重并排除 base mlir::Dialect(若计入 base 则为 68)。对原始符号表面使用更宽松的 regex 会膨胀该数字:匹配每个裸 …Dialect token 会把 namespaced 与 unqualified 拼写重复计数(例如 xla::xtile::XTileDialectXTileDialect),并拉入构造器/typeinfo,得到超过 200 的数字。vtable 计数是保守的一类一命中数字。Embedded-Library Atlas 负责完整方言树。

三层方言集合

这些方言分为三组,这种区分对重新实现者很重要:Group A'/B 的大部分内容在 TPU lowering 链中已链接但未使用;这是 registerAllDialects() 过度链接,是七个 LLVM 目标后端的 MLIR 对应物。

分组方言(已验证样本)TPU 路径上的状态
A — 上游 MLIR corefunc, arith, scf, cf, vector, memref, tensor, linalg, affine, math, complex, index已注册;部分用于早期 lowering
A' — LLVM/target 方言llvm, gpu, spirv(以及 X86/NVVM/ROCDL/AMDGPU 家族)已链接,未使用(过度链接后果)
B — HLO / 输入stablehlo, chlo, mhlo, vhlo输入 + 早期 lowering
C — TPU 专用(Google)tpu, llo, sparse_core (ScDialect, LlvmTpuDialect), mosaic_sc, xtileTPU 编译链

Group-C 方言由它们的具体 Dialect 类确认:

text
mlir::tpu::TPUDialect              ── the TPU dialect (tpu/Mosaic ops)
mlir::llo::LLODialect              ── low-level ops (LLO)
mlir::sparse_core::ScDialect       ── SparseCore ops
mlir::sparse_core::LlvmTpuDialect  ── SparseCore -> LLVM-TPU bridge dialect
mlir::mosaic_sc::MosaicSCDialect   ── Mosaic-SparseCore
xla::xtile::XTileDialect           ── XLA XTile tiling dialect
```text

> **怪癖 —** 在一个 *TPU* plugin 中存在 `mlir::gpu::GPUDialect`、`mlir::spirv::SPIRVDialect` 和 X86/NVVM target dialect,并不是 GPU code path 的证据。这是 `registerAllDialects()` 把整个上游方言集合拉入静态链接。同样的过度链接也解释了 AArch64/ARM/AMDGPU/PowerPC LLVM 后端。重新实现者应把 Group A' 视为 TPU 路径上的冗余负担,而不是能力。

---

## 嵌入式第三方库

### 目的

固定工具链编译所依赖第三方库的主要版本。本页给出 LLVM/MLIR 相邻清单及*可确定*版本;[Embedded-Library Atlas](embedded-library-atlas.md) 负责完整的约 60 个库的树和字节核算。

### 版本清单

版本通过三种方式 pin:带版本的**路径字面量**(决定性)、**带版本标签的命名空间**(决定性),或**feature-floor**(某个符号首次出现在已知发布版中,只能固定下限,不能固定精确版本)。

| 库 | 版本 | 版本证据 |
|---|---|---|
| LLVM | 23-dev (trunk) | `LLVM version g3_____-trunk 8918319853fbdf…` @ `0xb1fa070` |
| MLIR | = LLVM(同一提交) | 同一 monorepo commit;没有单独版本 |
| libc++ | = LLVM commit | `std::__u::` ABI tag(Google libc++ inline-namespace build) |
| Intel oneDNN | **v3.3** | 路径字面量 `third_party/intel_dnnl/v3_3/` |
| ICU | **icu_78** | 带版本命名空间 `icu_78::` |
| Protobuf | v32+(EDITION_2024 production) | edition enums `EDITION_2023`, `EDITION_2024`, `EDITION_2026`, `EDITION_UNSTABLE`(无 `EDITION_2025`) |
| Abseil | >= LTS 20240722(下限) | `absl::AnyInvocable`, `absl::log_internal::SetTimeZone`, `absl::CordBuilder`, `absl::StatusOr` |
| Eigen | 3.4.x branch | `Eigen::bfloat16`(现代),`Eigen::half`;没有 `EIGEN_*_VERSION` 字面量保留 |
| gRPC | google3-tip (>= 1.66-dev) | `chaotic_good` transport, `filter_fusion`(仅 trunk) |
| tcmalloc | 2024+(rseq per-CPU) | `google_malloc` ELF section, `__rseq_cs`, rseq cmpxchg family |

> **注意 —** **路径字面量**和**带版本命名空间** pin(oneDNN `v3_3`、ICU `icu_78`)是仅有的 `CERTAIN` 精确版本;它们作为构建嵌入的字面字符串保留下来。其他所有项都是**下限**:二进制表现出已知发布版新增的功能,但没有 `*_VERSION_MAJOR` 宏字面量保留,因此无法仅从该二进制读出精确 point release。Atlas 页面记录了收紧每个下限的逐库复现方法。
>
> **注意 —** 每个 libc++ 符号上的 `std::__u::` inline-namespace ABI tag 是 Google 构建定制(与 LLVM/MLIR 同一 monorepo)。由于 `llvm::`/`mlir::` 模板实例化在各处都模板化于 `std::__u::` 类型,libc++ 构建与 LLVM/MLIR 版本 pin 不可分割:它们是同一套工具链。

### Google 专用定制

在符号表面看不到对 LLVM/MLIR *core* 算法的行为 patch。Google 相对上游的定制都是标准 vendoring 模式,而不是 core fork:

1. **版本哨兵** — `9999.0.0` / `g3_____-trunk` 是上游版本宏的构建身份覆盖(遮蔽 `LLVM_VERSION_MAJOR`;不是行为变更)。
2. **树外 TPU 目标** — 整个 `llvm::TPU*` 后端(5 个硅片 subtarget、`InitMCInstrInfo` 中每个 6,166 个 MC opcode、`InstBits` encoder)是 Google 私有;上游没有 TPU target。
3. **AOT MLGO 模型** — `RegAllocEvictModel` + `InlinerSizeModel` 通过上游 release-model 机制烘入(构建期,不是源码 patch)。
4. **libc++ `__u` ABI tag** — Google libc++ 构建定制。
5. **TPU MLIR 方言** — `tpu`/`llo`/`sparse_core`/`mosaic_sc`/`xtile` 针对树内 MLIR API 编写;是在 stock MLIR 上叠加的树外方言,不是 core 修改。

---

## 字节占用

LLVM 和 MLIR 合起来是二进制中最大的单一代码类别。下方数字按符号桶统计(按命名空间汇总 `nm -S` 大小);界定它们的底层 ELF section 大小已直接确认:`.text` = `0x12bdb484` (314,422,404 B),`.lrodata` = `0x6c0e7d0` (113,305,552 B),`.rodata` = `0x39eaf28` (60,731,176 B)。

| 桶 | 合计字节 | 占 code+rodata 符号字节比例 |
|---|---|---|
| LLVM (`llvm::` + TableGen tables) | ~84.1 MB | ~19.8% |
| MLIR (`mlir::` + dialect templates) | ~72.2 MB | ~17.0% |
| MLGO models (RegAllocEvict + InlinerSize) | ~0.59 MB | ~0.14% |
| `kEigenUnaryLlIr` bitcode blob | 16,384 B | ~0.004% |
| **LLVM + MLIR + MLGO + bitcode** | **~156.9 MB** | **~36.9%** |

> **注意 —** 相比其 `.text`,MLIR 的 `.rodata` 占比很小,因为 MLIR 是模板密集型代码,大型常量表较少;LLVM 的 `.rodata` 很大,因为 TPU TableGen 表(`InstBits` 181,344 B、`TPUDescs`、`TPUInstrNameData`)和树内目标的 `InstBits` 表位于 rodata/lrodata。这些是*按符号桶统计*的大小:它们排除了 anonymous-namespace TU-local 残留,估计其中 70-80% 属于 LLVM/MLIR/XLA,因此真实占用会比 156.9 MB 略高。

---

## 交叉引用

- [Embedded-Library Atlas](embedded-library-atlas.md) — 负责完整的约 60 个库静态链接树和逐库字节核算;本页是以 LLVM/MLIR 为中心的切片。
- [Binary Forensics Overview](overview.md) — forensics 层的顶部;构建身份和分析方法。
- [ELF Anatomy](elf-anatomy.md) — section 大小、大代码模型 `.lrodata`/`.lbss`,以及支撑此处每个偏移的 build-id note。
- [Custom Sections](custom-sections.md) — `google_malloc`、`protodesc_cold`、`__rseq_cs` 和其他界定库桶的编译器发出 section。
- [The libtpu.so / sdk.so Two-Binary Split](two-binary-split.md) — 两个随附对象中哪一个承载 LLVM/MLIR 工具链。
- [The TPU Compiler](../compiler/overview.md) — 驱动这里编目的 LLVM/MLIR 组件的编译 pipeline。
- [LlvmTpu Intrinsic Catalog](../compiler/llvmtpu-intrinsic-catalog.md) — LLVM TPU 后端 lowering 的 intrinsic。
- [The tpu MLIR Dialect: Ops and the Op-Model Contract](../compiler/tpu-dialect-and-ops.md) — 深入说明 Group-C `tpu` 方言。
- [LowerToSparseCoreLlvm](../compiler/lower-to-sparsecore-llvm.md) — 把 MLIR 交给 LLVM TPU 后端的 `translateModuleToLLVMIR` AOT 路径。