sskj/docs/KIMI_K3_DEPLOY.md

# Kimi-K3 部署手册（运维版）

> 面向运维的一键部署手册：按顺序复制粘贴命令即可完成 Kimi-K3 在 4 节点
> RTX 6000D 集群上的部署。**不需要理解 SGLang 引擎参数。**
> 部署配置统一在 `deploy/profiles/pro6000/kimi3_pro6000_sglang_tp32ep32.env`，
> 压测统一走 `python -m sskj.bench`（见 `ops/README.md`）。

## 0. 概述

| 项 | 值 |
|---|---|
| 模型 | Kimi-K3（Moonshot AI，2.8T 参数 MXFP4 量化，~1.5TB） |
| 节点 | 6000D-5~8 = 174.1.60.5~8（每节点 8× RTX 6000D 85GB，共 32 卡） |
| 镜像 | `lmsysorg/sglang:kimi-k3`（专用镜像，普通 sglang 镜像不识别该模型架构） |
| 并行 | TP=32 × EP=32（跨 4 节点） |
| 服务地址 | `http://174.1.60.5:30000`（OpenAI 兼容） |
| 模型 ID | `kimi-k3` |
| 部署方式 | `python -m sskj.deploy start --profile pro6000/kimi3_pro6000_sglang_tp32ep32` |

## 1. 前置检查（每步都过再继续）

```bash
# 1.1 4 节点 8 卡全部空闲（应全部显示 0 MiB / 0%）
for i in 5 6 7 8; do echo "== 174.1.60.$i"; ssh 174.1.60.$i 'nvidia-smi --query-gpu=index,memory.used --format=csv,noheader'; done
# 注意：如显示被占用，联系模型团队确认（其他用户的作业需先让出）

# 1.2 磁盘空间（模型 1.5TB + 运行余量，/data 需 ≥ 2TB 可用）
ssh 174.1.60.5 'df -h /data'

# 1.3 RoCE 网卡就绪（应输出 4 个 ACTIVE 且 link_layer=Ethernet）
ssh 174.1.60.5 'cat /sys/class/infiniband/mlx5_*/ports/1/link_layer; cat /sys/class/infiniband/mlx5_*/ports/1/state'

# 1.4 节点互信（174.1.60.5 免密登录其余节点；若失败执行步骤 1.5）
ssh 174.1.60.5 'for i in 6 7 8; do ssh -o StrictHostKeyChecking=no 174.1.60.$i hostname; done'

# 1.5 若 1.4 失败，在 174.1.60.5 上执行（把 .5 的 root 公钥发到各节点）
ssh 174.1.60.5 'for i in 6 7 8; do ssh-copy-id -o StrictHostKeyChecking=no root@174.1.60.$i; done'
```

## 2. 模型下载（仅首次，~1.5TB，需数小时）

模型权重放 `/data/hf_models/Kimi-K3`（4 节点同一路径，本集群已就绪则跳过本节）。

```bash
# 在 174.1.60.5 上执行（国内走 ModelScope 最快；modelscope CLI 需先 pip install modelscope）
pip3 install -q modelscope
mkdir -p /data/hf_models
nohup modelscope download --model moonshotai/Kimi-K3 \
  --local_dir /data/hf_models/Kimi-K3 > /data/hf_models/download_k3.log 2>&1 &
# 查看进度
tail -f /data/hf_models/download_k3.log
```

下载完成后校验（96 个分片，共约 1.5TB）：

```bash
ssh 174.1.60.5 'ls /data/hf_models/Kimi-K3/*.safetensors | wc -l; du -sh /data/hf_models/Kimi-K3'
# 期望：96（safetensors 数量）、1.5T（总大小）
```

**分发到其余 3 台**（.5 上执行；源盘 NVMe 读是瓶颈，聚合约 3.5GB/s）：

```bash
ssh 174.1.60.5 'for i in 6 7 8; do rsync -aH --partial /data/hf_models/Kimi-K3/ 174.1.60.$i:/data/hf_models/Kimi-K3/ & done; wait'
```

