# 2017q1 Homework1 (compute-pi)
contributed by <`ierosodin`>
### Reviewed by `0xff07`
Pros
1. 用 95 % 信賴區間消除偏差的資料
2. openCL 的實作
3. getopt()
Cons
1. openCL 的編譯可以寫清楚一點，或是加入 Makefile 中
2. thread_benchmark.c 中，測試不同 thread 數目的程式略嫌冗長，可以考慮用迴圈或寫個 shell script 自動測試不同執行緒數目。
3. error.c 同上
## 開發環境
作業系統 : Red Hat 7
`$ lscpu`
Architecture: x86_64
CPU op-mode(s): 32-bit, 64-bit
Byte Order: Little Endian
CPU(s): 12
On-line CPU(s) list: 0-11
Thread(s) per core: 2
Core(s) per socket: 6
Socket(s): 1
NUMA node(s): 1
Vendor ID: GenuineIntel
CPU family: 6
Model: 45
Model name: Genuine Intel(R) CPU @ 3.30GHz
Stepping: 5
CPU MHz: 1277.976
BogoMIPS: 6600.19
Virtualization: VT-x
L1d cache: 32K
L1i cache: 32K
L2 cache: 256K
L3 cache: 15360K
NUMA node0 CPU(s): 0-11
## 軟體安裝
```
$ git clone https://github.com/sysprog21/compute-pi
$ cd compute-pi
$ make check
$ make gencsv
```
## 重現實驗
`$ time ./time_test_baseline`
```
N = 400000000 , pi = 3.141593
real 0m6.979s
user 0m6.978s
sys 0m0.001s
```
`$ time ./time_test_openmp_2`
```
N = 400000000 , pi = 3.141593
real 0m3.568s
user 0m7.131s
sys 0m0.002s
```
`$ time ./time_test_openmp_4`
```
N = 400000000 , pi = 3.141593
real 0m1.886s
user 0m7.536s
sys 0m0.001s
```
`$ time ./time_test_avx`
```
N = 400000000 , pi = 3.141593
real 0m3.340s
user 0m3.331s
sys 0m0.003s
```
`$ time ./time_test_avxunroll`
```
N = 400000000 , pi = 3.141593
real 0m2.030s
user 0m2.028s
sys 0m0.002s
```
## 折線圖
利用 result_clock_gettime.csv 所得到的數據畫折線圖, 可以發現時間與N呈正向關關係，下圖為 loop = 1 的折線圖：
改寫 Makefile 中的 gencsv ，將數據點提高：
```
for i in `seq 100 100 25000`; do \
printf "%d " $$i;\
./benchmark_clock_gettime $$i; \
done > result_clock_gettime.csv
```
詳細用法 : `$ man seq`
可以發現，雖然大致上呈正比關係，但有許多處的 data 值飄走了！
## 平均值
利用取平均的方式，降低 bad data 對整體的影響，下圖是取 25 次所畫出的結果：
100次：
可以發現，在取平均值之後， bad data 的影響明顯變小了
## 95%信賴區間
信賴區間是對這個樣本的某個總體參數的區間估計，代表著真實值有一定機率落在與該測量結果有關的某對應區間。95%信賴區間指的就是，真實值有百分之九十五的機率會落在這個區間。
這裡改寫 benchmark_clock_gettime.c , 在取樣100次後找出95%的信賴區間, 並刪除不適當的資料後取平均：
將標準提高到 90％ ，也就是平均值正負 1.645 個標準差，可以發現，關係圖已經趨近於一直線了！
### Threads 數量對效能影響
當 thread number = Thread(s) per Core * Core(s) per Socket 時，效能最好 (N = 25000)
Threads 12 vs 16 (N = 100~25000)
## ERROR
為了檢視我們所算來的結果是否正確，比對實驗值與標準值的差異：
發現其中四條曲線重疊，唯獨 avxunroll 會不斷的飄動(擺動週期為 N = 16 ），原因為 avxunroll 中每一次 for 迴圈為 16 筆資料，然而實驗的 N 並非 16 的倍數，因此會產生數值錯誤的問題(餘數不會算到）
分別在avx與avxunroll之中，增加了檢查N並非4的倍數與16的倍數，並加以修正，得到以下的結果圖：
## OpenCL
>第一次使用 GPU 來加速程式，真的是太厲害啦！！！[name=ierosodin][color=green]
OpenCL 是一種異質性計算裝置進行平行化運算所設計的標準 API ，在同一種語法下，可以命令 CPU 或 GPU 來進行運算。 GPU 含有數千個更小型且更高效率的核心，專為同時處理多重任務進行最佳化， OpenCL 提供兩種平行化的模式，包括 task parallel 以及 data parallel ，而 data parallel 是最簡單的平行化方法，也就是如果有一大筆資料要進行同樣的運算就很適合。
OpenCL 大致的流程：
1. 把 OpenCL 裝置初始化。
* clGetPlatformIDs(1, &platform_id, &ret_num_platforms);
* clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_GPU, 1, &device_id, &ret_num_devices);
* clGetDeviceInfo(device_id, CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(size_t), &workGroupSize, NULL);
2. 在 OpenCL 裝置上配置記憶體
* clCreateContext( 0, 1, &device_id, NULL, NULL, &ret);
* clCreateCommandQueue(context, device_id, 0, &ret);
3. 編譯要執行的 OpenCL 程式（稱為 kernel）
* clCreateProgramWithSource(context, 1, (const char **)&source_str, NULL, NULL);
* clBuildProgram(program, 0, NULL, NULL, NULL, NULL);
4. 將資料內容複製到 OpenCL 裝置上
* clCreateKernel(program, "Pi", &ret);
* clCreateBuffer(context, CL_MEM_READ_WRITE, workGroups, sizeof(float), NULL, &ret);
* clSetKernelArg(kernel, 0, sizeof(cl_mem), &mem_obj);
5. 執行編譯好的 kernel
* clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, &globalWorkSize, &localWorkSize, 0, NULL, NULL);
6. 把計算結果從 OpenCL 裝置上，複製到記憶體中
* clEnqueueReadBuffer(command_queue, mem_obj, CL_TRUE, 0, sizeof(float)*workGroups, pi_partical, 0, NULL, NULL);
Kernel：
1. 獲得 localWorkSize
* get_local_id(0);
2. 獲得 globalWorkSize
* get_global_id(0);
3. 同步所有的 thread
* barrier(CLK_LOCAL_MEM_FENCE);
```clike=
__kernel void Pi(__global float *workGroupBuffer, __local float *scratch, const uint niter, const uint chunk) // NumWorkGroups, workGroupSize, Total iterations, Chunk size
{
const uint lid = get_local_id(0);
const uint gid = get_global_id(0);
const float h = (1.0/(float)niter);
float partial_sum = 0.0;
// Each thread compute chunk iterations
for(uint i=gid*chunk; i<(gid*chunk)+chunk; i++) {
float x = h * ((float) i - 0.5);
partial_sum += 4.0 / (1.0 + x * x);
}
// Each thread store its partial sum in the workgroup array
scratch[lid] = partial_sum;
// Synchronize all threads within the workgroup
barrier(CLK_LOCAL_MEM_FENCE);
float local_pi = 0;
// Only thread 0 of each workgroup perform the reduction
// of that workgroup
if(lid == 0) {
const uint length = lid + get_local_size(0);
for (uint i = lid; i<length; i++) {
local_pi += scratch[i];
}
// It store the workgroup sum
// Final reduction, between block, is done out by CPU
workGroupBuffer[get_group_id(0)] = local_pi;
}
}
```
可以發現，當 N 小的時候，由於使用 GPU 需要花費時間成本在資料的搬移，因此速度較慢，但當 N 超過 40000 後，使用 OpenCL 則有大幅度的加速，且從圖中可以發現， OpenCL 的執行時間幾乎是常數，也就是無論切得再細，所花費的時間幾乎不變。（這樣就可以拉高 N 值來提高精確度！）
不過由於操作 GPU 時是使用單浮點數運算，所以當 N 小的時候，誤差相對也大許多。
* 不同的 chunk 數對效能的影響
嘗試將 chunks 設為 $2^x$ ，其中， $x$ 為小於 $log_2N$ 的最大值，以得到較小的 workGroups ，這樣可以減少資料在 GPU 與 CPU 之間的交換次數，結果發現，無論 $N$ 為多少，時間都是 0.000005秒。
>推測是資料運算時間還小於 CPU 一次 interrupt 最短的時間，因此運算時間皆相同。[name=ierosodin][color=green]
## getopt()
以往寫程式如果想要在執行時傳入參數，都會使用 argc argv ，缺點就是必須要很精確的排好參數的順序，還要寫一大串的偵錯判斷式。 getopt() 提供了另一種給參數的方式，由選項與選項參數所構成，執行程式在給參數時，只需要將選項與參數配好，就不需要按照順序傳入，程式判斷式的可讀性也會提升許多。
使用範例：`$ ./execute_file -n 1024 -c 32`
getopt() 的實作在 unistd.h 中，因此使用時要引入，其中包含了三個主要的參數：
* optarg 用來抓取選項的參數
* optind 用來記錄 getopt() 已找過參數的數量
範例程式：
```clike
optind = 0;
while ((opt = getopt(argc, argv, "nt:")) != -1) {
switch (opt) {
case 'n':
FUNC(optarg);
break;
case 't':
int a = atoi(optarg);
break;
default: /* '?' */
fprintf(stderr, "Usage: \n");
exit(EXIT_FAILURE);
}
}
```
getopt() 可以抓取執行命令中的參數， `nt:` 為要找的選項，當找到時會傳回選項，將該選項的參數記錄在 optarg 中，並將 optind 加一，直到找完為止，傳回 -1 。
>好的習慣：由於 optind 是用來記錄 getopt() 找過的參數數量，因此在找過一次後， optind 會變成最大值，若在程式中呼叫兩次 getopt() ，便會出現找不到參數的問題，因此在每次調用 getopt() 前，先將 optind 設為零可以避免這個問題的出現。[name=ierosodin][color=green]
## 參考資料
[王紹華的共筆](https://www.emailhelpdesk.us/support-for-eumx.html)
[信賴區間與信心水準的解讀](https://www.emailhelpdesk.us/support-for-emma.html)
[OpenCL benchmark. PI calculation](https://www.emailhelpdesk.us/support-for-runbox.html)
[man getopt](https://www.emailhelpdesk.us/support-for-pobox.html)
[Linux getopt twice](https://www.emailhelpdesk.us/support-for-atmail.html)
Sign in with Wallet
Connect another wallet