垃圾CPU，耗我時光——Jetson Nano 初體驗2

CPU與GPU性能測試

1. CPU性能測試：計算圓周率

bc 命令是任意精度計算器語言，一般在 linux 下當計算器用。它相似基本的計算器, 使用這個計算器能夠作基本的數學運算
man 一下 bc 便可知道，a 是 bc 的一個內置函數，表明反正切 arctan ，因爲 tan(pi/4) = 1 ，因而 4*arctan(1) = pi

計算圓周率的前一萬位（單線程）並與 Intel(R) Xeon(R) Platinum 8163 CPU 的CPU作對比

# jetson nano CPU 參數 lscpu Architecture: aarch64 Byte Order: Little Endian CPU(s): 4 On-line CPU(s) list: 0-3 Thread(s) per core: 1 Core(s) per socket: 4 Socket(s): 1 Vendor ID: ARM Model: 1 Model name: Cortex-A57 Stepping: r1p1 CPU max MHz: 1428.0000 CPU min MHz: 102.0000 BogoMIPS: 38.40 L1d cache: 32K L1i cache: 48K L2 cache: 2048K Flags: fp asimd evtstrm aes pmull sha1 sha2 crc32 # 計算圓周率的前一萬位（單線程） time echo "scale = 10000; 4*a(1)" | bc -l -q 3.1415926535897... real 5m22.161s user 5m21.496s sys 0m0.020s # Intel(R) Xeon(R) Platinum 8163 CPU 參數 lscpu Architecture: x86_64 CPU op-mode(s): 32-bit, 64-bit Byte Order: Little Endian CPU(s): 1 On-line CPU(s) list: 0 Thread(s) per core: 1 Core(s) per socket: 1 Socket(s): 1 NUMA node(s): 1 Vendor ID: GenuineIntel CPU family: 6 Model: 85 Model name: Intel(R) Xeon(R) Platinum 8163 CPU @ 2.50GHz Stepping: 4 CPU MHz: 2500.008 BogoMIPS: 5000.01 Hypervisor vendor: KVM Virtualization type: full L1d cache: 32K L1i cache: 32K L2 cache: 1024K L3 cache: 33792K NUMA node0 CPU(s): 0 Flags: fpu vme de pse tsc msr pae mce cx8 apic sep mtrr pge mca cmov pat pse36 clflush mmx fxsr sse sse2 ss ht syscall nx pdpe1gb rdtscp lm constant_tsc rep_good nopl cpuid pni pclmulqdq ssse3 fma cx16 pcid sse4_1 sse4_2 x2apic movbe popcnt tsc_deadline_timer aes xsave avx f16c rdrand hypervisor lahf_lm abm 3dnowprefetch invpcid_single pti ibrs ibpb stibp fsgsbase tsc_adjust bmi1 hle avx2 smep bmi2 erms invpcid rtm mpx avx512f avx512dq rdseed adx smap avx512cd avx512bw avx512vl xsaveopt xsavec xgetbv1 # 計算圓周率的前一萬位（單線程） time echo "scale = 10000; 4*a(1)" | bc -l -q 3.1415926535897... real 2m20.695s user 2m19.211s sys 0m0.047s 

單核 CPU 性能大概是 Intel(R) Xeon(R) Platinum 8163 的一半

2. CPU與GPU對比測試

• CUDA在線學習

2.1 四種計算機模型

GPU設計的初衷就是爲了減輕CPU計算的負載，將一部分圖形計算的功能設計到一塊獨立的處理器中，將矩陣變換、頂點計算和光照計算等操做從 CPU 中轉移到 GPU中，從而一方面加速圖形處理，另外一方面減少了 CPU 的工做負載，讓 CPU 有時間去處理其它的事情。
在GPU上的各個處理器採起異步並行的方式對數據流進行處理，根據費林分類法（Flynn's Taxonomy），能夠將信息流（information stream）分紅指令（Instruction）和數據（Data）兩種，據此又可分紅四種計算機類型：

• 單一指令流單一數據流計算機（SISD）：單核CPU
• 單一指令流多數據流計算機（SIMD）：GPU的計算模型
• 多指令流單一數據流計算機（MISD）：流水線模型
• 多指令流多數據流計算機（MIMD）：多核CPU

