109 年 國立成功大學人工智慧科技碩士學位學程《計算機組織與系統》

核心觀念

本題考查兩件事：

1. CPU 與 GPU 的硬體組織方式
2. 指令如何經過多級管線重疊執行

CPU 主要追求單一執行緒的低延遲與處理複雜控制流程；GPU 主要追求大量資料平行工作的高吞吐量。

---

解題方法

作答時依序說明：

1. CPU 的核心架構與指令管線。
2. GPU 的運算單元與執行緒階層。
3. GPU 的運算管線。
4. 兩者在平行化、記憶體、分支處理與適用工作負載上的差異。

---

一、CPU 架構與管線

1. CPU 架構

典型 CPU 可表示如下：

CPU
→ 多個複雜核心
→ 每個核心包含：

• 程式計數器（PC）與控制單元
• 指令快取（Instruction Cache）
• 指令解碼器
• 暫存器檔案（Register File）
• 整數 ALU、浮點單元、分支單元
• Load/Store 單元
• L1 指令快取與資料快取
• L2 快取

多核心 CPU 通常再共用較大型的 L3 快取與主記憶體控制器。

CPU 的特色是以較少但較複雜的核心，利用下列機制提高指令平行度：

• Superscalar：每個週期發射多條指令。
• Out-of-Order Execution：在不違反相依性的前提下，讓可執行指令先執行。
• Branch Prediction：預測分支方向，提前抓取指令。
• Speculative Execution：依據預測結果先行執行。
• Register Renaming：消除 WAR 與 WAW 假相依。
• Cache Hierarchy：利用多層快取降低記憶體存取延遲。

2. CPU 經典五級管線

教科書常用的 CPU 五級管線為：

$$\text{IF} \rightarrow \text{ID} \rightarrow \text{EX} \rightarrow \text{MEM} \rightarrow \text{WB}$$

階段	功能
IF（Instruction Fetch）	依據 PC 從指令快取抓取指令
ID（Instruction Decode）	解碼指令、讀取暫存器、判斷控制訊號
EX（Execute）	執行算術邏輯運算、計算位址或判斷分支
MEM（Memory Access）	執行 Load 或 Store 的資料快取存取
WB（Write Back）	將運算結果寫回暫存器

五級管線的重點是：不同指令可以同時處於不同階段。例如第一條指令在 EX 階段時，第二條指令可在 ID 階段，第三條指令可在 IF 階段。

若管線有 $k$ 級、每級耗時 $t$，連續執行 $n$ 條指令：

• 未管線化時間：

$$T_{\text{non-pipeline}}=nkt$$

• 理想管線化時間：

$$T_{\text{pipeline}}=(k+n-1)t$$

• 理想加速比：

$$S=\frac{nkt}{(k+n-1)t} =\frac{nk}{k+n-1}$$

當 $n$ 很大時，理想加速比接近 $k$。管線化主要提升吞吐量，不代表單一指令的延遲一定縮短。

現代 CPU 的管線通常比五級更深，實際階段可表示為：

$$\text{Fetch} \rightarrow \text{Decode} \rightarrow \text{Rename} \rightarrow \text{Dispatch} \rightarrow \text{Issue} \rightarrow \text{Execute} \rightarrow \text{Memory} \rightarrow \text{Writeback} \rightarrow \text{Retire}$$

現代 CPU 即使採用亂序執行，通常仍需依程式順序完成提交（Retire），以維持正確的架構狀態。

3. CPU 管線危障

CPU 管線必須處理三類危障：

• 結構危障（Structural Hazard）：多條指令同時爭用同一硬體資源。
• 資料危障（Data Hazard）：後續指令需要前一條指令尚未產生的結果。RAW 可透過 forwarding 或停滯處理；WAR、WAW 可透過暫存器重新命名處理。
• 控制危障（Control Hazard）：分支結果尚未確定，造成下一條指令位置不明。CPU 透過分支預測與錯誤路徑清除降低損失。

---

二、GPU 架構與管線

1. GPU 架構

典型 GPU 可表示如下：

GPU
→ 多個圖形或運算叢集
→ 多個 SM/CU
→ 每個 SM/CU 包含：