> 每节点需约 1.5T 空闲磁盘（`df -h /data` 确认）。

## 3. 镜像准备（已拉取则跳过）

```bash
# 4 节点并行拉取（约 9.6GB；如 Docker Hub 直连失败，各机 daemon.json 需配镜像加速：
#   "registry-mirrors": ["https://docker.m.daocloud.io", "https://docker.xuanyuan.me"]）
for i in 5 6 7 8; do ssh 174.1.60.$i 'docker pull lmsysorg/sglang:kimi-k3' & done; wait

# 验证镜像存在
ssh 174.1.60.5 'docker images lmsysorg/sglang:kimi-k3 --format "{{.Repository}}:{{.Tag}} {{.Size}}"'
```

## 4. 补丁与依赖分发

Kimi-K3 镜像在消费级 Blackwell（sm_120，RTX 6000D）上有一个内核不兼容点，
需要打补丁（补丁源文件在本仓库 `platforms/patches/pro6000/kimi_k3/patch_k3_sm120.py`，
修复 attn_res 融合内核误用数据中心 Blackwell 专属的 tcgen05 指令问题）。

```bash
# 在 174.1.60.5 上，把补丁分发到 4 节点 /tmp（部署时容器自动挂载并执行）
ssh 174.1.60.5 '
  for i in 5 6 7 8; do
    scp /data/yy/sskj/platforms/patches/pro6000/kimi_k3/patch_k3_sm120.py 174.1.60.$i:/tmp/patch_k3_sm120.py
  done
  for i in 5 6 7 8; do ssh 174.1.60.$i "sha256sum /tmp/patch_k3_sm120.py"; done
'
# 4 台输出的 sha256 必须一致（内容校验）
```

## 5. 一键部署

```bash
cd /data/yy/sskj

# 5.1 先预览要执行的命令（不实际启动）
PYTHONPATH=src python3 -m sskj.deploy start \
  --profile pro6000/kimi3_pro6000_sglang_tp32ep32 --dry-run

# 5.2 正式启动（4 节点同时拉起，耗时 8~12 分钟：模型加载 + CUDA graph 捕获）
PYTHONPATH=src python3 -m sskj.deploy start \
  --profile pro6000/kimi3_pro6000_sglang_tp32ep32

# 5.3 查看状态
PYTHONPATH=src python3 -m sskj.deploy status \
  --profile pro6000/kimi3_pro6000_sglang_tp32ep32
```

就绪判定（满足其一）：
- 步骤 5.3 `status` 显示容器 Up 且健康检查通过
- 或手动验证：`curl -s http://174.1.60.5:30000/health` 返回 `{"status":"ok"}`

## 6. 验证服务

```bash
# 6.1 模型列表
curl -s http://174.1.60.5:30000/v1/models | head -c 300
# 应包含 "id": "kimi-k3"

# 6.2 推理冒烟（正确性：23×47 应算得 1081，输出可能带思考标签属正常）
curl -s -m 300 http://174.1.60.5:30000/v1/chat/completions \
  -H "Content-Type: application/json" \
  -d '{"model": "kimi-k3", "messages": [{"role": "user", "content": "What is 23*47? Answer briefly."}], "max_tokens": 128, "temperature": 0.1}'

# 6.3 服务日志（确认无 ERROR）
ssh 174.1.60.5 'docker logs kimi3_pro6000_sglang_tp32ep32_node0 2>&1 | tail -20'
```

> ⚠️ **首次请求很慢（TTFT 可达 100+ 秒）是正常现象**：Triton/FlashKDA 内核首次
> JIT 编译。发一个任意请求预热后，后续请求恢复正常（16K 输入 TTFT ~6s）。

## 7. 停止 / 重启