2.2 CPU 與 GPU 結構差別

 

CPU 與 GPU 結構差別

(1)CPU設計理念：低延時

 

CPU設計理念：低延時

• ALU：CPU有強大的ALU（算術運算單元）,它能夠在不多的時鐘週期內完成算術計算。
  • 當今的CPU能夠達到64bit 雙精度。執行雙精度浮點源算的加法和乘法只須要1～3個時鐘週期。
  • CPU的時鐘週期的頻率是很是高的，達到1.532～4gigahertz(千兆HZ, 10的9次方).
• Cache：大的緩存也能夠下降延時。保存不少的數據放在緩存裏面，當須要訪問的這些數據，只要在以前訪問過的，現在直接在緩存裏面取便可。
• Control：複雜的邏輯控制單元。
  • 當程序含有多個分支的時候，它經過提供分支預測的能力來下降延時。
  • 數據轉發。 當一些指令依賴前面的指令結果時，數據轉發的邏輯控制單元決定這些指令在pipeline中的位置而且儘量快的轉發一個指令的結果給後續的指令。這些動做須要不少的對比電路單元和轉發電路單元。

(2)GPU設計理念：大吞吐量

 

GPU設計理念：大吞吐量

• ALU，Cache：GPU的特色是有不少的ALU和不多的cache. 緩存的目的不是保存後面須要訪問的數據的，這點和CPU不一樣，而是爲thread提升服務的。若是有不少線程須要訪問同一個相同的數據，緩存會合並這些訪問，而後再去訪問dram（由於須要訪問的數據保存在dram中而不是cache裏面），獲取數據後cache會轉發這個數據給對應的線程，這個時候是數據轉發的角色。可是因爲須要訪問dram，天然會帶來延時的問題。
• Control：控制單元（左邊黃色區域塊）能夠把多個的訪問合併成少的訪問。

GPU的雖然有dram延時，卻有很是多的ALU和很是多的thread. 爲了平衡內存延時的問題，咱們能夠中充分利用多的ALU的特性達到一個很是大的吞吐量的效果。儘量多的分配多的Threads.一般來看GPU ALU會有很是重的pipeline就是由於這樣。

2.3 Nvidia GPU架構

(1)硬件架構

• SP：最基本的處理單元，streaming processor，也稱爲CUDA core。最後具體的指令和任務都是在SP上處理的。GPU進行並行計算，也就是不少個SP同時作處理。
• SM：多個SP加上其餘的一些資源組成一個streaming multiprocessor。也叫GPU大核，其餘資源如：warp scheduler，register，shared memory等。SM能夠看作GPU的心臟（對比CPU核心），register和shared memory是SM的稀缺資源。CUDA將這些資源分配給全部駐留在SM中的threads。所以，這些有限的資源就使每一個SM中active warps有很是嚴格的限制，也就限制了並行能力。

 

Nvidia GPU硬件架構

(2)軟件架構

CUDA在軟件方面組成有：一個CUDA庫、一個應用程序編程接口（API）及其運行庫(Runtime)、兩個較高級別的通用數學庫，即CUFFT和CUBLAS。CUDA改進了DRAM的讀寫靈活性，使得GPU與CPU的機制相吻合。另外一方面，CUDA 提供了片上（on-chip）共享內存，使得線程之間能夠共享數據。應用程序能夠利用共享內存來減小DRAM的數據傳送，更少的依賴DRAM的內存帶寬。

• thread：一個CUDA的並行程序會被以許多個threads來執行。
• block：數個threads會被羣組成一個block，同一個block中的threads能夠同步，也能夠經過shared memory通訊。
• grid：多個blocks則會再構成grid。
• warp：GPU執行程序時的調度單位，目前cuda的warp的大小爲32，同在一個warp的線程，以不一樣數據資源執行相同的指令,這就是所謂 SIMT。

 

Nvidia GPU軟件架構

（3）軟硬件架構對應關係
從軟件上看，SM更像一個獨立的CPU core。SM（Streaming Multiprocessors）是GPU架構中很是重要的部分，GPU硬件的並行性就是由SM決定的。