• Warp Scheduler 或 Wavefront Scheduler
• 大量 SIMD/SIMT 運算通道
• 暫存器檔案
• Shared Memory
• L1 Cache
• Load/Store 單元
• 浮點、整數、特殊函數或 Tensor 運算單元

GPU 外部通常還有：

• L2 Cache
• 高頻寬全域記憶體（Global Memory）
• 記憶體控制器
• 與 CPU 連接的 PCIe 或其他高速互連介面

GPU 採用執行緒階層：

$$\text{Grid} \rightarrow \text{Block} \rightarrow \text{Warp} \rightarrow \text{Thread}$$

一個 Block 通常配置到同一個 SM/CU 執行，其執行緒可以共享 Shared Memory 並進行同步。Warp 是硬體排程與發射的基本單位；以常見 NVIDIA GPU 為例，一個 Warp 包含 32 個執行緒。

GPU 採用 SIMT（Single Instruction, Multiple Threads）。同一個 Warp 的執行緒通常執行相同指令，但每個執行緒持有不同資料。

核心觀念

本題考查兩種平行計算的分類方式：

1. 依「使用何種計算資源」分類：heterogeneous computing。
2. 依「如何分割與執行資料」分類：data parallel computing。

兩者並不互斥，可同時出現在同一個系統中。

解題方法

回答定義題時，應包含「定義、運作方式、典型硬體、實例」四個部分，並補充與相近概念的差異。

(a) Heterogeneous computing

異質計算是指在同一個計算系統中，結合不同種類、不同架構或不同專長的處理單元，依照工作特性將任務分派給最適合的處理單元執行。

常見的處理單元包括：

• CPU：擅長複雜控制流程、分支多、需要較強單執行緒效能的工作。
• GPU：擅長大量獨立資料的相同運算。
• FPGA：可依需求建立專用硬體資料路徑。
• DSP：擅長訊號處理與固定格式的數值運算。

例如，在 CPU＋GPU 系統中，CPU 負責作業系統、程式控制與任務管理，GPU 負責大量矩陣乘法或影像像素運算。此時可將總工作量 $W$ 分配給不同處理器：

$$W = W_{\mathrm{CPU}} + W_{\mathrm{GPU}}$$

實際執行時間除了計算時間，還須考慮資料傳輸與同步成本：

$$T_{\mathrm{total}} \approx \max(T_{\mathrm{CPU}},T_{\mathrm{GPU}}) +T_{\mathrm{transfer}} +T_{\mathrm{sync}}$$

異質計算的主要目標是提高效能、降低能源消耗，或同時兼顧兩者。其重點不只是「有多個處理器」，而是這些處理器具有不同的架構與能力。

相對地，若系統由許多相同架構的 CPU 核心組成，通常稱為 homogeneous computing，即同質計算。

(b) Data parallel computing

資料平行計算是將大量資料分割成多個彼此獨立的部分，讓多個處理單元同時對不同資料執行相同或相似的運算。

例如向量加法：

核心觀念

題目中的 $A$、$B$ 尺寸分別為 $M\times N$ 與 $N\times P$，因此可以相乘，結果矩陣 $C$ 的尺寸為 $M\times P$。

矩陣乘法的定義為：

$$C_{i,j}=\sum_{k=0}^{N-1}A_{i,k}B_{k,j}$$

其中：

$$0\le i<M,\qquad 0\le j<P$$

每一個 $C_{i,j}$ 都是 $A$ 的第 $i$ 列與 $B$ 的第 $j$ 行內積，包含 $N$ 次乘法與加法。題目主要考查：

• 矩陣乘法的索引與維度。
• CPU 與 GPU 的主機—裝置記憶體資料傳輸。
• GPU 的 thread、block、grid 配置。
• 將可平行化的計算工作分配給 GPU 執行。

解題方法

採用 CUDA C，讓 GPU 的每一個執行緒負責計算一個 $C_{i,j}$ 元素：

• threadIdx.x 對應矩陣欄索引 $j$。
• threadIdx.y 對應矩陣列索引 $i$。
• blockIdx.x 負責切分輸出矩陣的欄。
• blockIdx.y 負責切分輸出矩陣的列。

