Skip to content

设计决策

本文档只保留仍然直接影响 CUDA 实现或公开集成面的核心 ADR,删除了与外部流程框架绑定的历史治理层。


ADR-1:head_dim 仅支持 32 / 64 / 128

背景
Kernel 的 tiling、共享内存占用和寄存器压力都依赖编译期常量。

决策
仅支持 3264128 三个 head_dim,并通过显式模板实例化进行分发。

影响

  • 共享内存与启动参数更可预测
  • 测试面更集中
  • 非标准 head_dim 需要由上层做 padding 或适配

ADR-2:FP16 路径内部统一使用 FP32 累加

背景
Attention 分数累加和 online softmax 在半精度下容易出现精度损失与数值不稳定。

决策
GMEM 边界可使用 half,但内部累加、归一化和关键中间值统一使用 float

影响

  • 长序列下更稳定
  • 内部资源压力略增
  • FP16/FP32 两条路径的行为更一致

ADR-3:Python 集成停留在 C ABI 层

背景
仓库强调轻量、可嵌入和低依赖,不希望为绑定层引入大型运行时。

决策
保留 C ABI,并以 ctypes 风格的使用方式作为官方文档路径,而不是引入 pybind11 一类框架。

影响

  • 依赖面更小
  • 更容易被多语言复用
  • 运行时类型检查不如高层绑定框架友好

ADR-4:核心线程块大小固定为 128

背景
当前 occupancy、warp 分工和 shared memory tiling 都围绕 128 线程块做了收敛。

决策
继续保持 blockDim.x = 128,并与显式启动配置校验配套。

影响

  • 调度与性能特征更稳定
  • 维护成本更低
  • 若未来调整,需要重新调优并完整回归验证

ADR-5:仓库文档优先,移除外部流程框架

背景
仓库曾叠加多层 AI / 流程控制框架,导致说明重复、约束冲突、维护成本上升。

决策
把维护规则收敛到仓库原生文档:READMECONTRIBUTING、GitHub Actions 与 VitePress 页面;移除外部控制层。

影响

  • 文档与代码更容易保持一致
  • 删除了冗余流程负担
  • 后续维护者只需阅读仓库本身即可理解项目

Stable v0.3.0 baseline. Lean CUDA FlashAttention reference.