當一個kernel啓動後，thread會被分配到這些SM中執行。大量的thread可能會被分配到不一樣的SM，同一個block中的threads必然在同一個SM中並行（SIMT）執行。每一個thread擁有它本身的程序計數器和狀態寄存器，而且用該線程本身的數據執行指令，這就是所謂的Single Instruction Multiple Thread。

CUDA是一種典型的SIMT架構（單指令多線程架構），SIMT和SIMD（Single Instruction, Multiple Data）相似，SIMT應該算是SIMD的升級版，更靈活，但效率略低，SIMT是NVIDIA提出的GPU新概念。兩者都經過將一樣的指令廣播給多個執行官單元來實現並行。一個主要的不一樣就是，SIMD要求全部的vector element在一個統一的同步組裏同步的執行，而SIMT容許線程們在一個warp中獨立的執行。

2.4 CUDA C編程入門

（1）程序架構

CUDA程序構架分爲兩部分：Host和Device。通常而言，Host指的是CPU，Device指的是GPU。在CUDA程序構架中，主程序仍是由 CPU 來執行，而當遇到數據並行處理的部分，CUDA 就會將程序編譯成 GPU 能執行的程序，並傳送到GPU。而這個程序在CUDA裏稱作核（kernel）。CUDA容許程序員定義稱爲核的C語言函數，從而擴展了 C 語言，在調用此類函數時，它將由N個不一樣的CUDA線程並行執行N次，這與普通的C語言函數只執行一次的方式不一樣。執行核的每一個線程都會被分配一個獨特的線程ID，可經過內置的threadIdx變量在內核中訪問此ID。
在 CUDA 程序中，主程序在調用任何 GPU 內核以前，必須對核進行執行配置，即肯定線程塊數和每一個線程塊中的線程數以及共享內存大小。

CUDA 設備擁有多個獨立的存儲空間，其中包括：全局存儲器、本地存儲器、共享存儲器、常量存儲器、紋理存儲器和寄存器

CUDA線程可在執行過程當中訪問多個存儲器空間的數據，以下圖所示其中：

• 每一個線程都有一個私有的本地存儲器。
• 每一個線程塊都有一個共享存儲器，該存儲器對於塊內的全部線程都是可見的，而且與塊具備相同的生命週期。
• 全部線程均可訪問相同的全局存儲器。
• 此外還有兩個只讀的存儲器空間，可由全部線程訪問，這兩個空間是常量存儲器空間和紋理存儲器空間。全局、固定和紋理存儲器空間通過優化，適於不一樣的存儲器用途。紋理存儲器也爲某些特殊的數據格式提供了不一樣的尋址模式以及數據過濾，方便 Host對流數據的快速存取。

CUDA 假設線程可在物理上獨立的設備上執行，此類設備做爲運行C語言程序的主機的協處理器操做。內核在GPU上執行，而C語言程序的其餘部分在CPU上執行（即串行代碼在主機上執行，而並行代碼在設備上執行）。此外，CUDA還假設主機和設備均維護本身的DRAM，分別稱爲主機存儲器和設備存儲器。於是，一個程序經過調用CUDA運行庫來管理對內核可見的全局、固定和紋理存儲器空間。這種管理包括設備存儲器的分配和取消分配，還包括主機和設備存儲器之間的數據傳輸。

（2）CUDA C基礎

CUDA C是對C/C++語言進行拓展後造成的變種，兼容C/C++語法，文件類型爲".cu"文件，編譯器爲"nvcc"，相比傳統的C/C++，主要添加了如下幾個方面：

• 函數類型限定符：用來肯定某個函數是在CPU仍是GPU上運行，以及這個函數是從CPU調用仍是從GPU調用
  • device表示從GPU調用，在GPU上執行
  • global表示從CPU調用，在GPU上執行，也稱之爲kernel函數
  • host表示在CPU上調用，在CPU上執行