為了減少重複讀取 GPU 全域記憶體的次數，以下程式使用 tiled matrix multiplication。每個 thread block 先將 $A$ 與 $B$ 的子矩陣載入共享記憶體，再由同一個 block 內的執行緒重複使用這些資料。

CUDA C 程式

以下假設矩陣採用 row-major 一維陣列儲存：

• $A_{i,k}$ 儲存於 A[i * N + k]。
• $B_{k,j}$ 儲存於 B[k * P + j]。
• $C_{i,j}$ 儲存於 C[i * P + j]。

核心觀念

本題考查讀者－寫者問題。多個讀者可同時讀取共享資料；寫者寫入時，必須獨占資料，不能有其他讀者或寫者同時進入。

以下依題意將程式中的 wmnt 視為 wmut；若照字面執行，wmnt 未宣告，程式無法正確運作。wmut 控制共享資料的存取權，rmut 保護讀者計數器 rcount。當第一位讀者進入時，會取得 wmut；最後一位讀者離開時，才釋放 wmut。

解題方法與各小題

(1) 是否為最佳解法？

不是。 這是偏袒讀者的讀者－寫者解法。它能讓多位讀者並行讀取，也能避免寫者與讀者或另一位寫者同時存取共享資料，因此可維持基本互斥要求。

但若讀者持續到來，rcount 可能一直不歸零，寫者便可能長時間無法取得 wmut，甚至發生寫者飢餓。所以此解法不保證公平，也不能稱為所有需求下的最佳解法；若要避免飢餓，需採用公平排隊或讀寫鎖等策略。

(2) 三個執行緒進入臨界區的順序

題目給定的到達順序是 $R1、W1、R2$：

1. $R1$ 將 rcount 從 $0$ 加到 $1$，因此取得 wmut，接著進入讀取臨界區。
2. $W1$ 嘗試取得 wmut，但此時它由讀者持有，所以 $W1$ 無法進入。
3. $R2$ 將 rcount 從 $1$ 加到 $2$。因為不是第一位讀者，它不需再次取得 wmut，因此可進入讀取臨界區。
4. $R1$ 離開時，計數降為 $1$，不釋放 wmut。$R2$ 離開後，計數降為 $0$，才釋放 wmut，讓 $W1$ 進入寫入臨界區。

因此進入順序為：

$$R1 \rightarrow R2 \rightarrow W1$$

$R1$ 與 $R2$ 可同時讀取；$W1$ 必須等兩位讀者都離開後才能寫入。

核心觀念

本題考查作業系統（Operating System, OS）與檔案系統（File System）在記憶體快取（Cache / Buffer）、虛擬記憶體管理（Virtual Memory Management）與檔案 I/O 機制中的架構設計。主要包含四大核心觀念：

1. 檔案 I/O 快取層級（Cache Hierarchies for File I/O）：
   當資料從硬碟讀取至使用者空間應用程式時，除了 CPU 內部快取（L1/L2/L3）外，還會經過使用者空間緩衝區（User-Space Buffer）、作業系統核心頁快取（OS Page Cache / Buffer Cache）以及磁碟控制器快取（Disk Controller Cache）。若要避開 OS 快取的緩衝，可使用 Direct I/O 機制。
2. 統一頁快取架構（Unified Page Cache Architecture）：
   現代作業系統將系統的檔案快取（Page Cache）與虛擬記憶體管理（Virtual Memory Management）相結合，使用統一的 Page Frame 管理處理器虛擬分頁與檔案對應分頁（File-Backed Pages）。若應用程式透過傳統 read() 系統呼叫將檔案讀入 Heap，會在記憶體中形成兩份重複副本。
3. 雙重快取現象（Double Caching Phenomenon）：
   資料同時存在於 Kernel Page Cache 與 User-Space Heap 中即為「雙重快取」，會加倍消耗實體記憶體容量，進而引發記憶體換頁抖動（Memory Thrashing）。
4. 零複製與記憶體映射技術（Memory-Mapped I/O & Zero-Copy）：
   為消除雙重快取，可採用記憶體映射 mmap() 或直接 I/O（O_DIRECT）。mmap() 透過頁表映射將核心快取頁面直接掛載至進程的虛擬位址空間，實現無須重複複製的檔案存取。

---

解題方法與分項詳細推導

