""" ============================================================================= CS349E · 第 1 讲 · 01_energy_and_scaling.py —— 能量视角：为什么"搬数据"比"算"贵，为什么要专用硬件 ============================================================================= 【这个脚本要回答的问题】 今天所有 AI 芯片（GPU、TPU、各种推理加速器）长成这个样子，背后有一条主线： **芯片的功耗预算是固定的，所以性能 = 能效 × 功率**。想更快，就得让每次运算花更少的能量。 这个脚本用 Mark Horowitz 在 ISSCC 2014 给出的能量表（45 nm 工艺），一步步算清楚： 演示 1：各种操作的能量——运算很便宜，从 DRAM 读数据贵几百到几万倍。 演示 2：要做多少次运算才能"摊平"一次数据搬运的能量？（能量版的"算术强度"） 演示 3：Dennard scaling 的终结——为什么频率不再上涨、芯片出现"暗硅"。 演示 4：一条指令的"管理开销"——SIMD、GPU 的 warp、Tensor Core 怎样把它摊薄。 演示 5：把以上合在一起：一次 8 位乘加的总能量，在 CPU、GPU、专用阵列上差多少。 【怎么读输出】每个演示先打印表格，再打印"观察 Observations"。所有数字都是用能量表算出来的， 每次运行完全相同。45 nm 的绝对数字比今天的工艺大几倍，但**比例**是这节课要记住的。 【运行】python3 01_energy_and_scaling.py （不到 1 秒） 【依赖】只用 Python 标准库；从同目录导入 hw_specs、console。 """ from console import print_header, print_observations, print_table from hw_specs import ENERGY_45NM_PJ as E, ENERGY_LABELS # --------------------------------------------------------------------------- # 演示 1：能量表 # --------------------------------------------------------------------------- def demo1_energy_table(): """ 打印 Horowitz 能量表，并把每一项换算成"相当于多少次 8 位加法 / 32 位浮点乘法"。 【为什么要换算】绝对数字（0.03 pJ、1300 pJ）很难有感觉，换成倍数之后一眼就能看出： 一次 DRAM 读取 ≈ 几百次 32 位浮点乘法 ≈ 几万次 8 位加法。 【边界】内存类条目是"读 64 bit（8 字节）"的能量，我们另外算出"每字节"多少 pJ， 方便后面和"每次运算"比较。 """ print_header("演示 1：一次操作要花多少能量？（Horowitz, ISSCC 2014, 45 nm 0.9 V）") rows = [] for key, pj in E.items(): per_byte = f"{pj / 8:.2f}" if "read64" in key else "" rows.append([ENERGY_LABELS[key], f"{pj:g}", per_byte, f"{pj / E['int8_add']:,.0f}×", f"{pj / E['fp32_mult']:.2f}×"]) print_table(["操作", "pJ", "pJ/字节", "÷ 8 位加法", "÷ 32 位浮点乘法"], rows) dram_lo, dram_hi = E["dram_read64_low"], E["dram_read64_high"] print_observations([ f"读 8 字节 DRAM 要 {dram_lo:g}–{dram_hi:g} pJ，是一次 32 位浮点乘法（{E['fp32_mult']} pJ）的 " f"{dram_lo / E['fp32_mult']:.0f}–{dram_hi / E['fp32_mult']:.0f} 倍，是一次 8 位加法的 " f"{dram_lo / E['int8_add']:,.0f}–{dram_hi / E['int8_add']:,.0f} 倍。", f"片上 SRAM 也分大小：8 KB 的读一次 {E['sram_8KB_read64']:g} pJ，1 MB 的 {E['sram_1MB_read64']:g} pJ——" "存储越大，线越长，能量越高。这就是为什么芯片要做多级存储层次（寄存器 → 小 SRAM → 大 SRAM → DRAM）。", f"位宽影响很大：32 位整数乘法 {E['int32_mult']} pJ，8 位只要 {E['int8_mult']} pJ（{E['int32_mult'] / E['int8_mult']:.1f} 倍）。" "乘法器面积和能量大致随位宽的平方增长——这是低精度推理（第 3 讲）的物理基础。", f"一条最简单的指令总共 {E['instr_total']:g} pJ，其中真正'做加法'只占零头：" "取指令、读寄存器、控制逻辑才是大头（演示 4）。", ]) # --------------------------------------------------------------------------- # 演示 2：能量版的"算术强度" # --------------------------------------------------------------------------- def demo2_energy_breakeven(): """ 算出：每从某一级存储读 1 字节，要做多少次运算，运算本身的能量才和"读这 1 字节"的能量相当。 【思路】设一次运算的能量是 e_op（pJ/FLOP），读 1 字节的能量是 e_byte（pJ/byte）， 做 I 次运算/字节时：运算能量 = I·e_op，搬运能量 = e_byte。 二者相等时 I* = e_byte / e_op。I < I* 时总能量被"搬运"主导——和 roofline 的 "临界强度"完全同构，只不过把"时间"换成了"能量"。 【乘加怎么算】一次乘加（MAC）= 1 次乘法 + 1 次加法 = 2 FLOP，所以 e_op = (e_mult + e_add) / 2。 【边界】DRAM 用能量表给的低、高两个值，所以结果是一个区间。 """ print_header("演示 2：做多少次运算，才能'摊平'读 1 字节的能量？") mac_kinds = { "FP32 乘加": (E["fp32_mult"] + E["fp32_add"]) / 2, "FP16 乘加": (E["fp16_mult"] + E["fp16_add"]) / 2, "INT8 乘加": (E["int8_mult"] + E["int8_add"]) / 2, } sources = { "8 KB SRAM": (E["sram_8KB_read64"] / 8, E["sram_8KB_read64"] / 8), "1 MB SRAM": (E["sram_1MB_read64"] / 8, E["sram_1MB_read64"] / 8), "DRAM": (E["dram_read64_low"] / 8, E["dram_read64_high"] / 8), } def fmt_i(x): """小于 10 的数保留 1 位小数，否则取整并加千位分隔符。""" return f"{x:.1f}" if x < 10 else f"{x:,.0f}" rows = [] for mac_name, e_op in mac_kinds.items(): cells = [mac_name, f"{e_op:.3f}"] for src, (lo, hi) in sources.items(): cells.append(fmt_i(lo / e_op) if lo == hi else f"{fmt_i(lo / e_op)}–{fmt_i(hi / e_op)}") rows.append(cells) print_table(["运算类型", "pJ/FLOP"] + [f"I*（{s}）" for s in sources], rows) int8 = mac_kinds["INT8 乘加"] print_observations([ "表里的 I* 单位是 FLOP/byte：每从这一级存储读 1 字节，至少要做这么多次运算，运算能量才'追上'搬运能量。", f"对 INT8 乘加，DRAM 的 I* 是 {E['dram_read64_low'] / 8 / int8:,.0f}–{E['dram_read64_high'] / 8 / int8:,.0f}。" "Horowitz 在论文里说，要达到最高能效，'大约每次 DRAM 读取要完成上千次运算'——就是这个数量级。", "精度越低，运算越便宜，I* 越高：低精度让'算'变便宜了，但'搬'没变，数据复用（locality）就更关键。", "和下一个脚本的 roofline 对照：H100 的'时间'临界强度约 295 FLOP/byte，和这里'能量'的 I* 是同一个量级。" "芯片设计者在同时平衡这两本账。", ]) # --------------------------------------------------------------------------- # 演示 3：Dennard scaling 与暗硅 # --------------------------------------------------------------------------- def demo3_dennard(S=1.4, generations=6): """ 用一个玩具模型对比"Dennard 时代"和"后 Dennard 时代"每一代工艺的功耗密度变化。 【模型】动态功耗 P = C·V²·f（C 是电容，V 是电压，f 是频率），每代工艺把线宽缩小到 1/S（S ≈ 1.4， 即面积减半、同面积晶体管数 ×S² ≈ 2）。每个晶体管：C ∝ 1/S。 - Dennard 时代（约 1974–2005）：V ∝ 1/S，f ∝ S 每管功耗 ∝ (1/S)·(1/S²)·S = 1/S²；单位面积晶体管 ×S² → 功耗密度不变。**变快、变多、不变热。** - 后 Dennard 时代：漏电让电压降不下去（V 基本不变），频率也不涨了（f 不变） 每管功耗 ∝ (1/S)·1·1 = 1/S；单位面积 ×S² → 功耗密度 ×S 每代。 如果还想让频率继续 ×S，功耗密度就 ×S² 每代。 芯片能散的热（W/cm²）是有上限的，所以功耗密度超出多少倍，就只能让多少比例的晶体管同时工作—— 剩下的就是"暗硅"（dark silicon）。 【参数】S：每代线宽缩小倍数；generations：模拟多少代 【边界】这是教科书式的一阶模型，忽略了漏电功耗、3D 晶体管等细节；只用来看趋势。 """ print_header(f"演示 3：Dennard scaling 的终结（每代线宽 ÷{S}，同面积晶体管 ×{S * S:.2f}）") rows = [] for g in range(generations + 1): transistors = (S * S) ** g dennard = 1.0 post_fixed_f = S ** g post_rising_f = (S * S) ** g rows.append([g, f"{transistors:.1f}×", f"{S ** g:.2f}×", f"{dennard:.2f}×", f"{post_fixed_f:.2f}×", f"{1 / post_fixed_f * 100:.0f}%", f"{post_rising_f:.1f}×"]) print_table(["代", "晶体管数", "Dennard 频率", "Dennard 功耗密度", "后 Dennard 功耗密度（f 不变）", "能同时工作的比例", "若 f 继续涨"], rows) print_observations([ "Dennard 时代：每一代晶体管翻倍、频率 ×1.4，功耗密度却不变——性能是工艺'白送'的。", "大约 2005 年以后电压降不下去了：即使频率不再上涨，功耗密度也每代 ×1.4；" f"{generations} 代之后同样的散热能力只能让 {1 / S ** generations * 100:.0f}% 的晶体管同时工作。", "这就是 2005 年以后 CPU 频率停在 3–5 GHz、转向多核的原因，也是'暗硅'的来源：晶体管还在变多，但不能同时用。", "既然不能同时点亮所有晶体管，最划算的做法是把它们做成'只干一类活、但干得极省电'的专用电路——" "这就是 GPU 的 Tensor Core、TPU 的脉动阵列、各种推理芯片出现的根本原因。", ]) # --------------------------------------------------------------------------- # 演示 4：指令开销的摊薄 # --------------------------------------------------------------------------- def demo4_instruction_amortization(): """ 一条指令的固定开销（取指 + 译码 + 寄存器 + 控制，约 70 pJ）如果只换来 1 次运算，就太浪费了。 让一条指令做更多运算，就能把开销摊薄。 【模型】每次有用运算的能量 = e_op + 70 pJ / (每条指令完成的运算数) - 标量 CPU：一条指令 1 次运算 - SIMD（如 AVX-512 的 16 个 FP32 lane）：一条指令 16 次 - GPU warp：32 个线程执行同一条指令（SIMT），取指一次 - Tensor Core（Ampere 的 mma.m16n8k16）：一条 warp 指令完成 16·8·16 = 2048 次乘加 = 4096 FLOP - Hopper 的 wgmma.m64n256k16：一个 warpgroup 一条指令完成 64·256·16 = 262,144 次乘加 【简化】这里把 70 pJ 当成"每条指令的固定开销"套在所有架构上，只为看数量级；GPU 的真实指令 开销和 CPU 不同，而且操作数从寄存器 / 共享内存送到计算单元也要花能量（Tensor Core 靠 矩阵内部的复用把这部分也摊薄了）。运算统一用 FP16 乘加的能量。 """ print_header("演示 4：一条指令做多少活？——把'管理开销'摊薄") e_op = (E["fp16_mult"] + E["fp16_add"]) / 2 # 每 FLOP 的能量（FP16 乘加 / 2） overhead = E["instr_total"] designs = [ ("标量 CPU（1 FLOP/指令）", 1), ("SIMD 16 lane FMA（32 FLOP）", 32), ("GPU warp 32 线程 FMA（64 FLOP）", 64), ("Tensor Core mma.m16n8k16（4,096 FLOP）", 2 * 16 * 8 * 16), ("Hopper wgmma.m64n256k16（524,288 FLOP）", 2 * 64 * 256 * 16), ] rows = [] for name, ops in designs: per_op = e_op + overhead / ops rows.append([name, f"{per_op:.3f}", f"{overhead / ops / per_op * 100:.1f}%", f"{per_op / e_op:.2f}×"]) print_table(["设计", "pJ / 有用 FLOP", "开销占比", "÷ 纯运算能量"], rows) print_observations([ f"标量执行时，每次 FP16 运算真正的能量只有 {e_op:.2f} pJ，却要背 {overhead:g} pJ 的指令开销——" f"{(e_op + overhead) / e_op:.0f} 倍的浪费。", "SIMD 和 GPU 的 SIMT 用'一条指令驱动很多个计算单元'把开销降到几十分之一；Tensor Core 一条指令做几千次乘加，" "开销几乎可以忽略。", "Horowitz 论文里的原话：浮点运算大约只有一条简单指令能量的 1/10，所以做一个约 10 lane 的 SIMD 引擎，" "指令能量就被运算主导了——'这就是 GPU 走的路'。", "代价是灵活性：一条指令必须对一大块规整的数据做同样的事。AI 的矩阵乘法恰好满足这个条件。", ]) # --------------------------------------------------------------------------- # 演示 5：总账：一次 8 位乘加在三种架构上 # --------------------------------------------------------------------------- def demo5_total_energy(): """ 把"运算 + 指令开销 + 片上存储 + DRAM"四项加起来，看一次 INT8 乘加的总能量。 【模型】每次乘加的总能量（pJ）= e_mac + E_instr / r_instr + E_sram / r_sram + E_dram / r_dram r_x 表示"每做多少次乘加才需要一次 x"（复用因子）： - 通用 CPU：每次乘加都要一条指令，每次都从 32 KB 的 L1 cache 取操作数，每 10 次乘加碰一次 DRAM - GPU 风格（SIMT + Tensor Core）：指令摊到 ~1000 次乘加，片上 SMEM 复用 ~16 次，DRAM 复用 ~300 次 - 专用脉动阵列（TPU 风格）：指令摊到 ~10,000 次，数据在阵列里流动、复用 ~128 次，DRAM 复用 ~1000 次 DRAM 一次读 8 字节，INT8 乘加要读 1 字节权重，所以每字节 DRAM 能量 = dram_read64 / 8。 【简化】复用因子是"有代表性的假设"，不是测量值，只用来说明数量级；真实芯片的能效还取决于工艺、 电压、利用率等很多因素。DRAM 取低值 1.3 nJ / 64 bit。 """ print_header("演示 5：一次 INT8 乘加的总能量（运算 + 指令 + 片上存储 + DRAM）") e_mac = E["int8_mult"] + E["int8_add"] e_sram = E["sram_32KB_read64"] / 8 # 每字节 e_dram = E["dram_read64_low"] / 8 # 每字节 designs = [ ("通用 CPU", 1, 1, 10), ("GPU 风格", 1000, 16, 300), ("专用脉动阵列", 10000, 128, 1000), ] rows = [] totals = {} for name, r_instr, r_sram, r_dram in designs: parts = [e_mac, E["instr_total"] / r_instr, e_sram / r_sram, e_dram / r_dram] total = sum(parts) totals[name] = total rows.append([name, f"{r_instr:,}/{r_sram}/{r_dram:,}"] + [f"{p:.3f}" for p in parts] + [f"{total:.3f}"]) print_table(["架构", "复用 指令/SRAM/DRAM", "运算", "指令", "SRAM", "DRAM", "合计 pJ"], rows) print_observations([ f"通用 CPU 每次乘加约 {totals['通用 CPU']:.0f} pJ，几乎全花在指令和数据搬运上，真正的乘加只占 {e_mac:.2f} pJ。", f"同样的乘加在专用阵列上约 {totals['专用脉动阵列']:.2f} pJ，相差约 {totals['通用 CPU'] / totals['专用脉动阵列']:.0f} 倍——" "和 Horowitz 论文图 1.1.8 里'专用硬件比处理器高效 2–3 个数量级'一致。", "专用阵列剩下的能量里，DRAM 仍然占不小的比例：复用因子决定一切。这就是第 9 讲（硬件性能建模，Interstellar）" "要系统地研究'怎样安排循环才能让数据在便宜的存储里被复用最多次'。", "对 LLM decode 来说，权重的 DRAM 复用因子 ≈ 批大小 B（每读一次权重服务 B 个 token）。B = 1 时它就像" "'每次乘加都读一次 DRAM'——下一个脚本会从时间的角度看到同样的问题。", ]) if __name__ == "__main__": demo1_energy_table() demo2_energy_breakeven() demo3_dennard() demo4_instruction_amortization() demo5_total_energy()