• 執行配置運算符：執行配置運算符<<<>>>，用來傳遞內核函數的執行參數。格式以下：
  kernel<<<gridDim, blockDim, memSize, stream>>>(para1, para2, ...);
  • gridDim表示網格的大小，能夠是1,2,3維
  • blockDim表示塊的·大小，能夠是1,2,3維
  • memSize表示動態分配的共享存儲器大小，默認爲0
  • stream表示執行的流，默認爲0
  • para1, para2等爲核函數參數
• 五個內置變量：這些內置變量用來在運行時得到Grid和Block的尺寸及線程索引等信息
  • gridDim: 包含三個元素x, y, z的結構體，表示Grid在三個方向上的尺寸，對應於執行配置中的第一個參數
  • blockDim: 包含上元素x, y, z的結構體，表示Block在三個方向上的尺寸，對應於執行配置中的第二個參數
  • blockIdx: 包含三個元素x, y, z的結構體，分別表示當前線程所在塊在網格中x, y, z方向上的索引
  • threadIdx: 包含三個元素x, y, z的結構體，分別表示當前線程在其所在塊中x, y, z方向上的索引
  • warpSize: 代表warp的尺寸
• 變量類型限定符：用來肯定某個變量在設備上的內存位置
  • device表示位於全局內存空間，默認類型
  • share表示位於共享內存空間
  • constant表示位於常量內存空間
  • texture表示其綁定的變量能夠被紋理緩存加速訪問
• 其餘的還有數學函數、原子函數、紋理讀取、綁定函數等

2.5 CPU與GPU的矩陣乘法對比

（1）CPU單線程矩陣乘法

// CPU單線程矩陣乘法 #include <stdio.h> #include <stdlib.h> #include <sys/time.h> #include <unistd.h> #define w 2000 struct Matrix { int width; int height; float *elements; }; void matMul(float * M, float * N, float * P, int width){ for (int i = 0; i < width; i++){ for (int j = 0; j < width; j++){ float sum = 0; for (int k = 0; k < width; k++){ float a = M[i * width + k]; float b = N[k * width + j]; sum += a * b; } P[i * width + j] = sum; } } } int main(){ int width = w; int height = w; float * m = (float *)malloc (width * height * sizeof (float)); float * n = (float *)malloc (width * height * sizeof (float)); float * p = (float *)malloc (width * height * sizeof (float)); for (int i = 0; i < width * height; i++){ m[i] = 9.9; n[i] = 2.5; } struct timeval t1,t2; gettimeofday(&t1,NULL); double timeuse; matMul(m, n, p, w); gettimeofday(&t2,NULL); timeuse = t2.tv_sec - t1.tv_sec + (t2.tv_usec - t1.tv_usec)/1000000.0; printf("Use Time:%f\n",timeuse); return 0; } 

而後編譯運行

gcc  cpu_sigle.c -O3 -o cpu_sigle

./cpu_sigle

Use Time:52.641901 

（2）CPU多線程矩陣乘法