```bash
cd /data/yy/sskj

# 停止（4 节点容器全部移除）
PYTHONPATH=src python3 -m sskj.deploy stop \
  --profile pro6000/kimi3_pro6000_sglang_tp32ep32

# 重启 = 再次执行 5.2 的 start 命令（幂等，会先清旧容器）

# 手动兜底（start/stop 异常时）：4 节点各执行
for i in 5 6 7 8; do ssh 174.1.60.$i 'docker rm -f kimi3_pro6000_sglang_tp32ep32_node*'; done
```

## 8. 性能自检（可选）

```bash
# 稳态基准：16K 输入 TTFT / prefill 吞吐 / TPOT（先跑一次丢弃 JIT 冷启动）
cd /data/yy/sskj
scp -o ConnectTimeout=10 /data/flashkda_deploy/bench_16k_steady.py 174.1.60.5:/tmp/ 2>/dev/null || true
ssh 174.1.60.5 'python3 /tmp/bench_16k_steady.py 3'
# 参考值（稳态）：TTFT ~6s / prefill ~2700 tok/s / TPOT ~39ms / decode ~26 tok/s
```

正式压测（矩阵 + 自适应并发搜索）走 bench 层：

```bash
cd /data/yy/sskj/experiments/pro6000/kimi3_pro6000_sglang_tp32ep32
DRY_RUN=1 bash run_adaptive_concurrency_add16.sh        # 先看计划
bash run_adaptive_concurrency_add16.sh                  # 正式跑（放 tmux）
```

## 9. 故障排查速查

| 现象 | 原因 | 处理 |
|---|---|---|
| 首次请求 TTFT 100+ 秒 | Triton/FlashKDA 内核首次 JIT 编译 | 正常现象，预热一次即可 |
| `Not enough GPU memory for hybrid mamba state cache` | 某节点显存被其他作业占用 | 检查 4 节点 `nvidia-smi`，确认 8 卡全空闲后重启部署 |
| `ibv_create_cq failed: Cannot allocate memory` | 容器 memlock 限制 | 部署配置已带 `--ulimit memlock=-1`，检查 profile 未被改动 |
| `NCCL error: invalid usage`（graph 捕获时） | 误设了 `NCCL_ALGO=TREE` | 确认环境变量里没有 NCCL_ALGO |
| 启动后 `/health` 一直不通 | 分布式初始化失败/节点未就绪 | 看 master 日志；确认 4 节点 GPU 全空闲后 `stop` + 重新 `start` |
| 推理结果带 `<\|open\|>think` 标签 | K3 思考型模型正常输出格式 | 非故障；如需精简可调 chat template（联系模型团队） |
| 并发压测时部分请求失败 | flashkda 后端已知问题（未修复） | 当前默认 triton 后端无此问题；如误用 flashkda 见附录 |

## 附录 A：flashkda 可选后端（默认不启用）

FlashKDA（MoonshotAI CUTLASS KDA 内核，支持 sm_120）可作为 KDA prefill 的备选后端。
实测与默认 triton prefill **性能等价**（16K prefill 2693 vs 2695 tok/s），仅作备份用途。

启用方式（需模型团队协助）：
1. 构建 wheel（脚本见 `/data/flashkda_deploy/01_build_flashkda.sh`，产物
   `flash_kda-0.0.1-cp312-cp312-linux_x86_64.whl` 分发到 4 节点 `/tmp/`）
2. profile 增加 wheel 挂载与安装步骤，并在 LAUNCH_ARGS 加
   `--linear-attn-prefill-backend flashkda`

已知问题：flashkda 后端在 4/8 并发压测下服务异常（纯 triton 无此问题），修复前不建议生产使用。

## 附录 B：PD 分离部署（MoonCake RDMA，8 节点）

非 PD 单组部署之外的另一种形态：**Prefill/Decode 分离**（P 组 4 节点做 prefill，D 组 4 节点做 decode），KV 传输走 **MoonCake RDMA**。适合长上下文、吞吐优先的场景。

### B.1 拓扑