(1) 讀取檔案時使用的快取種類與不緩衝寫入的情境

• 讀取檔案時使用的快取（Caches used during file read）：
  資料從磁碟讀取至使用者應用程式（User-space Application）過程中，除處理器硬體快取外，涉及以下三個階層的快取：

  1. 使用者空間應用程式緩衝區（User-Space Application Buffer）：例如 C 語言標準函式庫 stdio 的 FILE* 內建緩衝區、Java 之 BufferedReader，或應用程式自行維護的高階快取。
  2. 作業系統頁快取／緩衝區快取（OS Page Cache / Buffer Cache）：位於核心空間（Kernel Space）的主記憶體（RAM）中，將磁碟資料區塊（Blocks）以頁面（Pages）形式緩衝於實體記憶體，避免頻繁發起物理磁碟存取。
  3. 磁碟控制器快取（Disk Controller Cache / Drive Buffer）：位於 SSD/HDD 控制板上的硬體 RAM 緩衝區，用於加速控制器的物理讀寫與資料緩衝。

• 檔案寫入不安置於 OS 軟體快取（Bypassing Software Caches）的情境：

  1. 使用直接 I/O（Direct I/O）：應用程式在透過 open() 系統呼叫開啟檔案時，指定 O_DIRECT 旗標。此時寫入操作會直接繞過作業系統的核心 Page Cache，由 DMA 將使用者空間的資料直接寫入磁碟控制器／儲存裝置。
  2. 原始磁區直接存取（Raw Disk I/O / Raw Partition Access）：應用程式（如資料庫管理系統 DBMS）繞過檔案系統，直接對未格式化的硬碟原始區段（Raw Partition）進行讀寫。

---

(2) 統一虛擬記憶體快取技術與讀取大檔案儲存於 Heap 的影響

• 虛擬記憶體技術名稱：
  統一頁快取（Unified Page Cache） 或 統一緩衝快取（Unified Buffer Cache）。
  早期 OS 將 Buffer Cache（針對檔案系統 Block）與 Page Cache（針對虛擬記憶體 Page）分開管理；現代 OS 將兩者整合，統一由虛擬記憶體管理系統以頁分頁架構（Page Frame）同時託管進程虛擬記憶體與檔案快取內容。

• Processes 讀取大檔案並存於 Heap 記憶體時會發生的狀況：

  1. 產生「雙重快取（Double Caching）」：
     檔案資料經由 read() 讀取時，OS 會先將資料從磁碟載入核心空間的 Page Cache 中；隨後將該資料從 Page Cache 複製一份到進程於使用者空間的 Heap Memory 中。同份資料在實體記憶體（RAM）內佔據兩倍空間。
  2. 實體記憶體耗盡與系統抖動（Memory Thrashing）：
     若讀取的檔案極大，雙重快取會使實體記憶體需求加倍。當 RAM 不足時，作業系統被迫頻繁發動頁面淘汰（Page Eviction）或將進程分頁寫入 Swap 磁碟區（Swapping），造成極高 I/O 開銷與系統效能劇烈下降（Thrashing）。

---

(3) 資料在不同快取中被緩衝兩次的現象名稱

• 現象名稱：
  雙重快取（Double Caching） 或 雙重緩衝（Double Buffering）。

---

(4) 避免雙重快取的技術與 I/O 方塊圖

• 避免雙重快取的技術（Technique to avoid double caching）：

  1. 記憶體映射 I/O（Memory-Mapped I/O，使用 mmap 系統呼叫）：
     作業系統建立頁表項目（Page Table Entries），將檔案的 Kernel Page Cache 直接映射至進程的虛擬位址空間（Virtual Address Space）。進程透過指標讀寫該記憶體區段時，即直接操作 Kernel Page Cache，無須發起 read() 將資料複製到 User Heap，完全消除雙重快取。
  2. 直接 I/O（Direct I/O，使用 O_DIRECT 旗標）：
     對於自身具備獨立 Cache 管理機制的應用程式（如資料庫引擎），透過 O_DIRECT 完全跳過 Kernel Page Cache，資料直接於磁碟與 User Buffer 之間傳輸。

• 使用 Memory-Mapped I/O (mmap) 的檔案 I/O 方塊圖（Block Diagram）：