//CPU多線程矩陣乘法 #include <stdio.h> #include <stdlib.h> #include <pthread.h> #include <sys/time.h> #include <string.h> #include <unistd.h> #include <sys/types.h> #include <sys/stat.h> #include <fcntl.h> #define LOG_ #define SIZE 8000 int * A, * B; // 計算矩陣 int * result, * result2, * result3, * result4; // 結果矩陣 /* int A[SIZE][SIZE]; int B[SIZE][SIZE]; int result[SIZE][SIZE]; int result2[SIZE][SIZE]; int result3[SIZE][SIZE]; int result4[SIZE][SIZE]; */ int size; // 矩陣階數 pthread_t tid2[2]; // 雙線程id pthread_t tid3[3]; // 三線程id pthread_t tid4[4]; // 四線程id /* 雙線程函數 */ void twoThread1(){ int i, j, k; for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ if (i % 2 == 0) result2[i * size + j] += A[i * size + k] * B[k * size + j]; // result2[i][j] += A[i][k] * B[k][j]; } } void twoThread2(){ int i, j, k; for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ if (i % 2 != 0) result2[i * size + j] += A[i * size + k] * B[k * size + j]; // result2[i][j] += A[i][k] * B[k][j]; } } /* 雙線程函數 end */ /* 三線程函數 */ void threeThread1(){ int i, j, k; for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ if (i % 3 == 0) result3[i * size + j] += A[i * size + k] * B[k * size + j]; // result3[i][j] += A[i][k] * B[k][j]; } } void threeThread2(){ int i, j, k; for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ if (i % 3 != 0 && i % 2 != 0) result3[i * size + j] += A[i * size + k] * B[k * size + j]; // result3[i][j] += A[i][k] * B[k][j]; } } void threeThread3(){ int i, j, k; for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ if (i % 3 != 0 && i % 2 == 0) result3[i * size + j] += A[i * size + k] * B[k * size + j]; // result3[i][j] += A[i][k] * B[k][j]; } } /* 三線程函數 end */ /* 四線程函數 */ void fourThread1(){ int i, j, k; for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ if (i % 2 == 0 && i % 4 != 0) result4[i * size + j] += A[i * size + k] * B[k * size + j]; // result4[i][j] += A[i][k] * B[k][j]; } } void fourThread2(){ int i, j, k; for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ if (i % 4 == 0) result4[i * size + j] += A[i * size + k] * B[k * size + j]; // result4[i][j] += A[i][k] * B[k][j]; } } void fourThread3(){ int i, j, k; for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ if (i % 2 != 0 && i % 3 == 0) result4[i * size + j] += A[i * size + k] * B[k * size + j]; // result4[i][j] += A[i][k] * B[k][j]; } } void fourThread4(){ int i, j, k; for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ if (i % 2 != 0 && i % 3 != 0) result4[i * size + j] += A[i * size + k] * B[k * size + j]; // result4[i][j] += A[i][k] * B[k][j]; } } /* 四線程函數 end */ int main(){ int i, j, k, m, n; // 循環變量 struct timeval t1, t2; double timeuse; // 計時 char sizeChars[8]; // 階數寫入字符串 char timeChars[16]; // 耗時寫入字符串 // 申請空間, 計算矩陣和結果矩陣 A = (int *)malloc (sizeof (int) * SIZE * SIZE); B = (int *)malloc (sizeof (int) * SIZE * SIZE); result = (int *)malloc (sizeof (int) * SIZE * SIZE); result2 = (int *)malloc (sizeof (int) * SIZE * SIZE); result3 = (int *)malloc (sizeof (int) * SIZE * SIZE); result4 = (int *)malloc (sizeof (int) * SIZE * SIZE); for (i = 0; i < SIZE; i++) for (j = 0; j < SIZE; j++){ /* A[i][j] = 1; B[i][j] = 2; result[i][j] = 0; result2[i][j] = 0; result3[i][j] = 0; result4[i][j] = 0; */ A[i * SIZE + j] = 1; B[i * SIZE + j] = 2; result[i * SIZE + j] = 0; result2[i * SIZE + j] = 0; result3[i * SIZE + j] = 0; result4[i * SIZE + j] = 0; } int fd; fd = open ("./pthreadTime.txt", O_WRONLY | O_CREAT, 0777); lseek(fd, 0, SEEK_SET); for (size = 200; size <= SIZE; size += 200){ printf ("當前階數: %d\n", size); sprintf (sizeChars, "%d, ", size); write (fd, sizeChars, strlen (sizeChars)); #ifdef LOG printf ("A矩陣: \n"); for (i = 0; i < size; i++){ for (j = 0; j < size; j++){ printf ("%d ", A[i * size + j]); // printf ("%d ", A[i][j]); } printf ("\n"); } printf ("B矩陣: \n"); for (i = 0; i < size; i++){ for (j = 0; j < size; j++){ printf ("%d ", B[i * size + j]); // printf ("%d ", B[i][j]); } printf ("\n"); } #endif /* 單線程 */ gettimeofday (&t1, NULL); for (i = 0; i < size; i++) for (j = 0; j < size; j++) for (k = 0; k < size; k++){ result[i * size + j] += A[i * size + k] * B[k * size +j]; // result[i][j] += A[i][k] * B[k][j]; } gettimeofday(&t2, NULL); timeuse = t2.tv_sec - t1.tv_sec + (t2.tv_usec - t1.tv_usec)/1000000.0; #ifdef LOG printf ("單線程結果矩陣: \n"); for (i = 0; i < size; i++){ for (j = 0; j < size; j++){ printf ("%d ", result[i * size + j]); // printf ("%d ", result[i][j]); } printf ("\n"); } #endif printf("單線程耗時: %fs\n", timeuse); sprintf (timeChars, "%lf, ", timeuse); write (fd, timeChars, strlen (timeChars)); for (i = 0; i < size; i++) for (j = 0; j < size; j++){ result[i * size + j] = 0; result2[i * size + j] = 0; result3[i * size + j] = 0; result4[i * size + j] = 0; /* result[i][j] = 0; result2[i][j] = 0; result3[i][j] = 0; result4[i][j] = 0; */ } /* 單線程 end */ /* 雙線程 */ gettimeofday (&t1, NULL); pthread_create (&tid2[0], NULL, (void *)twoThread1, NULL); pthread_join (tid2[0], NULL); pthread_create (&tid2[1], NULL, (void *)twoThread2, NULL); pthread_join (tid2[1], NULL); gettimeofday (&t2, NULL); timeuse = t2.tv_sec - t1.tv_sec + (t2.tv_usec - t1.tv_usec)/1000000.0; #ifdef LOG printf ("雙線程結果矩陣: \n"); for (i = 0; i < size; i++){ for (j = 0; j < size; j++){ printf ("%d ", result2[i * size + j]); // printf ("%d ", result2[i][j]); } printf ("\n"); } #endif printf("雙線程耗時: %fs\n", timeuse); sprintf (timeChars, "%lf, ", timeuse); write (fd, timeChars, strlen (timeChars)); for (i = 0; i < size; i++) for (j = 0; j < size; j++){ result[i * size + j] = 0; result2[i * size + j] = 0; result3[i * size + j] = 0; result4[i * size + j] = 0; /* result[i][j] = 0; result2[i][j] = 0; result3[i][j] = 0; result4[i][j] = 0; */ } /* 雙線程 end */ /* 三線程 */ gettimeofday (&t1, NULL); pthread_create (&tid3[0], NULL, (void *)threeThread1, NULL); pthread_join (tid3[0], NULL); pthread_create (&tid3[1], NULL, (void *)threeThread2, NULL); pthread_join (tid3[1], NULL); pthread_create (&tid3[2], NULL, (void *)threeThread3, NULL); pthread_join (tid3[2], NULL); gettimeofday (&t2, NULL); timeuse = t2.tv_sec - t1.tv_sec + (t2.tv_usec - t1.tv_usec)/1000000.0; #ifdef LOG printf ("三線程結果矩陣: \n"); for (i = 0; i < size; i++){ for (j = 0; j < size; j++){ printf ("%d ", result3[i * size + j]); } printf ("\n"); } #endif printf("三線程耗時: %fs\n", timeuse); sprintf (timeChars, "%lf, ", timeuse); write (fd, timeChars, strlen (timeChars)); for (i = 0; i < size; i++) for (j = 0; j < size; j++){ result[i * size + j] = 0; result2[i * size + j] = 0; result3[i * size + j] = 0; result4[i * size + j] = 0; /* result[i][j] = 0; result2[i][j] = 0; result3[i][j] = 0; result4[i][j] = 0; */ } /* 三線程 end */ /* 四線程 */ gettimeofday (&t1, NULL); pthread_create (&tid4[0], NULL, (void *)fourThread1, NULL); pthread_join (tid4[0], NULL); pthread_create (&tid4[1], NULL, (void *)fourThread2, NULL); pthread_join (tid4[1], NULL); pthread_create (&tid4[2], NULL, (void *)fourThread3, NULL); pthread_join (tid4[2], NULL); pthread_create (&tid4[3], NULL, (void *)fourThread4, NULL); pthread_join (tid4[3], NULL); gettimeofday (&t2, NULL); timeuse = t2.tv_sec - t1.tv_sec + (t2.tv_usec - t1.tv_usec)/1000000.0; #ifdef LOG printf ("四線程結果矩陣: \n"); for (i = 0; i < size; i++){ for (j = 0; j < size; j++){ printf ("%d ", result4[i * size + j]); } printf ("\n"); } #endif printf("四線程耗時: %fs\n", timeuse); sprintf (timeChars, "%lf\n", timeuse); write (fd, timeChars, strlen (timeChars)); for (i = 0; i < size; i++) for (j = 0; j < size; j++){ result[i * size + j] = 0; result2[i * size + j] = 0; result3[i * size + j] = 0; result4[i * size + j] = 0; /* result[i][j] = 0; result2[i][j] = 0; result3[i][j] = 0; result4[i][j] = 0; */ } /* 四線程 end */ } // 釋放空間 free (A); free (B); free (result); free (result2); free (result3); free (result4); A = NULL; B = NULL; result = NULL; result2 = NULL; result3 = NULL; result4 = NULL; // 關閉文件 close (fd); return 0; } 

