- title
- SMPDC Paper 08|CPU–GPU 異質協作:並行資訊流、工作切分、設備同步與效能邊界
- english_title
- CPU–GPU Heterogeneous Collaboration: Parallel Information Flow, Work Partitioning, Device Synchronization, and Performance Boundaries
- series
- SMPDC — Stellar Multi-Primary Persistent Diffusive Computing
- paper
- 08 / 11
- version
- 0.1
- status
- Full Research Manuscript — Candidate Heterogeneous Execution Contract; Conditional Claims and Reproducible Experiment Design; Real GPU Benchmark Pending
- date
- 2026-10-17
- language
- zh-TW
- primary_concept_originator
- Neo.K
- research_collaboration
- Aletheia / GPT-6
- institution
- EveMissLab / 一言諾科技有限公司
- canonical_source
- UTF-8 Markdown; inline math $...$ ; display math $$...$$
- predecessors
- SMPDC Paper 00 v0.1 (2026-10-16) · SMPDC Paper 01 v0.1 (2026-10-16) · SMPDC Paper 02 v0.1 (2026-10-16) · SMPDC Paper 03 v0.1 (2026-10-16) · SMPDC Paper 04 v0.1 (2026-10-16) · SMPDC Paper 05 v0.1 (2026-10-16) · SMPDC Paper 06 v0.1 (2026-10-16) · SMPDC Paper 07 v0.1 (2026-10-16)
- artifact_scope
- CPU–GPU execution mapping, GPU operator admission, host/device visibility, bounded batching, stream/event dependencies, memory leases, GPU failure and cancellation, typed protocol bridging, fair baselines, falsification
SMPDC Paper 08|CPU–GPU 異質協作
——並行資訊流、工作切分、設備同步與效能邊界
英文題名: CPU–GPU Heterogeneous Collaboration: Parallel Information Flow, Work Partitioning, Device Synchronization, and Performance Boundaries
研究系列: 群星多主節點持續擴散計算論(SMPDC)
主構想提出: Neo.K
研究協作: Aletheia/GPT-6
機構: EveMissLab/一言諾科技有限公司
日期/版本: 2026-10-17/v0.1
研究成熟度: L1,形式契約、條件命題、反例及驗證規格已建立;未宣稱已完成 SMPDC GPU Runtime、真實 CUDA benchmark、AI 訓練加速或跨平台重現。
直接承接: SMPDC Paper 00–07 [I0–I7]。
公開文獻查核: CUDA Programming Guide、CUDA Best Practices、CUDA Graphs、Nsight Systems、Legion、Kokkos 與 NCCL,查核至 2026-10-17 [E1–E12]。
作者: Neo.K
日期: 2026-10-17
版本: 0.1
摘要
SMPDC 前八篇(Paper 00–07)提出一個跨尺度但不混同尺度的研究綱領:多個 CPU 核心可承載各自具有持續身分與局部自治責任的主要工作流(Primary Workflow Node, PWN);這些工作流依可判定的相位資格建立耦合與有限擴散;訊息可採 R0 單路、R1 競速、R2 具來源條件的佐證,以及 R3 編碼冗餘;局部結果的驗證、推測狀態、權威提交與外部效果分別接受不同規範;PDAIS 抽象指令、SMPDC-Lang 編譯器及 OS/AI-free 控制層使這些約束具有軟體可執行的候選介面。本篇處理其中尚未建立完整實體映射的核心問題:CPU 持續主工作流如何在不失去局部主導責任的情況下,利用 GPU 加速大量可平行的資訊關係運算,同時保留正確的資料搬移、設備同步、證據血緣、資源所有權及正式提交邊界?
研究動機來自「CPU 核心不是 GPU 的單純前端,GPU 亦非 CPU 核心之間的物理高速導線」的兩重判斷。CPU 核心具有低延遲控制、複雜分支與本地工作流持續性的價值;GPU 適合能形成足夠批量、資料局部性合宜、平行性高的數值、圖關係、候選比較、聚合與部分編碼運算。GPU 可以加速在裝置上處理關係,但不會藉此消除 CPU 快取一致性、PCIe、主記憶體、資料重排及裝置同步的物理成本。尤其在 Windows 上,統一記憶體的共享語法不等於任意時刻主機與裝置可同時安全存取 [E2–E3]。
本篇提出 SMPDC GPU Collaboration Contract v0.1(縮寫 GCC08,僅限本文命名空間),將 GPU 協作拆成九個明確的邊界:工作切片資格、受限批次成形、可審計 offload 准入、資料與位址空間描述、BufferLease 所有權、stream/event 因果順序、裝置輸出的候選證據、CPU 主工作流的局部聚合、提交域的最終仲裁。定義 HeterogeneousWorkSliceV0_1、GpuBatchPlanV0_1、GpuExecutionTicketV0_1、DeviceBufferLeaseV0_1、DeviceVisibilityWitnessV0_1、GpuCandidateOutcomeV0_1 與 GpuCollaborationTraceV0_1 等候選契約,說明 Submitted、KernelDone、CopyDone、HostReadable、LocalVerified 和 Committed 之間的不可省略關係。
本文提出在明確前提下成立的安全及成本命題:只有保存版本與所有權守衛、將所需 copy/event 依賴納入同一因果偏序、以實體 buffer lease 綁定記憶體生命週期,才能宣稱裝置完成後的資料可被安全讀取;只有當 GPU 路徑的總端到端成本,在相同品質及資源限制下低於最佳成熟 CPU 或現成 GPU 基線,才能宣稱異質協作取得淨收益。為避免自我證成,本篇強制與傳統 CPU thread pool、CPU SIMD/平行演算法、成熟 CUDA streams、CUDA Graphs、kernel fusion,以及在適用情況下的現有 GPU library 比較;並設計反例、消融、故障注入與公平計量。第一輪實驗可在既有舊款 AMD CPU、32 GB RAM 與 RTX 3070 的桌機上進行,確切裝置能力由執行環境查詢,不把尚未購買的新 CPU/AI NAS 當成現有設備。
本篇的貢獻是候選跨協議的 CPU–GPU 執行契約與可證偽實驗架構,不是 GPU 平行運算、資料流、非同步 stream、CUDA Graphs、記憶體一致性或神經形態計算本身的發明;在沒有真實量測前,不宣稱任何通用加速比或 AI 訓練速度提升。
關鍵詞: CPU–GPU 異質計算、多主工作流、相位擴散、資料局部性、非同步 stream、CUDA event、CUDA Graphs、host/device 可見性、pinned memory、GPU 批次聚合、VRAM lease、冗餘資訊流、權威提交、端到端效能、可證偽研究。
1. 問題設定與論文邊界
1.1 「GPU 使連結更快」必須有明確工程意義
原始研究構想將各 CPU 大核/小核視為可能承擔主要工作流的節點,並提出在其下層以 GPU 大量並行運算協助節點連結及資訊流速。若把「連結」直接當成實體導線或把 GPU 描述成能跨越 CPU 的共享快取與片上互連,便會造成錯誤的硬體推論。本文將該構想精確化為三類不同操作:
- 資訊關係評估:對許多
PWN_i -> PWN_j候選邊計算權重、可達性、關聯分數、局部圖變換或傳遞優先級。 - 可平行的候選資料處理:對大量冗餘候選執行格式檢查、向量比較、算術聚合、索引化編碼/解碼子程序;實際來源認證與提交權限仍由可信端決定。
- 可重用的數值工作片段:AI 中的相似度矩陣、稀疏/分塊運算、批量特徵轉換與某些專家路由的數值核心。
它們可在適用條件下交由 GPU;CPU–CPU 實際訊息傳送依然遵循所使用的 host memory、cache-coherence 與 OS/Runtime 機制。邏輯上可畫為圓形、星系或高階圖,不要求晶片微觀物理拓撲與圖同構 [I0]。
1.2 五個研究問句
- RQ08-1: 在何種 batch 大小、算術強度、資料常駐率及相位截止時間下,GPU 候選取得正的端到端收益?
- RQ08-2: 如何在多個 PWN 與單一 GPU 之間建立公平、低競爭的批次/stream/buffer 協作,不讓 CPU 核心失去主工作流自治?
- RQ08-3: 如何以型別/Runtime guard 區分提交到 GPU、kernel 完成、copy 完成、host 可讀、局部驗證與正式提交?
- RQ08-4: R0–R3、推測分支、owner 遷移與設備故障如何影響 GPU 工作責任、證據血緣與安全回收?
- RQ08-5: 在完整核算排程、搬移、重排、同步、記憶體、驗證及能源後,相較成熟 CPU/CUDA 基線能否找到可重現的優勢適用域?
1.3 本篇不處理及不預設的事
本篇不重新制定 Paper 03 的證據真實性政策,不改寫 Paper 04 的權威提交域,不把每個 PWN 強制綁定固定實體核心,不假設特定 GPU 一定支援某種 async engine 或 memory feature,不重新發明 CUDA,不把 unified virtual address space 當成零拷貝或跨處理器強一致性保證,不預設所有 AI 訓練/推論可被 GPU 再加速,也不宣稱新 ISA 已存在。更低階的 CPU cache、NoC、CCD、NUMA、PCIe 實體映射與可能的新硬體需求保留給 Paper 09 [I0, I7]。
1.4 研究成熟度
| 層級 | 對本篇而言的證據 | 不允許的推論 |
|---|---|---|
| L0 概念 | CPU PWN + GPU 並行關係處理的候選拓撲 | 畫出 GPU 就代表更快 |
| L1 契約 | 本篇規則、條件命題、反例與基準設計 | 已完成實體效能驗證 |
| L2 參考實作 | Mock adapter/CUDA adapter 可重演、故障注入 PASS | 已證明所有 GPU 平台安全 |
| L3 單機實測 | 具硬體 manifest、公平基線與重複原始資料 | 適用於全部 AI 工作負載 |
| L4 跨平台複現 | 不同 CPU/GPU/OS 重現,界定適用域 | 普遍優於既有運算架構 |
Paper 08 v0.1 的成熟度為 L1。文中的數值範例均標為合成/解析示例,不是 RTX 3070 實測。
2. 對 SMPDC Paper 00–07 的不變更承接
2.1 九元系統及 GPU 作為外部 executor
沿用 Paper 00 的中觀狀態:
其中 映射 PWN 至 CPU 可用承載或未承載狀態;GPU 不是任意變成 裡的一顆物理 CPU core。本文新增的是具有投影性質的 GPU executor state ,不是修改 Paper 00 的九元 canonical source:
其中 為設備能力快照, 為記憶體 buffer 與 lease, 為已准入 work slice, 為 stream/event 依賴, 為設備候選結果, 為效能與錯誤觀測。完整系統仍由原九元狀態規範其邏輯權限, 是異質後端的具體化狀態。
2.2 Paper 01:PWN 身分不可綁死 GPU queue
flow_id、worker_id、physical_core_id、gpu_stream_id、device_id 是不同識別。PWN 可以為某工作片段請求 GPU,甚至在 GPU 持續執行時移往另一 OS worker;但 owner 世代與尚未完成的 GPU 義務必須被正確交接。GPU 的 stream 不是該 PWN 的正式身分,也不授予更高的權威 [I1]。
2.3 Paper 02:擴散 credit 不得因 GPU 批次而增加
DiffusionEnvelopeV0_1 的 message_id、lineage_id、delivery_attempt_id、event_epoch、authority_scope 與不可複製的傳播 credit 仍然有效。把 100 條訊息壓成一個 GPU buffer,不會自動產生新的 100 次邏輯傳播權;同一批次拆成三條 CUDA streams,也不是新 epoch 或無限免費的新傳送 [I2]。
2.4 Paper 03:GPU 只參與局部聚合與候選產生
R0/R1/R2/R3 表示不同的證據/冗餘義務,而不是 GPU 中四種算術 kernel。GPU 可加速 R1 的某些候選比較、R2 的大批特徵/規則計算、R3 的數值編碼,但GPU 產生相同 digest 不意味兩候選來源獨立。最終對 LocalReconciliationOutcomeV0_1 的 seal 必須保留來源、版本、獨立性與品質政策 [I3]。
2.5 Paper 04:局部計算與權威提交分離
GPU 的工作結果最強只能先成為被證明具必要可見性的候選值或暫態依賴。GpuCandidateOutcome 不能直接轉型成 Committed。若源快照過期,Paper 04 的 CommitProposalEnvelopeV0_1 必須重新驗證;CUDA event 成功也不改變其規則 [I4]。
2.6 Paper 05–06:必須語義保持地降階
PDAIS 的 GPU_* 家族、SPEC_ATTACH_GPU、RECONCILE_*、COMMIT_* 及 SMPDC-Lang 的 GPUEvent、GPUBufferLease、HostVisible、Tentative 都是本篇的語義入口。CUDA 呼叫只是其中一個後端映射。不得以 GPU opcode 或 memcpy 的可用性,取消原編譯器附著的效果/所有權檢查 [I5–I6]。
2.7 Paper 07:Control Plane 建議,不取代資源或權威
本篇維持 GPUAdapterControlV0_1:submit_candidate、query_event、obtain_host_visibility、release_lease 及 estimated_device_cost。Paper 08 只擴充資料平面與事件生命週期契約,沒有把 AI 估計、OS CPU Sets 意圖、CUDA 成功碼或 GPU kernel 的裝置權限提升成 canonical state 的提交權 [I7]。
3. 三種「快」:吞吐、關鍵路徑與傳輸
3.1 吞吐優勢不是單次訊息低延遲
GPU 擅長大量可同時執行且控制流相近的操作。然而單條 CPU 工作流需要向另一個 CPU 工作流傳遞一個非常小的布林判定時,放到 GPU 往返可能只增加排隊、傳送與同步。SMPDC 必須同時記錄每秒有效候選產出量、端到端延遲分布、截止時間錯過率及每單位能源的合法完成工作量。單一指標更好不代表整體更好。
3.2 GPU「並行連結」的合法語義
令 是當次待評估的關係邊集合,運算子 將局部特徵及上下文映射為候選耦合評分:
GPU 可以用批次方式計算 的部分或全部。之後由授權的 CPU PWN 對 進行相位資格重驗、版本比對、必要來源驗證與局部選擇。這不同於 GPU 直接使 的所有邊立即正式生效;當 GPU 只做候選評分時,CPU 與 GPU 必須保留該區分。
3.3 物理傳輸仍然存在
離散 GPU 常見資料路徑包括 host memory、PCIe 與 device VRAM。即使使用 UVA 或 managed memory,也不等於資料沒有移動、CPU cache 與 GPU L2 自動共享所有一致性語義,或任何 Windows 桌機具有 NVLink-C2C 等效能力 [E1–E3]。GPU 的吞吐增長可以減少資料處理所需時間,卻未必減少資料到達所需時間。
3.4 至少五個時間度量
令一個 GPU batch 的事件分別為 (准入)、 (提交)、 (kernel 完成)、 (可在 host 安全讀取)、 (局部候選可用)。定義:
這裡 可能包含 GPU queue waiting、kernel 啟動及 kernel 執行,不能把它直接命名為 pure_kernel_time。總局部候選延遲為:
這些時間以同一 host 單調時鐘標記端點;GPU event timing 可協助解開 device 部分,但 host clock 與 device clock 不可未校準便直接相減。正式提交另有 Paper 04 的決策和持久化延遲,不能藏入 candidate 指標。
4. GPU 運算資格與工作片段分類
4.1 能上 GPU 不代表應該上 GPU
提出至少八個與工作片段相關的特徵:可並行元素數 、每元素算術強度 、輸入資料 bytes 、輸出 bytes 、分支發散 、資料常駐率 、截止時間 、所需品質/決定性條件 。這些是 profile 輸入,不可只用 GPU 佔用率猜測。
4.2 工作片段分類
| 類別 | 候選 GPU 運算 | 常見風險 | V0 策略 |
|---|---|---|---|
| C1 大量關係評分 | 批次評分、多候選配對 | 小 batch 反而慢;資料重排 | 具門檻 admission |
| C2 向量/矩陣轉換 | embedding、相似度、特徵矩陣 | 若本就常駐 GPU,反向搬移浪費 | 優先常駐區處理 |
| C3 稀疏圖聚合 | segmented reduce、scatter/gather | 不規律存取、原子競爭 | 與 CPU 稀疏 baseline 對照 |
| C4 批次比較 | digest 前處理、向量一致性檢查 | 不能把比較當來源認證 | 只輸出候選比較紀錄 |
| C5 R3 編碼數值運算 | 明確碼制下 shard 變換 | 正確性及復原前提不完整 | 先 CPU 參考算法 |
| C6 AI 訓練數值核心 | 已宣告演算法下的批量張量計算 | GPU 已是瓶頸且無新平行度 | 與成熟框架比較 |
| C7 極短條件判斷 | 單筆 branch、權威 guard | launch/copy 主導 | 預設 CPU-only |
| C8 權威提交/外部效果 | 無 GPU 必要角色 | 版本、原子性、持久化責任 | 禁止 GPU 直接越權提交 |
4.3 可降階與可回退性
每個可交給 GPU 的 slice 至少具備 host_reference_operator 或等價的、可在 V0 執行的 oracle/參考算法。若設備不可用、品質模式不支援或成本估計缺乏可信範圍,執行層必須能產生 FallbackCPU、Deferred、Rejected 或 Unknown,不可把 GPU_REQUIRED 偷當成所有工作流的預設語義。
4.4 業務/研究品質先於運算品質
例如 AI 推論中的 gate scoring 可能在 GPU 上得到非常接近的分數,但若所用 floating precision 導致候選排名翻轉,實際選到不同 expert,整個工作流的語義便已改變。必須讓 quality_contract 定義允許的誤差、排名穩定性、NaN/Inf 處置及異常值策略;不得只在 GPU kernel 的單一局部輸出上確認近似正確。
5. HeterogeneousWorkSliceV0_1:跨硬體任務的共同語義
5.1 候選封套
HeterogeneousWorkSliceV0_1 {
slice_id: SliceId
flow_id: FlowId
flow_owner_generation: MonotonicGeneration
event_epoch: EventEpoch
phase_contract_ref: VersionedPhaseContract
coupling_admission_ref: Optional[AdmittedCouplingRef]
input_versions: VersionedReadSet
input_lineage_refs: List[LineageRef]
operation_id: TypedOperatorId
input_schema_id: SchemaId
output_schema_id: SchemaId
output_role: LocalCandidate | TentativeIntermediate
quality_contract: QualityContractId
allowed_backends: Set[CPU | CUDA | Mock | Other]
execution_deadline: Optional[MonotonicDeadline]
side_effect_class: Pure | IsolatedTentative
authority_scope: AuthorityScopeRef
retry_semantics: IdempotenceOrReplayPolicy
cost_budget_ref: VersionedBudgetRef
trace_correlation_id: TraceId
}
必要的 source digest、owner fencing 及 input lineage 可以透過此封套引用前篇狀態,不得由 adapter 為了方便另建一個不相容的身分系統。
5.2 不可接受的內容
side_effect_class 不能是 ArbitraryExternalIO;封套不能直接載入受保護的提交能力;不得只提供模糊字串例如 operator=verify 而不指定 verified algorithm/schema 及 source quality policy;不得讓 GPU 能根據結果擅自擴大傳播 credit 或開啟新權威 scope。
5.3 核心純度與狀態依賴
GPU kernel 最容易驗證的形式是純函數或對獨占暫態 buffer 的計算:
其中 、 的版本與精度政策已記錄。若 kernel 需要 atomics、非確定性縮減或可變全域 scratch,其效果須在算子簽章與品質契約中明示。is_pure=true 是有條件的工程聲明,不是函數名字包含 compute 就得到的證明。
5.4 Slice 拆分與可組合性
split(slice, partitions) 只有在輸出具可合法組合規則且每一分片保有原 slice_id 及子序號時成立。最基本的組合方式包括索引並列、已定序的串接、合法半群縮減及具特定精度政策的浮點 reduction。對非交換更新、需要全局排序的排名、跨分片一致性約束,不得無條件套用相同縮減器。
6. 分批、排隊與相位截止時間
6.1 為何 batching 是核心而不是小優化
GPU 的優勢通常需要足夠的並行工作量來攤薄 enqueue、launch、資料搬移與同步的固定成本。SMPDC 每個 PWN 持續產生工作,天然可以在相同 operator/dtype/品質契約下建立批次;但如果為達大 batch 而等候過久,會傷害短 deadline 的工作流。故本篇將 batch admission 視為策略問題,將其合法性與效率分開。
6.2 GpuBatchPlanV0_1
GpuBatchPlanV0_1 {
batch_id: BatchId
member_slice_ids: OrderedList[SliceId]
operator_family_id: OperatorFamilyId
compatible_schema_id: SchemaId
compatible_quality_contract: QualityContractId
independent_authority_scopes: List[AuthorityScopeRef]
membership_generation: MonotonicGeneration
earliest_deadline: Optional[MonotonicDeadline]
max_batch_wait_ns: NonNegativeDuration
max_batch_items: PositiveCount
max_input_bytes: NonNegativeBytes
max_output_bytes: NonNegativeBytes
max_device_mem_bytes: NonNegativeBytes
item_offsets: BoundedOffsetVector
partition_reconstruction_policy: VersionedPolicyRef
estimated_cost_interval: Optional[CalibratedCostInterval]
fallback_operator_ref: CpuReferenceOperator
}
若來源不同權限域的資料不能相互共置,雖然它們 dtype 一樣,也不得因 GPU batching 直接合併。independent_authority_scopes 是檢查用的集合,不是代表 GPU 因此擁有全部 domain 的權限。
6.3 兩級佇列
PWN 局部佇列記錄可交給異質後端的 slice,不放棄工作流的邏輯 owner。GPU admission 佇列只有在 OS/Runtime 守衛通過且保留 quota 後才產生可提交 batch。兩級佇列均需要上限、到期政策與回退;避免每個 PWN 同時直接對同一 GPU context 盲目提交細碎 kernel。
6.4 Batch 決策的三個門檻
- 大小門檻: 只是一個實測待校準候選,不得固定為通用常數。
- 時間門檻: 當等待導致 deadline 可能失敗,須提早發送或 CPU fallback;不能為了 GPU 滿載率犧牲明確任務約束。
- 成本門檻: 若目前校準的上界估計仍不優於 CPU,拒絕 offload;若資料不足,標示
UnknownEstimate而非假裝具有確定優勢。
注意:此處的 batch 成員是已持有准入資格的運算 slice,不是額外創造新的邏輯消息信用。
6.5 批次重建與輸出映射
每個輸出項必須映回 (batch_id, slice_id, slice_owner_generation, input_version_digest, output_offset, output_digest)。若 GPU 使用 prefix-sum、排序或重新排列提升吞吐,adapter 必須保留逆向映射以正確返回原 PWN;否則不同工作流的值可能被錯誤投遞,且此類錯誤在數值結果看似合理時極難偵測。
7. 成本模型與「GPU 選擇」的嚴格判定
7.1 端到端而不是 kernel-only 成本
令同等輸入、品質、deadline 與安全規則下,CPU 路徑的端到端候選完成耗時為 ;GPU 混合路徑為 。定義實際淨收益:
只有在 且品質與資源約束完全符合時,才能稱為此 workload 的端到端加速。若要對 throughput 作主張,應另定義 、 (每秒有效已驗證或正式提交工作)而不能把 偷換成吞吐提升。
7.2 非重疊的保守估計
當各階段串行且可精確分離時,可估:
如果使用多 stream、copy engine、CPU/GPU overlap,這不是一般情況的精確相加式。重疊的真實 wall-clock 應直接以同一 trace 的關鍵路徑長度或 host 端到端時鐘量測;不能把每個 stage 的累計時間直接相加,也不能重複扣除 overlap。模型可用來推導上界候選,不能替代實測 [E1, E4]。
7.3 基於搬移 bytes 的理論下界
若實際需要跨互連搬移 bytes,測得有效傳輸帶寬為 ,則單純傳輸耗時具有理想下界形式:
這裡 必須與方向、pinned/pageable、其他裝置流量及實際測量口徑一致,不能用 PCIe 名目規格當成已量到的端到端有效頻寬。若資料已常駐 GPU 或輸出只需小型摘要, 可能遠小於每次重新往返完整 tensor 的版本;這正是值得測試的實際設計空間。
7.4 算術強度與 roofline 直覺
設運算次數約為 ,裝置訪存 bytes,運算強度 。對 很低、GPU VRAM 訪問支配的 kernel,額外 GPU 算力不必然帶來線性改善;對 高且 GPU 能維持足夠 occupancy 的批次,才可能以更高吞吐攤薄 launch。這是判斷候選的工具,不是新的 roofline 定理。
7.5 排隊模型與突發流量
平均 arrival rate 接近實際 service rate 時,排隊延遲可能急升。故「平常 GPU 利用率只有 30%」不等於某一短 deadline 批次一定立即可執行;同樣,單一 GPU 的高 utilization 可能表示 queue 過深。控制器必須記錄 in-flight bytes、等待時間、VRAM 水位、batch age、可取消工作量與 PWN deadline miss。
7.6 預先載入與常駐性
當相同模型權重、索引、詞表或稀疏圖被重複使用,預先放在 VRAM 可能使每輪只需送入少量 delta。但須把第一次 warming、cache residency、失效刷新、記憶體租約及工作切換成本納入。若把資產事先載入的開銷藏到 benchmark 之前,便不符合公平比較。需要另列 cold、warm、steady-state 三類完整量測。
8. 裝置能力快照與後端資格
8.1 Capability 是 runtime observation,不是市場型號的推論
即便同一 GPU 型號,driver、CUDA Runtime、OS、顯示負載、VRAM 空間、device attribute、stream 優先度、memory pool 及計算模式都可能改變可用能力。後端必須查詢,例如 CUDA 設備名稱、compute capability、memory allocation capability、copy engine 數量、是否支援 concurrent kernel/managed access、裝置記憶體可用量,以及必要 API 是否真正可用。gpu_model 不足以取代 capability handshake [E1–E3]。
8.2 GpuDeviceCapabilitySnapshotV0_1
GpuDeviceCapabilitySnapshotV0_1 {
device_id: StableRuntimeDeviceId
driver_version: DriverVersion
runtime_version: RuntimeVersion
device_arch_id: ArchitectureReference
os_execution_mode: NativeWindows | NativeLinux | WSL | Other
total_vram_bytes: NonNegativeBytes
currently_available_vram_bytes: ObservedValueWithTimestamp
supported_transfer_modes: Set[Pageable | Pinned | Managed | Mapped]
supports_concurrent_copy_and_compute: ObservedCapability | Unknown
asynchronous_copy_engines: OptionalNonNegativeCount
supports_graph_capture: ObservedCapability | Unknown
supports_memory_pools: ObservedCapability | Unknown
managed_access_mode: Restricted | FullSoftware | FullHardware | Unknown
capability_generation: MonotonicGeneration
observed_at: MonotonicTimestamp
observation_validity_horizon: OptionalDuration
}
ObservedCapability 不能只依靜態 SKU 廣告建立;實際行為仍須在測試環境確認。對沒有觀測證據的 capability 使用 Unknown,而不是在 bool 欄位預設 true。
8.3 無效快照與能力降階
假設 admission 讀到 4 GB 可用記憶體,之後其他程序突然配置了 3 GB;原先快照不能用來保證 allocation 一定成功。因此准入中的 memory lease 必須處理競爭後失敗:ReservationProposed、Reserved、AllocationFailed 或 UnknownOutcome。任何 GPU adapter 必須允許在資源不足或 device reset 時降階,而不是把 cudaErrorMemoryAllocation 直接轉成工作流計算錯誤已提交。
8.4 共享桌面 GPU 的特殊性
RTX 3070 可能同時負責桌面顯示、其他視覺/AI 軟體和測試 kernel,Windows WDDM、顯示器負載及背景程序會干擾尾延遲。在 V1 測試時,應將顯示負載變化、GPU context 初始化、driver API 版本及記憶體使用量列為混雜因素。沒有必要假設或要求將日常使用的系統改成伺服器計算模式。
9. GpuExecutionTicketV0_1 與 device state machine
9.1 提交不是完成
GPU 操作通常是非同步 enqueue。成功提交表示 host API 已接受命令,不能推導 kernel 已執行、後續 H2D/D2H 已完成、host 端可讀或 PWN 已可提交。為此建立具有明確狀態的 ticket。
GpuExecutionTicketV0_1 {
ticket_id: TicketId
batch_id: BatchId
source_slice_ids: OrderedList[SliceId]
device_id: StableRuntimeDeviceId
capability_generation: MonotonicGeneration
flow_owner_generation_refs: Map[FlowId, OwnerGeneration]
event_epoch_refs: Set[EventEpoch]
operation_graph_digest: Digest
input_lease_refs: Set[BufferLeaseId]
output_lease_refs: Set[BufferLeaseId]
stream_sequence_ref: QueueSequenceRef
device_event_ref: Optional[DeviceEventRef]
submission_generation: MonotonicGeneration
submitted_at_host: Optional[MonotonicTimestamp]
declared_completion_condition: VersionedConditionId
recovery_key: StableRecoveryKey
trace_correlation_id: TraceId
}
其輸出語義不是 Result<T>,而是:
SubmitResult =
Submitted(ticket)
| Rejected(reason)
| Failed(error)
| Unknown(recovery_key)
若 host thread 在底層 API 回應前崩潰,結果必須可表達 Unknown;若重送的 work 具有外部可見副作用,不能在未完成恢復查詢之前盲目二次執行。
9.2 裝置事件生命週期
Draft
-> AdmissionAccepted
-> BuffersReady
-> Queued
-> Submitted
-> DeviceExecuting
-> KernelDone
-> TransferDone
-> HostReadable
-> CandidateImported
-> LocalVerified
任何活躍狀態可進入:
Rejected | Failed | TimeoutUnknown | AbandonedPendingCleanup
LocalVerified 仍須另行進入 Paper 04 的 CommitProposal,
並經 CommitValidate/Decide/Publish 才可能成為 Committed。
這是一個邏輯狀態機,不是說 CUDA API 原生就有相同的十種狀態。某些路徑可能沒有必要的 D2H(例如下一步仍全部在同一 GPU 且 output_role 是 DeviceIntermediate),但只要 CPU 要讀結果,便必須走可見性分支。
9.3 三種完成門檻
KernelDone:經正確 stream/event dependency 觀測到某 kernel 的必要 device command 完成,不包含尚未安排的 copy。TransferDone:對指定 input/output buffer 的所需傳輸與 device 相依操作已在正確事件鏈上完成。HostReadable:指定 host address 對指定版本、作用域、快照與同步機制可被 CPU 安全讀取,且在該讀取期間沒有相衝突的持有者。
把 cudaEventQuery(event)==success 誤當成所有上述操作完成,可能漏掉事件記錄位置之後提交的 D2H copy。因此 HostReadable 必須綁定具體的最後寫入、copy 和 owning lease,而不是只綁定一個任意 event handle [E1, E4]。
9.4 裝置成功與業務正確性不同
cudaSuccess、GPU event complete、資料 digest 一致、R2 witness accepted、LocalVerified、Committed 是不同判準。GPU 只應附上硬體/執行證據,不能替代 R2 的來源獨立性、Paper 04 的跨鍵一致性或正式權威決策。
9.5 查詢契約與不確定狀態
query_event(ticket_id) 必須區分 Pending、CompletedDevice、TransferCompleted、Failed、Unknown。當 adapter 喪失裝置 context、ticket ID 尚未被持久化或 event handle 無法查詢時,回傳 Unknown 而非 Failed;否則上層可能錯誤地認為原工作完全未執行。若算子是純函數且輸出暫態,重新執行可以安全地建立新候選,但必須重新計算資源與血緣、維持原正式提交鍵的防重契約。
10. Host/Device 記憶體模型與可見性
10.1 位址相同不代表存取無成本
Unified Virtual Addressing 可讓部分記憶體使用一致的虛擬位址表示;Managed Memory 可隱藏部分搬移程式碼;但位址方便性、物理常駐位置、cache coherence、合法並行讀寫、資料搬移成本仍是不同概念。Windows 環境的 managed memory 支援與 CPU/GPU 同時存取有具體限制,須依 runtime 和裝置能力查詢,不能直接將 Linux HMM 或 Grace Hopper 的行為投射到一般桌機 [E2–E3]。
10.2 四種主要主機與裝置配置方式
| 方式 | 主要優點 | 必須計入/檢查的風險 |
|---|---|---|
| Pageable host + explicit copy | 容易實作 | cudaMemcpyAsync 可能涉及 staging/同步,不能假設真正非同步 |
| Pinned host + explicit async copy | 常有利於 H2D/D2H 與 overlap | 鎖定頁面、記憶體壓力、配置/回收成本 |
| Managed / Unified Memory | 簡化編程與地址管理 | residency、fault、migration、OS 及 coherence 限制 |
| Device-resident buffer | 避免每輪重複 H2D/D2H | VRAM 配額、失效刷新、與其他工作流的隔離 |
特定硬體亦可能支援 mapped host/zero-copy 等功能,但名字叫 zero-copy 不代表零延遲、零互連流量或總是更快。需要實測其有效頻寬與重複遠端訪存成本 [E1–E3]。
10.3 DeviceBufferLeaseV0_1
DeviceBufferLeaseV0_1 {
lease_id: BufferLeaseId
allocation_id: AllocationId
memory_domain: HostPageable | HostPinned | Device | Managed | Mapped
device_id_optional: Optional[StableRuntimeDeviceId]
base_allocation_generation: MonotonicGeneration
address_range_ref: OpaqueMemoryRegionRef
byte_size: NonNegativeBytes
data_schema_id: SchemaId
logical_owner_flow_id: FlowId
owner_generation: MonotonicGeneration
usage_rights: Read | Write | ReadWrite
active_ticket_refs: Set[TicketId]
active_stream_use_refs: Set[StreamUseId]
last_required_event_ref: Optional[DeviceEventRef]
outstanding_host_readers: NonNegativeCount
validity_state: Reserved | Active | PendingRetire | Retired | Unknown
memory_visibility_policy: VersionedPolicyRef
audit_trace_ref: TraceId
}
這是 Runtime 的語義描述,不是要求每次底層 CUDA malloc 都攜帶上述全部欄位。其用途在於證明 buffer 不能在最後使用者完成之前被釋放或重用。若使用 cudaMallocAsync/cudaFreeAsync,更須遵守 stream ordered allocation/free 的順序;跨 streams 使用應補充 event 依賴,而不是想像 allocator 會自動替應用建立任意正確因果邊 [E6]。
10.4 Lease 拆分與 DMA 期間禁止釋放
一個 slice 可在預先確認不重疊的 address ranges 上建立多個不可互相覆寫的借用;但若分區有別名、scatter 存在共享索引或不同 kernel 同時寫入相同位置,應提高驗證責任。release_lease 對仍有 in-flight DMA 或 kernel 使用的 buffer 應返回 PendingRetire,而不能按 CPU thread 已返回便立即 free。
10.5 映射與 host 存取
若 output 是 GPU VRAM 中的向量,CPU PWN 想驗證它可選擇:
- 執行必要 D2H copy,等待對應 copy 完成並取得
HostReadable; - 將後續純數值 kernel 留在 GPU,僅將可驗證摘要與血緣送回 host;
- 在實際支援的 coherent shared memory/mapped memory 模式使用依法保證的 host visibility 與同步規則。
第三種不是可以無條件使用的捷徑。當裝置不支援所宣稱的共享並行存取時,adapter 應拒絕 MappedConcurrentAccess,避免違反平台特定 CUDA memory model [E2–E3]。
10.6 Host visibility 的形式責任
對 buffer 與某最後必要寫入/copy 事件 ,只有在存在合法的 happens-before 鏈、buffer lease 仍有效、寫入版本匹配且沒有與 host read 衝突的持有者時,可簽發:
這張憑證是 Runtime 的保證義務,不能由使用者 cast 出來;其是否有效取決於底層 driver/OS 的記憶體模型與實際同步,而不是形式符號本身。
11. CUDA Streams、Events 與可觀察因果偏序
11.1 Stream order 是局部順序,不是全域同步
在相同 CUDA stream 中,已提交的 kernel、copy、event 具有相應序列依賴;不同 streams 的操作可以並行,也可能因資源及隱式同步而序列化。不能把「使用兩條 stream」當成必然可重疊的效能證據 [E1, E4]。
11.2 三種依賴關係
- Local stream order: 在同一 stream 上,先前操作在相應順序規則中排於後續操作之前。
- Cross-stream event dependency: stream B 透過
cudaStreamWaitEvent(或合法等價方式)依賴在 stream A 錄製的 event,建立指定指令間順序。 - Host-visible completion: host 端使用相應 event 或 stream completion query/wait,加上 memory visibility 合法性,才能在 CPU 上讀值。
這些不是同一種同步。只在 CPU 控制圖上畫箭頭,不會自動產生 CUDA stream 關聯。
11.3 依賴偏序
定義 表示執行事件 必須在 前完成。以一次 H2D、kernel、D2H 為例:
當操作分在不同 streams,需以 events 或 graph dependencies 跨 streams 明確建立此偏序。若程式選擇將前三者放入相同 stream,順序可以較簡單,但未必達成 copy/compute overlap。若想重疊,通常需要多批次 pipeline 而不是違反同一批次的真實依賴。
11.4 為何 default stream 需警惕
某些 default stream 模式有隱式同步規則,另有 per-thread default stream 選項。不同編譯選項和程式庫的 default stream 行為可能影響實際並行性;研究應記錄所用模式,優先在同等依賴語義下建立明確 streams/events,不能把 stream 0 當成所有版本都有同一種非同步效果 [E1]。
11.5 CUDA Graphs 不可缺席於比較組
CUDA Graphs 已能以顯式圖或 stream capture 定義可重用 GPU 工作與依賴,降低反覆 kernel launch 的 CPU 開銷;部分動態圖也能透過 update 或 graph selection 支援,不能宣稱既有 GPU 系統一概只會重複逐 kernel launch [E5]。因此 SMPDC 的候選批次實作至少需要比較:
- 一般 CUDA streams with proper async enqueue;
- 相同 kernel 的合法融合/批次版本;
- 可重用時的 CUDA Graphs;
- SMPDC 增加跨 PWN 分組、相位守衛與結果路由的全成本版本。
如果第 4 項輸給第 3 項,應誠實將結果列為負面實驗;不應以不優化的 naive CUDA 當唯一對手。
11.6 高頻提交下的圖復用門檻
設 graph 建置/實例化成本為 ,每輪相較逐 kernel 方式節省 , 預估復用次數為 。僅在最簡化、其他開銷不變且 的假設下,有候選打平條件:
若圖須頻繁因相位、batch shape、buffer address 或 dtype 改變而重建,需把 update 與 reinstantiate 成本納入。實際 gain 不能直接由公式推論,因 GPU 排程、融合與 cache 也會改變 [E5]。
12. DeviceVisibilityWitnessV0_1 與事件見證
12.1 提交、完成及可見性 witness 的型別分離
DeviceVisibilityWitnessV0_1 {
witness_id: WitnessId
ticket_id: TicketId
device_id: StableRuntimeDeviceId
event_ref: DeviceEventRef
completed_operation_set_digest: Digest
completion_state: KernelDone | TransferDone | HostReadable
memory_region_ref: OpaqueMemoryRegionRef
buffer_lease_id: BufferLeaseId
buffer_generation: MonotonicGeneration
output_schema_id: SchemaId
source_version_digest: Digest
synchronization_rule_id: VersionedRuleId
signed_by: TrustedAdapterIdentity
observed_at_host: MonotonicTimestamp
trace_correlation_id: TraceId
}
completed_operation_set_digest 必須對應實際已成功完成的必要 command chain,不能只列出期望流程。signed_by 在研究 v0.1 代表具可信身份的 adapter 出具及可查驗的審計紀錄,不預設現有 GPU 原生具備密碼學硬體簽章。
12.2 Witness 條件
若只是 cudaEventQuery(event_A)==success,而事件 A 發生在 D2H 之前,則對 CPU 可讀的 HostReadable 見證不成立。反之,如果 host buffer 的 D2H event 已成功等待,但 source input snapshot 已過期,則 host visibility 可能成立,正式提交正確性仍未成立。此區分防止裝置時間線與權威時間線被混合成一張模糊的 complete=true 標籤。
12.3 可觀測完成證明與內容真實性不同
Device completion witness 回答「哪些已准入、具體 event-chain 的操作已完成、資料何時可安全讀取」;內容正確與否仍需品質 policy、CPU/可靠參考驗證器、輸入版本及操作定義確認。即使使用 ECC VRAM,亦不等於演算法正確、來源可靠或外部效果冪等。
12.4 與 GPUEvent 型別的關係
Paper 05–06 的 GPUEvent<Device,Work> 是較低層次的事件 handle;DeviceVisibilityWitnessV0_1 是驗證了事件依賴、buffer lease 與版本後的較強證書。兩者不能隱式互換。編譯器可以插入等待/copy/guard,卻不能在無 runtime 證據時編造 witness。
13. GpuCandidateOutcomeV0_1 與 CPU 主節點回收
13.1 候選輸出契約
GpuCandidateOutcomeV0_1 {
ticket_id: TicketId
slice_id: SliceId
originating_flow_id: FlowId
flow_owner_generation_at_admission: MonotonicGeneration
flow_current_owner_generation_checked: MonotonicGeneration
input_version_digest: Digest
input_lineage_refs: List[LineageRef]
operation_graph_digest: Digest
quality_contract_id: QualityContractId
output_schema_id: SchemaId
output_location: DeviceIntermediate | HostReadableBuffer
device_visibility_witness_ref: Optional[WitnessRef]
host_readable_buffer_lease_ref: Optional[BufferLeaseId]
result_digest: Optional[Digest]
candidate_state: Produced | Importable | Conflict | Failed | Unknown
trace_correlation_id: TraceId
}
Produced 可表示裝置輸出已存在,Importable 表示已具備相關 host-readable 條件、可以交給 PWN 的局部候選處理。Importable 不是 LocalVerified;後者仍需依 Paper 03 的品質、來源及比較政策。
13.2 主節點合併不是移到 GPU 執行
本系列原始構想要求主要節點負責合併、判斷、去重。這可以解讀為 PWN 保有局部決策責任:GPU 可做大量可平行的前處理、局部歸約或比較,最後的 RECONCILE_SEAL 由合法的 PWN/可信 Runtime 執行,必要時結合 CPU 驗證器。即使部分比較 kernel 在 GPU 上完成,也不改變 logical owner 對局部結果的責任。
13.3 Output import 時的 owner generation
若 slice 由 PWN-A 的 owner generation 7 送出,GPU 稍後完成時,PWN-A 已由 generation 8 的新 worker 接手。adapter 應檢查原 lease/fence、尋找合法的 successor receiver,移轉 ticket ownership 或隔離過期結果。不得把已完成 output 直接交給舊 worker 並允許它繞過 fencing 修改 PWN 本地 canonical 狀態 [I1, I7]。
13.4 多個 GPU results 如何形成一個局部結果
多 GPU batch 或同一 GPU 的數個 streams 所返回的候選,必須依各自 slice_id、版本、血緣、reduction order 與品質約束重組。若某片失敗、某片通過、某片 timeout,結果不一定可以計算成完整 aggregated outcome;應以型別化 Partial/NeedsRetry/Hold 表達,不得偷偷以零值、上一輪值或重複項填補。
13.5 傳回 CPU 的資訊量與驗證需求
把所有中間 tensor 完整 D2H 可以簡化驗證,但可能抹去 GPU 的效能收益。將檢查規則中 GPU 可執行的純數值部分下沉,讓 host 只接收候選摘要、錯誤證據與可按需回讀的資料,是值得研究的方向。然而若摘要不足以驗證所宣稱的結論,就必須回讀更多資料或降低確定性等級;不能以「省資料搬移」當成跳過必要證據的理由。
14. 異質執行的最小狀態機與 adapter API
14.1 延伸 Paper 07,不替換它
保留 Paper 07 既有 GPUAdapterControlV0_1,本篇補充如下資料平面接口:
SmpdcGpuDataAdapterV0_1 {
describe_device() -> GpuDeviceCapabilitySnapshotV0_1 | Unknown
admit_slice(slice, budget, authority_guard)
-> AdmittedSlice | Rejected | Deferred | Unknown
plan_batch(admitted_slices, policy)
-> GpuBatchPlanV0_1 | Incompatible | Unknown
acquire_buffers(batch, memory_policy)
-> LeaseSet | Pending | Failed | Unknown
enqueue(batch, lease_set, execution_policy)
-> Submitted(GpuExecutionTicketV0_1) | Rejected | Failed | Unknown
query(ticket)
-> Pending | KernelDone | TransferDone | Failed | Unknown
import_output(ticket, output_ref, required_visibility)
-> HostReadableCandidate | Pending | Failed | Unknown
retire(ticket_and_leases)
-> Retired | PendingRetire | Failed | Unknown
trace(ticket)
-> EventTrace | IncompleteTrace | Unavailable
}
plan_batch 可以是純政策,但 admit_slice、acquire_buffers 與 enqueue 均有資源效果,必須受 budget、owner、authority 與設備 generation guard 約束。
14.2 參考執行序列
PWN A,B,C: phase-ready slices
-> `FlowRegistry` + ownership/version guards
-> `GpuBatchPlan` candidate
-> batch admission / budget reservation
-> buffer leases and H2D staging
-> explicit CUDA stream/graph dependency
-> device kernel(s)
-> event/query/required D2H
-> HostReadable witness
-> per-slice candidate import
-> PWN local reconciliation and evidence seal
-> Paper 04 commit proposal/authority decision
-> lease retirement and trace closure
錯誤時須指出停在哪個階段,保留責任人與未結清 lease/ticket。這是避免同一輪工作發生 buffer leak、in-flight data race 或無法回收的必要條件。
14.3 最小可執行策略(概念偽程式碼)
for each ready PWN:
slice = compile_or_construct_typed_slice(PWN)
if not phase_and_owner_guards_hold(slice): reject(slice)
else if cpu_preferred_by_policy(slice): run_cpu_reference(slice)
else enqueue_for_batch_admission(slice)
for each admissible GPU batch:
if not stable_budget_or_quality: fallback_or_defer()
leases = reserve_and_allocate_buffers(batch)
if leases unavailable: fallback_or_defer()
ticket = submit_with_explicit_dependency_graph(batch, leases)
if ticket is Unknown: query_or_reconcile_before_retry()
else register_outstanding_obligations(ticket, leases)
for each completed_ticket:
visibility = verify_required_copy_and_visibility(ticket)
if not visibility: keep_pending_or_fail_closed()
else:
for each slice_output:
if owner_generation_still_valid_or_handoff_complete:
import_candidate_to_pwn(slice_output)
else quarantine_stale_result(slice_output)
retire_leases_only_after_all_uses_and_readers_finish()
偽程式碼沒有包含全部實體 API 與例外處理,不可直接當作生產程式;它指定了哪些檢查不能因最佳化而省略。
14.4 三種部署方式
A. CPU 各 PWN 直接提交 GPU: 容易實作,但較難集中批次、quota 和跨 worker 問題。B. 單一 GPU broker: 有利集中 batching/memory lease 管理,但 broker 可能成為瓶頸或單點故障。C. 多個受限 brokers 共用 policy/ledger: 可分擔控制負荷,但必須協調 device context、resource ledger、ticket ownership 與公平性。這三者應當是可消融的實驗候選,不預設某個全域最佳。
15. GPU 對 R0–R3 冗餘模型的角色
15.1 R0:單路高吞吐算子
R0 是基線,也是很多 AI batch 的實際路徑。若 CPU PWN 只需要將一組候選進行高維向量評分,GPU 可以只執行一個批次核、回傳一次結果。為證明 R1 的必要性,必須先用成熟 R0 實作得到吞吐和 tail latency;不能把冗餘本身當成免費的計算加速。
15.2 R1:複製/競速與 GPU 資源競爭
對同一 operator 可安排 CPU 與 GPU 競速或兩個合法執行路徑,但 first-valid 只可能在某些 tail latency 條件下降低等待;被放棄的 GPU kernel 仍可能消耗 compute、copy 與 VRAM,且一般 kernel 並不保證能即時撤銷。R1 最終結果只有一份同源候選,不得提升為 R2 獨立佐證 [I3]。
15.3 R2:異源證據可包含 GPU,但設備不同不保證方法獨立
即使一份候選在 CPU 計算,另一份在 GPU 計算,兩者仍可能共享演算法、錯誤資料、權重或錯誤規格。因此 hardware_backend_is_different 不是 IndependentWitness 的充分條件。可實驗不同演算法、不同資料轉換或不同數值精度的交叉檢查,但其可認證獨立性仍需由 Paper 03 的 provenance policy 判定。
15.4 R3:編碼與解碼可以 parallelize,解碼條件不能偷改
明確的 R3 擦除碼能讓 GPU 處理足夠大的 shard matrix、GF 運算或合適的大量資料轉換;但當 shard 不足、索引衝突或來源與完整性不合時,不能輸出 Recovered。若資料很小,CPU 或現成 library 可能更有效;因此需比較不同 shard size、 組合以及 encode/decode 的實際 copy cost。
15.5 冗餘與 GPU 批次對 VRAM 的放大
若 R1 建立 份副本,記憶體需求不一定恰為 倍(可共享唯讀資料),但同時 in-flight output、scratch、descriptor 與候選緩衝可能增加。GpuBatchPlan 必須對 peak VRAM、pinned host bytes、outstanding events 設硬上限;一個副本的取消不代表其尚未完結的 device lease 可立即釋放。
16. 推測性 GPU 工作、取消與權威一致性
16.1 GPU 工作可以先行,但仍屬暫態
如果 Paper 04 的 SPEC_BEGIN 建立分支 ,GPU 可處理其具版本的純運算子,並以 SPEC_ATTACH_GPU 將 ticket 與 buffer 依賴記入分支。這允許 CPU 主工作流不必同步等待 kernel 完成便安排其他合法工作。然而這個「提前」只能提早計算,不能讓推測性結果在正式權威狀態中提前生效 [I4–I6]。
16.2 CPU branch 失效時不能假裝 GPU kernel 已消失
當上游版本衝突導致分支失效,可立即阻止後續 candidate import 和 commit proposal,但 GPU 指令可能已經開始或排在 device queue 中。對已提交 GPU work,取消有三種不同意義:
- Logical cancel: 禁止將輸出作為可提交候選;可即時實作。
- No further enqueue: 停止後續批次或圖節點提交;需依後端能力實作。
- Physical abort: 中止正在執行的 kernel;一般不得假定可無代價且立即完成。
沒有能力可靠地執行第三項時,合法策略是保持 buffer leases 至 device 完成/恢復邊界後再回收,並把 wasted GPU work 納入成本。
16.3 Copy/Kernel 使用者的失效閉包
假設 產生 device buffer ,而 、 依其啟動後續 kernel。若 因輸入快照失效,必須追蹤那些輸出不可被合法重新使用。若另有經證明獨立於失效 input 的純計算,才可從失效閉包排除。不能以「GPU 很快,重跑就好」取代依賴管理,因重跑仍要消耗 launch、copy、計算及 VRAM。
16.4 提交世代與輸入版本守衛
GpuCandidateOutcome 需綁定 snapshot_id 或等價的 input_version_digest。在 Paper 04 COMMIT_VALIDATE 之前,必須確認該結果的 read set/predicate set 仍符合所要求的隔離契約。模型 step 中的 optimizer state、random seed、parameter version,不能因 kernel 順利完成就視為仍與當前模型狀態一致。
16.5 重要反例:CPU 更新後的 GPU 遲到
Step 01 PWN-A snapshots model_version = 90
Step 02 GPU ticket T90 submitted for scoring
Step 03 PWN-B commits model_version = 91
Step 04 GPU T90 kernel finishes and D2H completes
Step 05 T90 output becomes HostReadable (physical fact)
Step 06 PWN-A tries to use T90 as if model_version = 91
Expected: REJECT / RECOMPUTE / TENTATIVE-ONLY
Forbidden: convert T90 output to Committed(model_version=91)
此反例中,裝置可見性與演算法版本正確性分屬不同問題。它應成為有限狀態模擬器的固定測試案例。
17. 數值一致性、reduction 決定性與品質門檻
17.1 GPU 的浮點數縮減順序可能改變結果
浮點加法在有限精度下通常不具嚴格結合律。多個 PWN 的輸入若分批改變求和樹、warp/block 分配或裝置實作,即使都「正確執行」,也可能出現小的 bitwise 差異。對高敏感 AI routing、閾值分類或形式驗證用摘要,這些差異可能改變控制流程。Paper 03 已要求區分逐位重現、數值近似和語義品質 [I3]。
17.2 允許三種品質模式,但不得混用
| 模式 | 要求 | 典型適用情況 |
|---|---|---|
BitExact |
在宣告硬體/版本域內逐位一致 | hash 相關位元運算、某些固定點運算 |
NumericTolerance |
絕對/相對/ULP 等明確容忍條件 | 特徵轉換、數值評分 |
TaskSemantic |
下游路由/分類/模型效果符合指定規範 | AI 專家選擇、近似推論 |
NumericTolerance 不自動保證 TaskSemantic。若候選排名差距很小,即使向量元素只差 也可能改變 top-1 expert,造成後續執行圖不同;需要獨立評估 downstream decision stability。
17.3 NaN、Inf、overflow、underflow 及未初始化資料
每個 GPU operator 應宣告非法數值政策:Reject、PropagateWithTag、ClampWithJustification 或 ExplicitUndefined;不可在比較時默默將 NaN 設為零或被視為與所有值相等。資料出界/未初始化使用須屬 implementation failure;不能以寬容的浮點誤差界掩飾 memory safety 缺陷。
17.4 隨機與非決定性的血緣
對隨機取樣、dropout、競速結果等,需要記錄 seed、RNG counter/stream identity、模型/算子版本與是否允許非決定性 kernel。相同 seed 在不同 GPU kernel 版本或不同執行拓撲下不一定得到同樣位元結果。若研究要求 bitwise reproducibility,必須事前限制硬體、驅動、演算法及 deterministic 模式,而不是事後挑選接近輸出當作符合。
17.5 不能把驗證器偷換成 GPU 輸出本身
若 GPU 產生輸出,然後在相同 kernel 上用相同運算再算一次,通常只能測量重複性,而不能提供獨立的數學正確性證據。對需要獨立驗證的 R2,應適用不同且已被明確驗證的算法/程式或可檢查證書。當完整 CPU oracle 成本太高時,可明確降低保證等級,但不能仍標記為 ProofVerified。
18. 八項條件性命題及證明概要
以下命題只在完整前提下討論,沒有宣稱形式系統機械證明或硬體實測成立。每一命題均須在 Paper 11 的 V0 模型檢查或 V1 實測中接受反例搜索。
18.1 命題 A:受保護 buffer 不會在最後有效使用之前被回收
前提: 所有讀寫 kernel、DMA、host reader 都被登記為 lease 使用者;最終事件關係正確且可觀察;回收器只在無 outstanding user、owner generation 合法且所需事件完成時標記 Retired;Runtime 不存在繞過 lease 的非受控指標。
命題: 對符合前提的同一 allocation generation,在所有受控存取完成前,受控 Runtime 不會合法重用或釋放該 allocation。
論證概要: Retire transition 的 guard 明確要求 outstanding user set 為空;任意仍存活使用者保留阻止回收的 witness,故可用狀態不變式作有限步歸納。若 kernel 可持有未登錄 alias、driver 出現不可觀測 DMA、或外部程式繞過 Runtime,前提失效,不能套用命題。
18.2 命題 B:合法 host import 必有裝置依賴先行證據
前提: 所有必要 device writes、copies 與 event wait 都正確進入追蹤依賴圖;HostReadable 只能由受信任 adapter 在滿足同步及平台 memory model 後簽發;CPU import 需要該 witness。
命題: 在受控 trace 中,CandidateImported 不會先於其所需的 HostReadable 成立。
論證概要: import transition 對 witness 存在硬性 guard;編譯期/Runtime 限制使繞過 guard 的 transition 不可到達。這是協議安全,不是「CUDA 會自動替錯誤程式修好同步」的宣稱。
18.3 命題 C:GPU 不能在缺少權威能力時創造 Committed
前提: GPU operator output type 只有 DeviceIntermediate 或 LocalCandidate;commit transition 必須由 Paper 04 的權威域和合法 guard 完成;不存在不受約束的 FFI/extern cast。
命題: GPU kernel 完成事件無法透過合法型別規則直接轉為 Committed<T,K>。
論證概要: GpuCandidateOutcome 與 CommitDecision 屬不同構造子,沒有純轉型規則;唯一升格路徑需權威仲裁。若直接暴露 raw pointer 至 canonical database write API,則此保證不能成立。
18.4 命題 D:批次成形不增加邏輯擴散 credit
前提: 所有成員片段維持 Paper 02 的 credit 身分;batch 的 split/merge 操作只改變承載表示,未創造新的 send grant;broker 支援原子入場帳本。
命題: 同一 event epoch 的合併/分拆不會因 batch 本身導致可合法傳送的 logical credit 總量增加。
論證概要: 對每個 batch membership 更新可用前後 credit 帳本等價證明;非法 credit.clone() 不在指令集合法操作中。這不限制為了實際冗餘而另行、合法申請更多 credit;該申請須由原協議審批。
18.5 命題 E:過期 owner 的 GPU 回傳不能直接修改 canonical PWN state
前提: PWN 有 monotonic owner generation/fence;candidate import 與 local state mutation 檢查當前 generation;舊 worker 沒有旁路權限。
命題: GPU 遲到結果若只持有舊 owner generation,無法在未完成合法 handoff/重新驗證時直接修改新 owner 的受控狀態。
論證概要: 新世代 fence 使舊 generation 的權限不被接受。物理 GPU 計算可能仍完成,但其候選值須隔離或交接。此命題不保證在不可信共享記憶體上無法被惡意程式直接寫入,需要隔離/OS 內存權限的外部假設。
18.6 命題 F:GPU 准入不會無條件破壞硬資源上限
前提: GpuBatchPlan 對可觀測的 buffer/ticket/quota 需求估算完整;resource ledger 與實際 allocation 雙重 guard;所有配置都經受控 admission,且無未受控外部佔用者。
命題: 系統不會合法核准超過其自身保留的硬上限的 in-flight ticket/buffer lease。
論證概要: 由原子 ledger reservation 對 admit transition 做歸納即可;但對實際全系統顯存 OOM 不提供絕對保證,因顯示程序、driver 保留及其他 process 可在檢查與分配間搶占資源。OOM 必須視為可處理狀態,而非理論失敗不可能事件。
18.7 命題 G:GPU 速度優勢必須在同等正確性下才有意義
前提: 基線和實驗組使用相同輸入、品質/deadline/isolaton/durability 政策,且完整測量 end-to-end wall time、失敗與資源成本。
命題: 若 且品質/安全/資源條件均滿足,則該次定義的工作負載上 GPU 混合路徑有正的端到端延遲收益;此結論不推及未測量的工作負載。
論證概要: 由 的定義直接得到,屬受條件限制的測量判準,而非架構必然加速的定理。
18.8 命題 H:可局部失敗不等於整個工作流必須死亡
前提: GPU slice 是純/隔離暫態運算,CPU fallback 擁有相同 operator semantics,原 slice 的重送、版本、credit、budget 與提交鍵被保留或依法補發;故障沒有破壞基礎權威狀態。
命題: 對受控 GPU 故障,可將剩餘候選計算降階至 CPU 而不必因此推導 canonical state 已被錯誤提交。
論證概要: 因 GPU 路徑只輸出暫態候選,未越過提交邊界;在新的合法 worker generation 上重算並重新驗證即可。此命題不保證能滿足原 deadline,也不保證重算零成本。
19. 三十二條最低安全不變量(G08)
以下規則是規範性測試標的。V0 模擬器可先覆蓋可模型化子集;V1 針對真正裝置能力再延伸。
19.1 主節點與資訊信用(G08-01~G08-06)
- G08-01:
flow_id不以 CUDA stream/kernel ID 取代。 - G08-02: GPU 後端不能自行擴大 PWN
authority_scope。 - G08-03: 工作流 owner generation 在輸出 import 時必須重新驗證。
- G08-04: Batch merge/split 保留原 credit lineage,不會憑空產生新的 logical send。
- G08-05: 不符合
phase_contract的 slice 不得被當作合法輸入。 - G08-06: 跨 domain batching 受到能力與隔離政策限制,而不只是 schema 相容。
19.2 Buffer、位址及生命週期(G08-07~G08-13)
- G08-07: 同一 buffer 的所有受控 device 使用均附上有效 lease。
- G08-08: Lease 綁定 allocation generation,避免 free/realloc 後的 ABA 混淆。
- G08-09: 仍有 in-flight kernel/DMA 不得完成
Retired。 - G08-10: CPU 讀取 host buffer 前必須取得滿足該平台 memory model 的可見性 witness。
- G08-11: 不能以
cudaMallocManaged的語法可用性假設任意 CPU/GPU 同時存取合法。 - G08-12: 跨 streams 同時讀寫相同區域須具有正確依賴/合法 race policy。
- G08-13: HostPinned bytes 與 VRAM allocation 的資源成本需納入帳本。
19.3 Ticket、event 與完成語義(G08-14~G08-20)
- G08-14:
Submitted不等於KernelDone。 - G08-15:
KernelDone不等於TransferDone。 - G08-16:
TransferDone不等於LocalVerified。 - G08-17: 需要 CPU 讀取時,不得缺少相應依賴的
HostReadable。 - G08-18: 事件 witness 綁定具體 ticket、memory region、buffer generation 與 source version。
- G08-19:
Unknown不能被默認處理為FailedBeforeExecution。 - G08-20: CUDA graph update/replay 不得繞過更新後的版本、權限及 buffer 使用契約。
19.4 候選品質與權威(G08-21~G08-27)
- G08-21: GPU 比較一致不會自動構成 R2 的獨立證據。
- G08-22: GPU kernel 的數值容差模式須與下游 quality contract 一致。
- G08-23: 過期 snapshot 的 output 不得直接成為正式狀態。
- G08-24:
GpuCandidateOutcome不能直接 cast 成Committed。 - G08-25: 推測失效的分支不得合法引入不可撤銷外部效果。
- G08-26: GPU 的部分 batch 失敗不會被默默補零後視為完整成功。
- G08-27: 對持有不同權威作用域的輸出,重組與發布必須遵守各自的隔離政策。
19.5 資源、測量與回退(G08-28~G08-32)
- G08-28: 不准使用未量測的 GPU 帶寬/launch 常數宣稱通用性能。
- G08-29: 性能報告保留 GPU queue、transfer、sync、CPU/GPU 計算及驗證的完整成本。
- G08-30: 需要時可退回滿足相同正確性契約的 CPU 執行,不使回退放寬品質政策。
- G08-31: CPU/GPU 複製競速需把失敗/被取消工作消耗的資源計入。
- G08-32: CUDA 版本或裝置 feature 缺失必須明確記錄,不能假裝測試了未支援的最佳化。
20. 故障與反例矩陣(F08-01~F08-36)
在本篇中,安全 failure 是能重現不變量違反或越權的事件;效能 failure 是在相同 correctness contract 下未取得淨收益。兩者不能以同一個「不夠快」原因統一解釋。
| 案例 | 錯誤情境 | 必須產生的處理/反例結論 |
|---|---|---|
| F08-01 | kernel 已 enqueue,程式立即讀 output | 等待必要同步/拒絕 host import |
| F08-02 | event 記在 D2H 之前卻宣稱 host visible | 拒發 HostReadable witness |
| F08-03 | 在另一 stream 使用 buffer 無 event order | 檢查器拒絕或測試檢出 race |
| F08-04 | DMA 未完即釋放 pinned buffer | lease 留待回收/檢出非法釋放 |
| F08-05 | cudaFreeAsync 早於另一 stream 的最後使用 |
拒絕缺失的跨 stream 依賴 |
| F08-06 | buffer 被重新配置,舊 ticket 使用相同位址 | allocation_generation fencing |
| F08-07 | Windows managed buffer 與 active kernel 並行 CPU 存取 | 依平台能力拒絕非法模式 |
| F08-08 | 系統型號標稱支援能力,但 runtime query 不支持 | feature negotiation 降階 |
| F08-09 | Pageable memcpy 被假定為完全 asynchronous | 改記錄真實 staging/同步 |
| F08-10 | H2D 拷貝尚未完成,GPU kernel 先讀 | 拒絕錯誤依賴圖 |
| F08-11 | 多 stream 互相等待形成 circular dependency | 偵測 deadlock/超時,不得假裝完成 |
| F08-12 | default stream 造成未預期 implicit synchronization | trace 揭露效能回退 |
| F08-13 | GPU graph replay 使用已失效 buffer pointer | version/lease guard 拒絕 |
| F08-14 | graph update 後 output schema 已變 | 拒絕不相容重新使用 |
| F08-15 | 三個 PWN 同時下發超出 VRAM 上限的 batch | admission reject/OOM fallback |
| F08-16 | 顯示或其他工作負載搶佔 GPU 記憶體 | 顯示資源競爭/容許 allocation failure |
| F08-17 | 批次排序丟失 slice_id 與 offset 映射 | 拒絕 candidate import |
| F08-18 | CPU worker handoff,舊 worker 接收 GPU 遲到結果 | owner_generation guard 拒絕 |
| F08-19 | GPU device reset,所有 tickets 狀態不明 | Unknown/recovery/quarantine |
| F08-20 | GPU kernel timeout 但底層仍在執行 | 不可立刻回收所有 buffers |
| F08-21 | R1 winner 已到,loser GPU 工作仍消耗資源 | wasted-work 仍列入總成本 |
| F08-22 | CPU 與 GPU 的 R2 結果同算法同錯誤來源 | 不發 IndependentWitness |
| F08-23 | R3 shards 索引重複卻宣稱足夠 份 | 拒絕恢復 |
| F08-24 | 某 GPU reduction 引入 NaN | 按 quality policy 拒絕/標註 |
| F08-25 | GPU 近似分數導致 AI expert 排名翻轉 | downstream semantic quality violation |
| F08-26 | 驗證全部使用同一 GPU 函數兩次 | 不宣稱獨立正確性 |
| F08-27 | 參考模型版本 v90,結果以 v91 提交 | Paper 04 版本拒絕 |
| F08-28 | 某批次少一片,剩餘片被補零後稱完全成功 | Partial/Hold,不可升格 |
| F08-29 | GPU 用來執行單一小布林運算 | 比 CPU-only 更慢的合理負對照 |
| F08-30 | profiler 顯示 kernel 變快,但 H2D 增加更多 | 不宣稱端到端加速 |
| F08-31 | Graphs 建置時間被隱藏而未攤提 | cold/warm/steady 分開報告 |
| F08-32 | GPU 利用率上升,正式提交吞吐量反降 | 不宣稱總系統收益 |
| F08-33 | AI planner 將未量測 VRAM 用量寫成零 | guard 拒絕 Unknown -> 0 的轉型 |
| F08-34 | 實驗只留下成功快樣本,刪除 OOM/timeout | 不符合公平實驗;原始資料保留 |
| F08-35 | RTX 3070 不支援某特定功能,卻以替代路徑冒充原測試 | 標示 NOT_SUPPORTED、另立基線 |
| F08-36 | GPU ticket 已成功但外部效果重試產生兩次 | 遵循 Paper 04 outbox/外部冪等邊界 |
20.1 四種反例的重要分組
- 安全類(如 F08-01~08、17~20、27、36): 一旦發生需停止對正式狀態的繼續提交、保留 evidence trace。
- 效能類(如 F08-12、29~32): 可作為失敗的正當科學結論,不必掩蓋。
- 方法學類(如 F08-33~35): 指向論文可重現性或數據造假風險,須修正實驗而非只調整控制參數。
- 品質類(如 F08-22~26): 說明 GPU 裝置正確執行不代表最終 AI 行為品質不變。
20.2 故障注入目標
最小 V0 要在每個狀態轉移之前/之後允許注入 crash、delay、drop_response、reorder、allocate_fail、stale_generation、wrong_schema、expired_snapshot 及 visibility_missing,並保留可重演 seed。V1 只做安全、受控的失效注入,不建議在日常桌面主機上強制 GPU driver reset 或隨意殺死影響桌面顯示的系統程序。
21. 可證偽效能實驗:公平基線與實驗矩陣
21.1 不能只拿不成熟的 naive baseline 比較
SMPDC 的跨協議整合若要形成有效研究貢獻,應與相同硬體/相同輸入上的成熟最佳可行候選比較,而不只是與完全串行、每次強制同步的低效程式比較。過往 StarPU、Legion、Kokkos 及 CUDA Graphs 都提供了異質排程、資料局部性或減少 launch 負擔的實際先例 [E5, E8–E9]。即便不能在同一平台原封不動安裝某 framework,也必須保留等價的控制實驗:batch、stream、graph、SIMD、資料常駐與 CPU 端最佳化不能只給 SMPDC 一組使用。
21.2 基準組 B0–B8 與實驗組 E0–E4
| ID | 執行設計 | 比較目的 |
|---|---|---|
| B0 | CPU 單執行緒參考算法 | 完整正確性 oracle,不作最強性能基線 |
| B1 | CPU thread pool/work stealing | 成熟多核心動態工作分配 |
| B2 | CPU 多核+SIMD/優化資料布局 | 公平檢查一般 CPU 能力 |
| B3 | CPU 事件驅動/per-core queue | 與持續工作流的既有低延遲做法比較 |
| B4 | CUDA 逐 batch stream + 正確事件依賴 | 基本 GPU offload,無無謂逐 kernel 強制 sync |
| B5 | CUDA pipeline/雙緩衝/pinned transfer(當能力支援) | 轉輸重疊與資料搬移控制 |
| B6 | CUDA Graphs 或等價可重用圖(當適用) | 排除 SMPDC 只是 graph reuse 重新命名 |
| B7 | 最佳適用 GPU library/kernel fusion/成熟框架版本 | 排除自製 kernel 無故落後 [E5, E12] |
| B8 | CPU/GPU 最佳基線中,按事先規範挑選的適用路徑 | 反映現有工程的務實選擇 |
| E0 | SMPDC 多 PWN,但 GPU 關閉 | 隔離多主/相位/協議開銷 |
| E1 | E0 + 固定 batch GPU adapter | 檢查單純 GPU 接入增量 |
| E2 | E1 + 相位/期限感知動態批次 | 檢查相位策略增量 |
| E3 | E2 + 常駐/複製路徑、R0–R3 部分設定 | 檢查資料流及冗餘邊際效益 |
| E4 | E3 + Paper 07 動態控制建議(先規則後可選 AI) | 檢查控制策略額外成本與收益 |
若 GPU 無法支援某基線,必須標記 NOT_SUPPORTED 並說明具體 capability;不可直接將缺席基線記為 SMPDC 獲勝。B7 在不同 workload 可能由不同現成工具擔任;應事前宣告 selection protocol,避免事後挑選較弱版本。
21.3 統一工作負載 W0–W10
| ID | 工作負載 | 主要控制變因 | 核心測量 |
|---|---|---|---|
| W0 | 單條微小消息判斷 | 每訊息 bytes、發生頻率 | 負對照:GPU 不應被假定獲勝 |
| W1 | 大量獨立關係評分 | N、維度、精度 | 高吞吐批次是否有淨益 |
| W2 | 多 PWN 動態耦合圖 | edge count、稀疏度、重配速率 | phase-aware batching 收益 |
| W3 | 稀疏鄰接聚合 | degree distribution、熱點 | 原子競爭與訪存成本 |
| W4 | R1 CPU/GPU 雙路競速 | latency variance、尾端故障率 | p99/wasted GPU energy |
| W5 | R2 異源數值佐證 | independent algorithm 成本 | 證據獨立性/質量正確性 |
| W6 | R3 分片編碼/恢復 | shard bytes、 | 真實 H2D/D2H + encode/decode |
| W7 | 推測分支 GPU 預演 | invalidation ratio、branch depth | rollback cost/有效提交率 |
| W8 | 經常重複的 GPU 子圖 | reuse count、shape churn | Graphs vs dynamic scheduling |
| W9 | AI 模型外部 router/MoE gate | token batch、expert imbalance | step throughput + semantic policy |
| W10 | CPU 持續 Agent + GPU embedding/檢索核 | arrival burst、tokenization load | 完成任務 p95/p99、deadline miss |
W9–W10 是模型化或輕量可執行的 AI workload 候選;不得把 synthetic vector benchmark 宣稱成完整大型模型預訓練的成功證據。真正 AI 訓練研究需 Paper 10 加入模型品質與收斂速度驗證。
21.4 比較公平性契約
每一成對實驗至少固定:相同 input hash、同樣輸入到達序列(或相同隨機 seed)、相同品質政策、相同工作完成定義、相同 CPU/GPU/RAM/VRAM 配額、相同 deadline、相同佇列上限、相同冷/熱啟動條件、相同功率模式與盡可能一致的背景負載。所有 CPU-only 組也應得到相應的 thread pinning/SIMD/cache locality 最佳化機會。若一組的 cache 是 warm 而另一組是 cold,必須分開報告而不混在單一 speedup 中。
21.5 統計及可重現性
固定 baseline 版本後,對每一設定多次重複並保留原始樣本;報告中位數、p95、p99(若樣本量足夠)、bootstrap 信賴區間或其他已宣告不確定性評估、失敗率、實際 deadline misses,且說明量測時鐘與 timer resolution。對 timeout/OOM 等截尾樣本不能只保留成功的快樣本,亦不能把未測 energy 記為 0。跨不同功率設定與驅動版本的資料不得直接合併,需分層或明確建模混雜因素。
21.6 主要效能指標
定義符合相同 quality contract 且已正式確認完成的工作數 ,測試持續時間 ,則:
若能源量測可用、完成合法工作共消耗 joule,則:
對純 kernel benchmark,合法完成定義可以是 LocalVerified,但必須另外標示;若跨入權威提交域,不能以 KernelDone 作為分子。對大型訓練模型則需要有效 optimizer step、token/s、time-to-quality 及模型損失/評估表現,而不只是 kernel TFLOPS。
21.7 預註冊失敗條件
若某個 workload 在公平最強基準 B8 下,比 E4 明顯更快,且 E4 的額外協議沒有改善可觀測的安全或可靠性指標,應承認該 workload 不值得引入完整 SMPDC。若少數 W1/W2 得益,但 W0/W7 顯著惡化,應將結果寫為適用域而不是宣稱系統普遍優越。若所有測試的端到端 gains 均為零或負,當前實作路線在性能目標上被實驗反駁,下一輪只能根據事前接受的優化假說重試,不能改寫成功定義。
22. V0:不依賴 GPU 的參考語義與模型測試
22.1 V0-A:有限狀態事件模擬器
V0-A 以離散事件模擬 PWN、GpuExecutionTicket、DeviceBufferLease、StreamDependency、HostVisibility 與 Paper 04 commit proposal。每個 transition 帶 generation、guard、runtime timestamp、error injection 與 trace correlation。它要能重現本篇安全不變量的正例與反例,而不需要現實 CUDA 支援。
22.2 最小模型狀態
MockGpuStateV0_1 {
flow_owners: Map[FlowId, MonotonicGeneration]
buffers: Map[BufferLeaseId, BufferState]
tickets: Map[TicketId, TicketState]
stream_edges: Set[(EventId, EventId)]
events_completed: Set[EventId]
host_visibility: Set[(BufferLeaseId, BufferGeneration)]
credit_ledger: CreditLedger
faults: FaultQueue
trace: EventLog
}
模擬器不得隱式把 events_completed 的任何 event 當成 host-read readiness,而須為具體 buffer 檢查其可見性依賴;也不得因 CPU worker 被銷毀而自動釋放所有 GPU ticket buffer。
22.3 V0-B:類型與協議 conformant 驗證
用 Paper 05–06 的 GPU_* 指令與 PDAIS-IR 驗證以下非法程式:
- 沒有
GPUAlloc/有效 lease 就進行GPUDispatch; - 尚未 copy/event 完成便將
GPUEventcast 為HostVisible; - 在 producer 和 consumer 不同 streams 中缺乏依賴;
- 在 kernel still in-flight 時執行
GPURelease並假裝立刻成功; - 把
GpuCandidateOutcomecast 為Committed; - 不同
flow_id的輸出經錯誤 batch offset 回傳; - generation 已改變卻容許舊 worker 寫入;
- 以相同 R1 來源的兩份 GPU 結果構造
IndependentWitness。
若編譯器尚無法以靜態方式證明,必須執行動態 guard 並可在 trace 中驗證其拒絕行為;「靜態檢查沒報錯」不代表 runtime safety 已完成。
22.4 V0-C:成本模擬器不是實測
給定合成的 CPU/GPU service-time 分布、PCIe bandwidth、batch wait policy 與 failure events,可以比較不同 admission 策略的預期 queueing、tail latency、wasted work 等。在 V0-C 的報表上必須明顯標示 SIMULATED_PARAMETERS;假設的 transfer bandwidth 不得引用為真實 RTX 3070 測試結果。V0 提供安全性與設計空間的探索,不能替代 V1 的實機證據。
23. V1:現有桌機上的真實 CPU–GPU 實驗
23.1 可用的首輪硬體背景
本次研究記錄的現有桌機為舊款 AMD CPU、32 GB RAM、NVIDIA RTX 3070,主電腦尚未升級;不存在已購買的 MINISFORUM N5 Pro AI NAS。CPU 精確型號、核心數、PCIe generation/link width、GPU 實際可用 VRAM、顯示 driver、CUDA Runtime、Windows 或 Linux 版本均應由當次測試 manifest 讀取,不可憑估計或過去購物比較填入。這套硬體已足夠做 CPU-only、host/device copy、CUDA stream/event、Graph 與 GPU batching 的初步實驗;不必為 Paper 08 先購買高價 P/E CPU 或多 CCD 平台。
23.2 V1-0:環境與 capability discovery
先以只讀方式觀測硬體/driver/系統資訊,建立 device query、copy mode 支援、managed memory 限制、graph capture 支援與 stream memory pool 能力清單。用最小 CUDA sample 驗證 H2D、kernel、D2H、event wait 的實際可用性。若用 WSL 或 Windows 原生編譯器,必須以真正運作的 execution backend 和版本為準,不能假定兩者的 memory model 或 profiler 特性相同。
23.3 V1-1:正確性與資料可見性
以極簡純運算子(例如每元素有限整數變換、明確 overflow policy)建立 CPU oracle 和 GPU 版本,先做:
- 單 stream H2D -> kernel -> D2H -> event wait;
- 缺少 event 的非法程式由 verifier/test guard 拒絕;
- cross-stream 等待及 buffer 生命週期測試;
- ticket retry/timeout/stale generation 的模擬 fault injection;
- C08/G08 invariant 與 trace 檢查。
僅當該步成功,才進入複雜 PWN、R1–R3 與 AI 工作負載,以免把 GPU 應用的基本 data race 當成新理論的拓撲問題。
23.4 V1-2:CPU與GPU共同成本量測
在不同 、不同輸入 bytes、不同複用率、不同 batch wait、不同 stream 數、不同工作流數的網格上運行 B0–B8、E0–E3;保留 host 端端到端時鐘和必要的 CUDA events。GPU event elapsed time 只用於 device 局部階段,不得直接替代 host latency [E1]。
23.5 V1-3:動態工作流和尾延遲
對 W2/W4/W7/W10 注入突發流量、長短混合任務、在安全範圍內的暫態延遲與 CPU owner handoff。檢查 GPU batch 成形是否造成長 deadline miss、R1 loser 的 wasted VRAM/GPU time 是否影響 p99,以及 PWN 在 kernel 未完成時能否繼續處理其他合法主工作流事件。
23.6 V1-4:性能工具與測量擾動
Nsight Systems 可觀測 CUDA API、copy、kernel、stream timeline,適合查明是否存在額外隱式同步、idle gap 或 serialization [E7]。對高頻小 kernel,profiler 本身也可能增加儀器負擔;應同時提供不用 profiler 的 host wall-clock 結果及用 profiler 的機制分析,不應把已受 trace overhead 影響的數值當成純延遲。
23.7 維持工作站可用的界線
不要在日常桌機強制將 GPU 驅動 reset、把所有系統 CPU affinity 變更、強行移除桌面 display driver、關閉必要保護,或以無限制 allocation 填滿 VRAM。高負載長期實驗應設定最大溫度/功率/記憶體壓力政策(需基於實機安全範圍),並備有安全中止與完整 log 保存程序。任何故障注入需隔離到受控測試進程/容器或 VM,並保證不傷害其他使用者資料。
24. AI 訓練、推論與持續 Agent 的實際適用域
24.1 AI 是重要應用目標,不是預先成功證明
SMPDC 的動機包含 AI 訓練與推論,但現有 AI 框架已大幅使用 GPU 大型矩陣運算、算子融合、圖式執行及非同步資料供應。對訓練中的密集 GEMM/attention kernel,額外 CPU 節點協議可能根本無法改善瓶頸,甚至增添延遲。較有希望的應用通常出現在 CPU 端仍需大量動態路由、稀疏/不規則處理、實驗性驗證、多個持續狀態或較高的 batch ready churn;該希望仍須比較公平基線。
24.2 MoE 專家選擇與動態路由
可將 gate scoring、expert route selection 的候選評估、負載預估及 dispatch metadata 分割為 host 控制與 device 數值子圖。對大 batch,若路由特徵常駐 GPU,繼續在 GPU 處理可能較理想;若 CPU 每 token 都要回收完整 routing vector 再重新傳到 GPU,可能浪費端到端頻寬。評估必須包含 expert imbalance、top-k 決策翻轉、device gather/scatter、模型品質與 token throughput。
24.3 長時間持續 Agent 的異質分工
Agent 的工具呼叫/檔案處理/決策流程不一定適合 GPU;某些向量化檢索、影像轉換、批次 embedding、相似度運算則可能適合。讓 CPU PWN 維持長期狀態與因果責任,對可平行化片段使用 GPU,能提出合理的實驗假說。但 GPU kernel 本身不能證明 Agent 的計畫正確、社會互動可靠或外部效果已結清。
24.4 具證據要求的 AI 研究流水線
對證明檢查、符號/數值混合研究,GPU 可以作為特定大量數值候選生成器;邏輯證書、形式證明、版本及權威提交通常需另一套可信驗證器。GPU 生成的百萬個候選就算全部通過 hash checksum,也不構成數學定理成立。這是對模型成果與硬體完成證據的必要區分。
24.5 單 GPU 與多 GPU 的研究界線
本篇只要求單 GPU 的 CPU–GPU 混合協作作為 V1 最低實驗。若未來擴至多 GPU/分散式訓練,NCCL 已有成熟的 AllReduce、Broadcast、AllGather、ReduceScatter 等裝置間 collective primitives [E10]。SMPDC 不可把現有 GPU collective 通訊重新命名為「相位擴散」就宣稱新增硬體速度;必須測量其跨協議選擇、權限與工作流持續性是否在同等 correctness contract 下帶來額外收益。
24.6 模型訓練真正的完成口徑
對 AI 訓練最重要的不只是每秒 GPU kernel 次數,還包含有效梯度步數、資料是否重複或遺漏、精度/loss scaling、optimizer 狀態版本、checkpoint 可恢復性、最終評估品質與 time-to-target-quality。若 GPU offload 路線在每秒 KernelDone 上改善,卻增加 gradient stale update 或 rollback,便可能沒有端到端訓練效益。Paper 10 將把此層做成完整 AI 應用驗證,而不是讓 Paper 08 提前聲稱已完成。
25. 與既有系統的嚴格比較與新穎性邊界
25.1 CUDA streams/events 與 CUDA Graphs 已解決哪些問題
CUDA streams 提供非同步命令序列,events 可支援跨 stream 依賴及觀測完成;CUDA Graphs 減少重複工作提交成本,也能在條件允許時採 update/capture;stream-ordered allocator 定義與 stream 相關的 allocation/free 順序 [E1, E4–E6]。因此 GpuExecutionTicket 若只重包 cudaEvent_t,並未創造新方法。SMPDC 的可驗證增量必須是將上述設備原語與 PWN identity、phase、R0–R3、fencing、commit、trace 及資料跨域權限聯合約束的語義與可測量結果。
25.2 Legion、StarPU、Kokkos、SYCL 的前例
Legion 對 heterogeneous processing 與 data locality 的 region-based 管理有成熟研究;Kokkos 明確區分 execution space 與 memory space 以及 deep copy 的同步/非同步條件 [E8–E9]。StarPU 與 SYCL 等亦具有異質執行、資料管理及排程能力。這意味「CPU+GPU 異質協作」本身不是新穎發明。本系列須提出、實作並測試多主持續工作流與其有界擴散/冗餘/提交語義的特定交叉約束,且在實驗負面時保留負面結論。
25.3 三種可能成立的增量,不是已證實的結論
- 語義增量: GPU 執行被納入統一的可檢查效果系統,降低跨後端的不安全資料可見性或權威混淆錯誤。
- 系統增量: 針對特定持續 PWN+相位就緒模式,batch/placement 控制可能改善尾端延遲或穩態吞吐。
- 工程增量: 可重用的跨 OS/CPU/GPU adapter、trace schema、fencing 與故障恢復程式,降低系統整合成本。
第一項偏向可驗證程式系統,第二項偏向性能,第三項偏向軟體工程品質。不能因只達成一項便宣稱其他兩項已得到證明。
25.4 明確不主張的新穎性
本篇不主張發明:GPU 平行 kernel、device streams、CUDA Graphs、pinned memory、unified memory、dataflow、host/device synchronization、CPU affinity、work stealing、CRDT、erasure coding、MVCC、GPU collectives、神經形態計算或 CPU/GPUs 的共享物理連線。對這些成果的尊重是設計公平基線的必要前提。
26. 與 Paper 09–11 的正式交接
26.1 Paper 09|記憶體與微觀硬體映射
必須承接: DeviceBufferLeaseV0_1、GpuDeviceCapabilitySnapshotV0_1、GpuExecutionTicketV0_1、DeviceVisibilityWitnessV0_1、GpuCollaborationTraceV0_1。研究 CPU cache/NUMA/CCD/P-E/SMT、device memory、PCIe/互連、資料常駐及 buffer placement 的真實成本。需建立不同拓撲的 HardwareCostProfile,並說明哪些相位擴散收益可以透過現有 cache/NUMA locality 得到,哪些需要新互連才能成立;不能單靠圓形網絡圖主張某晶片路由已更快。
26.2 Paper 10|AI 訓練與推論適用域
必須承接: W9/W10、R0–R3、quality contract、模型版本、optimizer/seed lineage、有效 step/token throughput 及 time-to-quality。需與成熟 PyTorch/CUDA/模型專用推論實作比較,並明確區分 CPU 主控策略、數值算子、routing、資料搬移和 GPU 計算瓶頸。
26.3 Paper 11|系統驗證與可證偽實驗
必須承接: 本篇 B0–B8/E0–E4、G08-01~G08-32、F08-01~F08-36、V0/V1 manifest、統計與故障注入。要建立正式驗收條件,使「新架構沒有達到性能收益」可以成為合法結論。若硬體不可用或某 API 不受支援,須明示而不能替代性地寫成 PASS。
26.4 對後續硬體或新指令的限制
只有在現有最佳基線下識別出穩定、可重複的特定 bottleneck(如跨核心訊息路由、host/device copy、頻繁 CPU launch、stream/event overhead 或共享記憶體 contention),才合理提出微觀新 ISA/專用橋接硬體的候選規格;還需估計面積、功率、複雜性、安全、編譯器支持及製造代價。中宏觀 SMPDC 的存在本身不構成「需要新晶片」的充分證據。
27. 研究結論與不可越界的主張
SMPDC Paper 08 將原先的「CPU 各核心持續主工作流,透過相位呼叫耦合,以 GPU 並行加速資訊關係」落為一個可以對既有硬體實作、比較及反駁的異質協作協議候選。它的核心不是讓 GPU 取代 CPU,而是維持 CPU 主工作流對狀態與決策的局部責任,讓 GPU 僅處理足夠大、可並行且適合設備資料路徑的計算子圖。
為達此目標,本篇定義的關鍵工程分離包括:PWN identity 與 GPU queue 分離、logical message 與 batch buffer 分離、GPU submission 與 kernel completion 分離、device completion 與 host visibility 分離、局部候選與權威提交分離、效能估計與實測分離,以及 AI planning 與 OS/Runtime 資源權限分離。這些分離不是修辭,而是會影響 memory safety、resource leaks、race、故障恢復與實際效能的規範性條件。
本篇最重要的可證偽假說為:在特定具有持續局部狀態、動態相位資格、足夠批次的關係運算與可受控資料常駐特性的 workload 上,以具型別異質執行與嚴格 PWN/phase/credit/commit 接口組成的 SMPDC 實作,可能在公平成熟基線的約束下改善某一項端到端指標;但對微小、序列依賴強、搬移成本過高或已由成熟 CUDA 圖式執行充分優化的 workload,可能沒有優勢或明顯更差。
目前尚未完成這項假說的性能實驗。Paper 08 v0.1 的正式完成項目為:候選協議、狀態機、條件命題、失效反例、資料型別、比較矩陣與原稿;它不構成已完成的 GPU 工程產品,也不構成新的 CPU/GPU 物理架構已被證明優越的報告。
附錄 A|GpuCollaborationTraceV0_1、可追溯事件與示例
A.1 Trace 封套
GpuCollaborationTraceV0_1 {
trace_id: TraceId
run_id: RunId
source_revision_digest: Digest
event_id: MonotonicSequenceOrUUID
local_monotonic_time_ns: NonNegativeInteger
clock_domain_id: ClockDomainId
event_class: TypedEventClass
flow_id_optional: Optional[FlowId]
flow_owner_generation_optional: Optional[Generation]
event_epoch_optional: Optional[EventEpoch]
message_id_optional: Optional[LogicalMessageId]
lineage_id_optional: Optional[LineageId]
slice_id_optional: Optional[SliceId]
batch_id_optional: Optional[BatchId]
ticket_id_optional: Optional[TicketId]
gpu_device_id_optional: Optional[DeviceId]
cuda_stream_id_optional: Optional[StreamId]
cuda_event_id_optional: Optional[EventId]
buffer_lease_id_optional: Optional[BufferLeaseId]
buffer_allocation_generation_optional: Optional[Generation]
input_version_digest_optional: Optional[Digest]
quality_policy_ref_optional: Optional[QualityPolicyRef]
device_capability_generation_optional: Optional[Generation]
host_visibility_witness_ref_optional: Optional[WitnessRef]
authority_scope_ref_optional: Optional[AuthorityScopeRef]
outcome: Pending | Accepted | Rejected | Failed | Unknown
reason_code_optional: Optional[TypedReasonCode]
input_bytes_optional: Optional[Bytes]
output_bytes_optional: Optional[Bytes]
device_memory_reserved_bytes_optional: Optional[Bytes]
cpu_time_ns_optional: Optional[NonNegativeDuration]
device_elapsed_time_ns_optional: Optional[NonNegativeDuration]
copy_elapsed_time_ns_optional: Optional[NonNegativeDuration]
ledger_balance_digest_optional: Optional[Digest]
collector_confidence: Observed | Derived | Estimated | Unknown
}
A.2 示意 trace:CPU 主節點透過 GPU 完成大批關係評分
下表是概念追蹤,NOT_EXECUTED;它不含任何真實測得的微秒數。
| 序列 | 事件 | 必要條件 | 可推出什麼、不可推出什麼 |
|---|---|---|---|
| 1 | PWN-A/B 同時 PhaseReady | 各自 owner/epoch 正確 | 可以考慮建立批次;未必能上 GPU |
| 2 | BatchPlan B accepted | quality/authority/budget 兼容 | 可以配置資源;沒有完成任何 kernel |
| 3 | BufferLease L allocated | 地址與 allocation generation 登錄 | 允許受控搬移,不允許提早 free |
| 4 | H2D issued | source pinned/pageable 記錄 | 命令已提交;資料未必已到達 GPU |
| 5 | Kernel issued | 合法依賴 H2D | 計算已排入佇列;並非立即產生候選 |
| 6 | KernelDone event observed | event 位置與流已驗證 | kernel 完成;不代表 D2H 完成 |
| 7 | D2H transfer event observed | 對應 copy 依賴合法 | 可以檢查 host 可讀條件 |
| 8 | HostReadable(L) issued | 同步與 lease/版本均合法 | CPU 可讀指定 buffer,不代表內容可信 |
| 9 | PWN-A local verification | 來源及品質政策通過 | LocalOutcome 可輸出;還不是正式提交 |
| 10 | COMMIT_VALIDATE/DECIDE | 目前版本及授權通過 | 才可能產生唯一受控提交決策 |
| 11 | Lease retire | 無 outstanding device/host user | VRAM/host buffer 可安全回收 |
A.3 示意 trace:版本過期的 GPU 結果
run=SIMULATED / NOT_EXECUTED
1 Flow F generation 4, snapshot 51 -> ticket T51 submitted
2 GPU kernel completes, D2H completes, HostReadable(L,51) is valid
3 canonical state already advanced to snapshot 52
4 PWN attempts CommitProposal using T51
5 authority guard -> REJECT_VERSION_CONFLICT
6 T51 remains a historically valid device computation,
but is not a valid current canonical commit
7 outstanding leases retire only after all users finish
這個示例顯示至少兩種獨立有效性:設備可見性可以成功,正式狀態版本合法性卻同時失敗。它不代表 GPU 算錯了。
A.4 Trace 關聯及時計邊界
CPU host timestamp、CUDA events、NSight timeline 與跨執行緒事件偏序必須記錄各自 clock domain。若需要以 device time 與 host time 估算 overlap,應使用明確 correlation ID 與工具提供的同步/校準方式,而非隨意把不同時鐘的 timestamp 相減。正式報表應同時給出端到端 host wall time、GPU event duration 與 profiling 工具計算的歸因區間,並揭露未校準部分。
附錄 B|有限狀態 Mock GPU 協議微型參考程式
B.1 目的與限制
以下程式只測試一小部分抽象狀態機:flow generation、一次性 credit、ticket 完成順序、HostReadable、BufferLease 最後使用者與 owner handoff。它不執行 CUDA、不模擬 GPU memory model、不宣稱機械證明或 SMPDC V1 benchmark。用 Python 僅為可快速檢查 Paper 08 核心規則的示例;無 runtime device driver 及外部系統權限。
B.2 可執行參考程式(Python 3,標準函式庫)
from dataclasses import dataclass, field
@dataclass
class Lease:
generation: int
users: set[str] = field(default_factory=set)
host_visible_tickets: set[str] = field(default_factory=set)
retired: bool = False
@dataclass
class Ticket:
flow: str
owner_generation: int
lease_id: str
credit_id: str
kernel_done: bool = False
copy_done: bool = False
closed: bool = False
class MockGpuProtocol:
def __init__(self):
self.flow_owners = {"A": 1}
self.leases = {"L": Lease(generation=1)}
self.tickets = {}
self.used_credits = set()
def submit(self, flow, owner_generation, lease_id, credit_id):
assert self.flow_owners[flow] == owner_generation, "stale owner"
assert credit_id not in self.used_credits, "duplicate credit"
lease = self.leases[lease_id]
assert not lease.retired, "retired buffer"
ticket_id = f"T{len(self.tickets) + 1}"
self.used_credits.add(credit_id)
lease.users.add(ticket_id)
self.tickets[ticket_id] = Ticket(
flow=flow, owner_generation=owner_generation,
lease_id=lease_id, credit_id=credit_id,
)
return ticket_id
def mark_kernel_done(self, ticket_id):
ticket = self.tickets[ticket_id]
assert not ticket.closed, "already closed"
ticket.kernel_done = True
def mark_copy_done(self, ticket_id):
ticket = self.tickets[ticket_id]
assert not ticket.closed, "already closed"
assert ticket.kernel_done, "D2H before kernel completion"
ticket.copy_done = True
def witness_host_readable(self, ticket_id):
ticket = self.tickets[ticket_id]
assert not ticket.closed, "already closed"
assert ticket.kernel_done and ticket.copy_done, "not visible"
lease = self.leases[ticket.lease_id]
assert not lease.retired and ticket_id in lease.users
lease.host_visible_tickets.add(ticket_id)
return (ticket_id, ticket.lease_id, lease.generation)
def import_candidate(self, ticket_id, witness):
ticket = self.tickets[ticket_id]
assert not ticket.closed, "closed ticket"
assert self.flow_owners[ticket.flow] == ticket.owner_generation, "stale import"
lease = self.leases[ticket.lease_id]
assert witness == (ticket_id, ticket.lease_id, lease.generation)
assert ticket_id in lease.host_visible_tickets and not lease.retired
return {"kind": "LocalCandidate", "ticket": ticket_id}
def handoff(self, flow, expected_generation, next_generation):
assert self.flow_owners[flow] == expected_generation
assert next_generation > expected_generation
self.flow_owners[flow] = next_generation
for ticket in self.tickets.values():
if ticket.flow == flow and not ticket.closed:
ticket.owner_generation = next_generation
def close_ticket(self, ticket_id):
ticket = self.tickets[ticket_id]
assert not ticket.closed, "already closed"
assert ticket.kernel_done and ticket.copy_done, "in-flight work"
self.leases[ticket.lease_id].users.remove(ticket_id)
ticket.closed = True
def retire(self, lease_id):
lease = self.leases[lease_id]
assert not lease.retired, "already retired"
assert not lease.users, "outstanding users"
lease.retired = True
def expect_reject(fn):
try:
fn()
except AssertionError:
return True
raise AssertionError("expected rejection was not enforced")
def run_tests():
p = MockGpuProtocol()
checks = 0
t = p.submit("A", 1, "L", "credit-01")
checks += 1
assert expect_reject(lambda: p.witness_host_readable(t)); checks += 1
assert expect_reject(lambda: p.retire("L")); checks += 1
assert expect_reject(lambda: p.mark_copy_done(t)); checks += 1
p.mark_kernel_done(t)
assert expect_reject(lambda: p.witness_host_readable(t)); checks += 1
p.mark_copy_done(t)
proof = p.witness_host_readable(t)
assert proof == (t, "L", 1); checks += 1
assert p.import_candidate(t, proof)["kind"] == "LocalCandidate"; checks += 1
assert expect_reject(lambda: p.submit("A", 1, "L", "credit-01")); checks += 1
assert expect_reject(lambda: p.submit("A", 0, "L", "credit-02")); checks += 1
t2 = p.submit("A", 1, "L", "credit-02")
checks += 1
assert expect_reject(lambda: p.import_candidate(t2, proof)); checks += 1
p.mark_kernel_done(t2)
p.mark_copy_done(t2)
proof2 = p.witness_host_readable(t2)
assert proof2 != proof; checks += 1
assert p.import_candidate(t2, proof2)["kind"] == "LocalCandidate"; checks += 1
p.handoff("A", 1, 2)
assert p.import_candidate(t, proof)["kind"] == "LocalCandidate"; checks += 1
assert expect_reject(lambda: p.submit("A", 1, "L", "credit-03")); checks += 1
p.close_ticket(t)
assert expect_reject(lambda: p.retire("L")); checks += 1
p.close_ticket(t2)
assert expect_reject(lambda: p.import_candidate(t2, proof2)); checks += 1
p.retire("L")
assert p.leases["L"].retired; checks += 1
assert expect_reject(lambda: p.submit("A", 2, "L", "credit-04")); checks += 1
assert expect_reject(lambda: p.retire("L")); checks += 1
return checks
if __name__ == "__main__":
n = run_tests()
print(f"Mock protocol PASS: {n}/{n} assertions")
B.3 這些測試沒有提供哪些保證
附錄程式在本輪以 Python 3 實際執行,20/20 項示例斷言通過;這只確認該有限 mock 的本輪控制路徑。程式中的 Python assert 只用來驗證示例控制流程,不是能抵抗惡意環境的安全 guard;執行時不應以 python -O 取消必要斷言。真實系統還需要不可偽造的能力與 buffer handle、跨執行緒 atomic ledger、GPU driver 的非同步錯誤處理、device event API、OS 資源隔離、CUDA stream-order 和 host/device memory model,以及充分的 fault injection。此示例沒有覆蓋 Paper 08 全部 32 條不變量,不能僅因單次 PASS 就把本篇成熟度升級至 L2/L3。
附錄 C|V0/V1 研究 Manifest、預註冊與驗收清單
C.1 最小 YAML manifest(候選)
manifest_version: "smpdc-gpu-collaboration/0.1"
run_id: "example-not-executed"
execution_status: "NOT_EXECUTED"
source_artifact:
filename: "SMPDC_Paper08_CPU_GPU_Heterogeneous_Collaboration_v0.1_2026-10-17.md"
sha256: "TO_BE_COMPUTED"
implementation:
repository_revision: "TO_BE_RECORDED"
runtime_adapter_version: "TO_BE_RECORDED"
compiler_version: "TO_BE_RECORDED"
hardware:
cpu_model: "TO_BE_OBSERVED"
cpu_physical_cores: null
cpu_logical_processors: null
host_memory_bytes: null
gpu_model: "TO_BE_OBSERVED"
gpu_device_capability: "TO_BE_OBSERVED"
gpu_vram_bytes: null
pcie_link_info: "TO_BE_OBSERVED"
software:
os_name_and_build: "TO_BE_OBSERVED"
gpu_driver_version: "TO_BE_OBSERVED"
cuda_runtime_version: "TO_BE_OBSERVED"
graph_capture_supported: "UNKNOWN"
managed_memory_mode: "UNKNOWN"
cpu_affinity_mode: "TO_BE_RECORDED"
experiment:
workload_id: "W1"
baseline_id: "B2"
treatment_id: "E2"
input_hash: "TO_BE_COMPUTED"
quality_contract_id: "TO_BE_DEFINED"
deadline_contract_id: "TO_BE_DEFINED"
phase_policy_digest: "TO_BE_COMPUTED"
batch_policy_digest: "TO_BE_COMPUTED"
isolation_policy_digest: "TO_BE_COMPUTED"
gpu_memory_budget_bytes: null
pinned_memory_budget_bytes: null
repetitions: 0
warmup_iterations: 0
measurement_seed: null
profiler_mode: "TO_BE_RECORDED"
background_workload_policy: "TO_BE_RECORDED"
results:
end_to_end_latency_ms: null
p50_ms: null
p95_ms: null
p99_ms: null
legal_completions_per_second: null
candidate_quality_violations: null
deadline_misses: null
gpu_kernel_total_ms: null
h2d_copy_total_ms: null
d2h_copy_total_ms: null
gpu_queue_total_ms: null
cpu_runtime_total_ms: null
gpu_energy_j: null
cpu_energy_j: null
measurement_uncertainty: "NOT_MEASURED"
raw_trace_bundle_digest: "TO_BE_COMPUTED"
safety:
g08_invariants_checked: []
g08_invariant_violations: null
f08_fault_cases_executed: []
unknown_ticket_outcomes: null
buffer_leases_pending_retire: null
C.2 驗證閘門
- C08-GATE-01: Canonical UTF-8 source 與引用可校驗。
- C08-GATE-02: V0 mock/reference model 的單步規則有明確狀態轉移。
- C08-GATE-03:
Submitted、KernelDone、HostReadable、LocalVerified、Committed可在 trace 中清楚區別。 - C08-GATE-04: 在已覆蓋的子集中 G08 invariant 全部通過;未覆蓋者標示
NOT_TESTED。 - C08-GATE-05: 已選的 F08 反例可重演,證明預期拒絕行為,而不是只測 happy path。
- C08-GATE-06: B0/B1/B2 至少具有正確 CPU oracle 與公平性能對照。
- C08-GATE-07: GPU driver/capability、stream/events、transfer mode 可被實機查詢。
- C08-GATE-08: B4/B5/B6 依實際支援度完成,不能使用不合法的虛假 Graphs 版本。
- C08-GATE-09: CPU-only、GPU-only 及 SMPDC end-to-end 都計入 bytes moved、sync、CPU 控制與驗證。
- C08-GATE-10: timeout、OOM、device error、被取消的未結清 GPU work 均被保留。
- C08-GATE-11: 能源缺值標記為
NOT_MEASURED,不能當零。 - C08-GATE-12: 任何 speedup 結論具有 workload/driver/quality/硬體適用範圍。
- C08-GATE-13: 不以理論上的 GPU 吞吐值或廠商行銷數字取代測得的 application-level latency。
- C08-GATE-14: Paper 09–11 交接符號與契約的相容性已確認。
C.3 固定負面對照
在所有報告中至少保留一個 W0,其操作極小、依賴序列化且資料原本在 host。該對照可用來直接反駁「GPU 一定讓資訊流速更快」。同時保留 W8,使完整 GPU graphs baseline 有機會展現成熟系統已能消除大部分 CPU launch overhead。若 SMPDC 在這兩個對照上沒有額外收益,不得更改工作量或刪除測試以維持某種「全域加速」敘事。
附錄 D|來源、數學符號、版本治理及研究誠信
D.1 Canonical 形式化符號表
| 符號 | 含義 | 主要出處 |
|---|---|---|
| SMPDC 九元中觀狀態 | Paper 00 [I0] | |
| 持續主要工作流集合 | Paper 00–01 [I0–I1] | |
| PWN 至 CPU 實體承載的可變映射 | Paper 00–01 [I0–I1] | |
| 與互動資格相關的相位資訊 | Paper 02 [I2] | |
| 訊息及傳遞狀態 | Paper 00–02 [I0–I2] | |
| 本篇 GPU 後端具體化狀態 | Paper 08,非原九元新公理 | |
| 某一批次的關係/數值評分算子 | Paper 08 | |
| 同口徑 CPU/GPU 候選完成時間 | Paper 08 | |
| 合法完成工作吞吐量 | Paper 08 | |
| 每焦耳合法完成工作量 | Paper 08 | |
| 單路、競速、佐證、編碼冗餘 | Paper 03 [I3] | |
HostReadable |
host 遵循契約可安全讀特定 buffer 的 witness | Paper 05–08 [I5–I7] |
D.2 研究結論口徑與實驗不可替代性
此版本已完成: 形式契約、候選狀態機、具型別資料結構、八項條件性命題、32 項不變量、36 類故障反例、基線矩陣、V0/V1 驗證設計與附錄微型範例。
此版本尚未完成: C++/CUDA 的真實 SMPDC GPU runtime、GPU driver profiler trace、在現有 RTX 3070 上的受控端到端 A/B 實測、AI 訓練品質/收斂評估、獨立審查、形式定理的機械化證明、任何新晶片或實體 ISA。
凡涉及先前稿件中的「PASS」,應以該稿自己提供的測試證據為準;它不會自動使 Paper 08 的真實 GPU 效能實驗成立。內部理論只是概念起源及接口材料,不能作為公開硬體加速事實的直接證據。
D.3 Canonical source 規則
本稿唯一正式原始碼為 UTF-8 Markdown 檔,不以聊天渲染頁面為原稿;數學使用 $...$ 和 $$...$$ 作為唯一 canonical delimiter。正式原稿不經 unicode_escape round trip、不將 LaTeX 替換為 Unicode 數學字後當 source、不任意改變 backslash/公式空白與分隔符。正式發布或提交前須 validate 檔案完整性、符號及引用,再建立 commit。此版的編修應以檔案 diff 和原始碼驗證為準。
參考文獻與原始來源
內部研究源(本系列直接依賴)
- [I0] Neo.K × Aletheia(2026-10-16)。《SMPDC Paper 00|群星多主節點持續擴散計算論:形式總綱》,v0.1。
SMPDC_Paper00_Stellar_Multi_Primary_Persistent_Diffusive_Computing_v0.1_2026-10-16.md。 - [I1] Neo.K × Aletheia(2026-10-16)。《SMPDC Paper 01|多主工作流節點論》,v0.1。
SMPDC_Paper01_Multi_Primary_Workflow_Nodes_v0.1_2026-10-16.md。 - [I2] Neo.K × Aletheia(2026-10-16)。《SMPDC Paper 02|相位耦合與有界持續擴散動力學》,v0.1。
SMPDC_Paper02_Phase_Coupling_and_Bounded_Persistent_Diffusion_v0.1_2026-10-16.md。 - [I3] Neo.K × Aletheia(2026-10-16)。《SMPDC Paper 03|冗餘資訊流與主節點聚合協議》,v0.1。
SMPDC_Paper03_Redundant_Information_Flow_and_Primary_Node_Reconciliation_v0.1_2026-10-16.md。 - [I4] Neo.K × Aletheia(2026-10-16)。《SMPDC Paper 04|推測運算與確定性提交》,v0.1。
SMPDC_Paper04_Speculative_Execution_and_Deterministic_Commit_v0.1_2026-10-16.md。 - [I5] Neo.K × Aletheia(2026-10-16)。《SMPDC Paper 05|持續擴散指令集》,v0.1。
SMPDC_Paper05_Persistent_Diffusive_Instruction_Set_and_Executable_Semantics_v0.1_2026-10-16.md。 - [I6] Neo.K × Aletheia(2026-10-16)。《SMPDC Paper 06|擴散程式語言與可驗證編譯器》,v0.1。
SMPDC_Paper06_Diffusive_Programming_Language_and_Verifiable_Compiler_v0.1_2026-10-16.md。 - [I7] Neo.K × Aletheia(2026-10-16)。《SMPDC Paper 07|作業系統與 AI 動態控制架構》,v0.1。
SMPDC_Paper07_OS_and_AI_Dynamic_Control_Plane_v0.1_2026-10-16.md。
公開技術與學術前例(查核截至 2026-10-17)
- [E1] NVIDIA, CUDA C++ Best Practices Guide, Chapters: Timing; Data Transfer; Pinned Memory; Asynchronous and Overlapping Transfers. https://docs.nvidia.com/cuda/cuda-c-best-practices-guide/index.html。用途:CPU/GPU timer、資料搬移及可重疊傳輸的性能與正確性前提。
- [E2] NVIDIA, CUDA Programming Guide, Unified and System Memory. https://docs.nvidia.com/cuda/cuda-programming-guide/02-basics/understanding-memory.html。用途:unified memory 模式、平台及硬體能力區分。
- [E3] NVIDIA, CUDA Programming Guide, Unified Memory. https://docs.nvidia.com/cuda/cuda-programming-guide/04-special-topics/unified-memory.html。用途:Windows/WSL 的 managed memory/coherence 限制及設備支援。
- [E4] NVIDIA, CUDA C++ Programming Guide, archived v13.0.1, stream ordering, events and explicit synchronization. https://docs.nvidia.com/cuda/archive/13.0.1/cuda-c-programming-guide/index.html。用途:CUDA API/stream/event 的語義性原始文獻;實機仍以安裝版本的文件為準。
- [E5] NVIDIA, CUDA Programming Guide, CUDA Graphs. https://docs.nvidia.com/cuda/cuda-programming-guide/04-special-topics/cuda-graphs.html。用途:圖式工作提交、捕捉、重用與更新的成熟基線。
- [E6] NVIDIA, CUDA Runtime API Reference Manual, Stream Ordered Memory Allocator, v13.4.2. https://docs.nvidia.com/cuda/archive/13.4.2/cuda-runtime-api/cuda_runtime_api/group__CUDART__MEMORY__POOLS.html。用途:
cudaMallocAsync/cudaFreeAsync及跨 stream 生命週期義務。 - [E7] NVIDIA, Nsight Systems User Guide. https://docs.nvidia.com/nsight-systems/UserGuide/index.html。用途:CUDA API、H2D/D2H copy、kernel/stream timeline 與儀器化風險。
- [E8] Bauer, M. et al., Legion: Expressing Locality and Independence with Logical Regions, SC 2012. https://legion.stanford.edu/publications/index.html。用途:異質 task/data locality/mapping 的既有研究。
- [E9] Kokkos, deep_copy API and Execution Space Semantics. https://kokkos.org/kokkos-core-wiki/API/core/view/deep_copy.html。用途:memory-space copy、同步/潛在非同步的成熟抽象。
- [E10] NVIDIA, NCCL Collective Operations. https://docs.nvidia.com/deeplearning/nccl/user-guide/docs/usage/collectives.html。用途:成熟多 GPU 集體運算前例;非 SMPDC 原創 primitive。
- [E11] Hoffman, H. and Oh, F., Constant Time Launch for Straight-Line CUDA Graphs and Other Performance Enhancements, NVIDIA Technical Blog, 2024-09-11. https://developer.nvidia.com/blog/constant-time-launch-for-straight-line-cuda-graphs-and-other-performance-enhancements。用途:圖式重複 launch 與 CPU submit overhead 的實測先例;不得直接移植其 benchmark 數值。
- [E12] NVIDIA, Kernel Fusion in NVIDIA CUDA: Optimizing Memory Traffic and Launch Overhead, technical blog, 2026. https://developer.nvidia.com/blog/kernel-fusion-in-nvidia-cuda-optimizing-memory-traffic-and-launch-overhead/。用途:fusion 與圖式提交不同層次,作為正當效能基線。
原稿結案標記(Paper 08 v0.1)
- 資料性質: 候選研究原稿;不是效能驗證報告。
- 核心交付:
GCC08,7 組具型別資料/狀態契約、8 項條件性命題、32 條安全不變量、36 項故障反例、B0–B8/E0–E4 與 W0–W10 實驗矩陣。 - 可實驗但尚未聲稱完成: Windows/Linux + NVIDIA RTX 3070 的 CPU/GPU 原型、CUDA streams/events/Graphs、OS adapter、AI 工作負載及 V1 perf/energy 測試。
- 下一篇: SMPDC Paper 09|記憶體與微觀硬體映射:快取、CCD/NUMA、實體互連、資料局部性與成本邊界。