```
                    ┌───────────────────┐
  client ──31000──▶ │  router (MiniLB)  │ 174.1.60.5
                    └───────┬───────────┘
                    ┌───────▼───────────┐
  P 组 (prefill)   │  174.1.60.1~4      │  TP32×EP32，4 节点 32 卡
  └── prefill 计算  │  计算完产出 KV      │
                    └───────┬───────────┘
                    ┌───────▼───────────┐
      MoonCake RDMA│  mlx5_0~3 (RoCE)  │  4 链路并行
                    └───────┬───────────┘
                    ┌───────▼───────────┐
  D 组 (decode)    │  174.1.60.5~8      │  TP32×EP32，4 节点 32 卡
  └── decode 生成   │  接收 KV 逐 token  │
                    └───────────────────┘
```

### B.2 运维准备（一次性，新环境才需执行）

已部署过的集群可跳过本节。步骤顺序：

1. **免密与文件就位**：确保 8 节点互信、docker 可用；把 `patch_k3_sm120.py`、`flash_kda-*.whl`、mooncake wheel 分发到 8 节点（路径见附录 B.2 下方说明）。
2. **sskj 仓库**：`.5` 和 `.1` 都需 clone `/data/yy/sskj`（deploy_pd.sh 的 deploy 层命令依赖）。
3. **拉镜像**：8 节点 `docker pull lmsysorg/sglang:kimi-k3`。
4. **GPU 空闲**：`nvidia-smi` 全 0 MiB，否则联系模型团队确认。

详细命令见 `experiments/pro6000/kimi3_pro6000_pd_rdma/README.md` 的「运维一键部署」章节（含每步的可执行命令与验证）。

### B.3 一键部署

```bash
cd /data/yy/sskj/experiments/pro6000/kimi3_pro6000_pd_rdma
bash deploy_pd.sh start      # mc-master → P 组 → D 组 → router
bash deploy_pd.sh status
bash deploy_pd.sh stop
bash deploy_pd.sh restart
```

### B.4 验证

```bash
# 端到端（router 自动双发 P/D）
curl -s http://174.1.60.5:31000/generate -H "Content-Type: application/json" \
  -d '{"text":"Hello","sampling_params":{"max_new_tokens":16}}'
```

长输入验证（2026-08-11 实测）：1680 / 5280 tokens 输入均正常，P 组 prefill ~112 tok/s，D 组 decode ~26 tok/s。

### B.5 关键注意事项（务必遵守）

1. **禁止设置 `PYTORCH_CUDA_ALLOC_CONF=expandable_segments:True`**
   expandable_segments 分配的 GPU 段，mooncake RDMA 注册报 `Bad address [14]`
   （GitHub kvcache-ai/Mooncake **#2511**）。移除后 KV 注册从 2376 条失败 → 0。
2. **P 组必须先于 D 组启动**（prefill 需先注册 bootstrap）。
3. **容器内必须升级 mooncake 到 0.3.12.post1**（镜像自带 0.3.11.post1 无 dmabuf 修复 #2035）。
   wheel 在 `/data/flashkda_deploy/wheels/`，profile 的 BOOTSTRAP 已含 pip install。
4. NCCL/GLOO 走计算网（`bond1` + `NCCL_IB_HCA=mlx5_0..3`），不要走管理网 bond0。
5. mc-master（mooncake 元数据服务）需先在 174.1.60.1 运行（deploy_pd.sh 自动处理）。

### B.6 与单组部署的取舍

| 维度 | 单组 TP32×EP32 | PD 分离（MoonCake RDMA） |
|---|---|---|
| 节点 | 174.1.60.5~8（4 节点） | P 174.1.60.1~4 + D 174.1.60.5~8（8 节点） |
| KV 传输 | 无（本地） | MoonCake RDMA（计算网 4 链路） |
| 传输后端 | - | mooncake（必须 0.3.12.post1） |
| 适用 | 单组吞吐、简单 | 长上下文、PD 分离 |
| profile | kimi3_pro6000_sglang_tp32ep32 | kimi3_pro6000_pd_prefill / _pd_decode |
| 编排 | `python -m sskj.deploy` | `deploy_pd.sh` |