編譯

gcc cpu_mul.c -O3 -o cpu_mul 

出現問題

undefined reference to `pthread_create' 

問題的緣由：pthread不是linux下的默認的庫，也就是在連接的時候，沒法找到phread庫中哥函數的入口地址，因而連接會失敗。

解決：在gcc編譯的時候，附加要加 -lpthread參數便可解決。

再編譯

gcc cpu_mul.c -O3 -lpthread -o cpu_mul 

執行

./cpu_mul
當前階數: 200
單線程耗時: 0.068800s
雙線程耗時: 0.064988s
三線程耗時: 0.065114s
四線程耗時: 0.065234s
當前階數: 400
單線程耗時: 0.616225s
雙線程耗時: 0.592269s
三線程耗時: 0.590107s
四線程耗時: 0.590583s
當前階數: 600
單線程耗時: 2.106992s
雙線程耗時: 2.067765s
三線程耗時: 2.078379s
四線程耗時: 2.079277s
當前階數: 800
單線程耗時: 5.433704s
雙線程耗時: 5.232933s
三線程耗時: 5.241303s
四線程耗時: 5.245513s
當前階數: 1000
單線程耗時: 11.862752s
雙線程耗時: 11.370975s
三線程耗時: 11.380489s
四線程耗時: 11.376795s
當前階數: 1200
單線程耗時: 21.028770s
雙線程耗時: 20.471958s
三線程耗時: 20.481492s
四線程耗時: 20.683800s
當前階數: 1400
單線程耗時: 33.121682s
雙線程耗時: 30.445559s
三線程耗時: 30.390110s
四線程耗時: 30.387628s
當前階數: 1600
單線程耗時: 57.306933s
雙線程耗時: 56.083187s
三線程耗時: 56.102355s
四線程耗時: 56.104007s
當前階數: 1800
單線程耗時: 70.554218s
雙線程耗時: 64.973096s
三線程耗時: 64.993218s
四線程耗時: 64.977718s
當前階數: 2000
單線程耗時: 98.518510s
雙線程耗時: 94.792153s
三線程耗時: 94.810620s
四線程耗時: 94.876091s

然而從htop裏面看的多線程時實際只有兩個線程，有一個線程在工做，另外一個在sleep，非常奇怪

編譯時去掉優化級別

gcc cpu_mul.c -lpthread -o cpu_mul2 

當前階數: 200
單線程耗時: 0.236366s 雙線程耗時: 0.276663s 三線程耗時: 0.352680s 四線程耗時: 0.365158s 當前階數: 400 單線程耗時: 2.031096s 雙線程耗時: 2.332579s 三線程耗時: 2.967604s 四線程耗時: 3.046181s 當前階數: 600 單線程耗時: 7.005228s 雙線程耗時: 7.990198s 三線程耗時: 10.129884s 四線程耗時: 10.384419s 當前階數: 800 單線程耗時: 16.948757s 雙線程耗時: 19.568093s 三線程耗時: 24.364115s 四線程耗時: 24.274340s 當前階數: 1000 單線程耗時: 38.138426s 雙線程耗時: 42.429860s 三線程耗時: 52.424720s 四線程耗時: 53.720896s 

從htop裏面看也是兩個線程，一個CPU佔用100%，另外一個CPU佔用0%，在圍觀......結果未優化的性能更差，一樣只有兩個線程，且一個線程在圍觀。。。真的是醉了。。。有機會再研究多線程的編程吧！
（3）GPU 器件參數輸出

vim device.cu nvcc device.cu -o device 

#include <stdio.h> int main() { int nDevices; cudaGetDeviceCount(&nDevices); for (int i = 0; i < nDevices; i++) { cudaDeviceProp prop; cudaGetDeviceProperties(&prop, i); printf("Device Num: %d\n", i); printf("Device name: %s\n", prop.name); printf("Device SM Num: %d\n", prop.multiProcessorCount); printf("Share Mem Per Block: %.2fKB\n", prop.sharedMemPerBlock / 1024.0); printf("Max Thread Per Block: %d\n", prop.maxThreadsPerBlock); printf("Memory Clock Rate (KHz): %d\n", prop.memoryClockRate); printf("Memory Bus Width (bits): %d\n", prop.memoryBusWidth); printf("Peak Memory Bandwidth (GB/s): %.2f\n\n", 2.0 * prop.memoryClockRate * (prop.memoryBusWidth / 8) / 1.0e6); } return 0; } 

結果：

./device
Device Num: 0
Device name: NVIDIA Tegra X1
Device SM Num: 1
Share Mem Per Block: 48.00KB
Max Thread Per Block: 1024
Memory Clock Rate (KHz): 12750
Memory Bus Width (bits): 64
Peak Memory Bandwidth (GB/s): 0.20

(4)GPU 矩陣乘法

//GPU 矩陣乘法 #include <stdio.h> #include <stdlib.h> #include <sys/time.h> #include <unistd.h> #define w 2000 struct Matrix { int width; int height; float *elements; }; __device__ float getElement(Matrix *A, int row, int col) { return A->elements[row * A->width + col]; } __device__ void setElement(Matrix *A, int row, int col, float value) { A->elements[row * A->width + col] = value; } __global__ void matMulKernel(Matrix *A, Matrix *B, Matrix *C) { float Cvalue = 0.0; int row = threadIdx.y + blockIdx.y * blockDim.y; int col = threadIdx.x + blockIdx.x * blockDim.x; for (int i = 0; i < A->width; ++i) { Cvalue += getElement(A, row, i) * getElement(B, i, col); } setElement(C, row, col, Cvalue); } int main() { int width = w; int height = w; Matrix *A, *B, *C; cudaMallocManaged((void**)&A, sizeof(Matrix)); cudaMallocManaged((void**)&B, sizeof(Matrix)); cudaMallocManaged((void**)&C, sizeof(Matrix)); int nBytes = width * height * sizeof(float); cudaMallocManaged((void**)&A->elements, nBytes); cudaMallocManaged((void**)&B->elements, nBytes); cudaMallocManaged((void**)&C->elements, nBytes); A->height = height; A->width = width; B->height = height; B->width = width; C->height = height; C->width = width; for (int i = 0; i < width * height; ++i) { A->elements[i] = 1.0; B->elements[i] = 2.0; } dim3 blockSize(32, 32); dim3 gridSize((width + blockSize.x - 1) / blockSize.x, (height + blockSize.y - 1) / blockSize.y); struct timeval t1,t2; gettimeofday(&t1,NULL); double timeuse; matMulKernel << < gridSize, blockSize >> >(A, B, C); cudaDeviceSynchronize(); gettimeofday(&t2,NULL); timeuse = t2.tv_sec - t1.tv_sec + (t2.tv_usec - t1.tv_usec)/1000000.0; printf("Use Time:%fs\n", timeuse); return 0; } 

vim gpu_mul.cu
nvcc gpu_mul.cu  -o gpu_mul

Use Time:1.466122s (w=2000) Use Time:267.060169s (w=8000) 

（5）結果對比

經過對矩陣乘法運算的結果對比，一樣2000階的矩陣乘法，單核CPU須要52.64s，而GPU只須要1.466s，矩陣乘法運算速度遠快於CPU，另外編程時也要根據其GPU編程模型來優化

參考資料

• NVIDIA Jetson TX2 查看系統參數狀態
• CUDA編程入門
• CUDA編程(四): CPU與GPU的矩陣乘法對比