Re: [問題] openCL kernal中如何知道index屬於誰?

[LPH66 (かつて交わした約束)]

※ 引述《hardman1110 (笨小孩)》之銘言：
: 推 LPH66: index 是 kernel 的編號, 不是陣列的編號                   08/22 17:04
: → LPH66: kernel 做事的概念是「幾號 kernel 去拿陣列的哪幾格做事」  08/22 17:06
: → LPH66: 只是一般來說會因為方便讓 kernel 以自己編號去取元素       08/22 17:06
: → LPH66: 當然其他狀況也有可能是會成一個 kernel 抓很多格做事       08/22 17:07
: → LPH66: 這時去取哪幾格時就要自己去算哪個 kernel 該取哪幾格       08/22 17:07
: → LPH66: 另外你最後的問題, 去找 OpenCL 標準文件裡面寫得很清楚     08/22 17:09
: → LPH66: 或者我印象中有看過一個是把各種函數整理成的 cheat sheet   08/22 17:09
: → LPH66: (包含 host 端和 kernel 端的函數都有)                     08/22 17:09
: → hardman1110: 我傳進去的陣列型態大小都不一樣，我該如何知道切了   08/22 17:11
: → hardman1110: 多少kernal(work item) 然後每個work item包含每個    08/22 17:12
: → hardman1110: 陣列的哪幾格?                                      08/22 17:13
: 推 LPH66: 這是你在決定的事; 你要決定每個 work item 的工作量是多少  08/22 17:25
: → LPH66: 每個人要怎麼拿到他所要的工作量做事                       08/22 17:25
: → hardman1110: 查到clEnqueueNDRangeKernel可控制要切多少workitem   08/22 17:26
: → LPH66: 有的時候甚至可以寫說「幾號以後不用做那邊的事」之類的     08/22 17:26
: → hardman1110: 但每個workitem對應到陣列哪一段就不清楚原理了       08/22 17:26
: → LPH66: 然後再 enqueue 你要的數量的 workitem 去做事              08/22 17:26
: → LPH66: 例如如果我要一個 workitem 做一格, 那就是照編號取元素     08/22 17:27
: → LPH66: 如果我要一個 workitem 做兩格, 那可以編號乘二再取兩格     08/22 17:27
: → LPH66: 重點是你要搞懂你是怎麼分配工作下去的                     08/22 17:28
: → LPH66: 一個簡單的比方是你請了一堆工讀生分擔工作                 08/22 17:29
: → LPH66: 那你總要告訴工讀生「你的工作範圍在哪裡」                 08/22 17:29
: → LPH66: 而做出一個簡單易懂的分配方式就是你在問的"對應"           08/22 17:29
: → LPH66: 讓每個工讀生能只從自己分到的編號直接領到工作做           08/22 17:30
好像越推越長了所以回文總結 & 換個方式敘述一下

這裡其實有一個借鏡另外一個平行工作的 framework 的想法

在 OpenMP 裡面, 常見的一個平行法是把工作包在一個 for 裡

例如:

#pramga omp parallel for
for(int index = 0; index < 128; index++)
{
    C[index] = A[index] + B[index];
}

以 OpenCL 的話來說, 這個 for 的內容就是一個 kernel

而 for 本身即是 enqueue 的動作, index 就會變成 global_id 等等的編號

同樣的東西寫成 OpenCL 可能會像是:

// host
...
clEnqueueNDRangeKernel(..., func, ..., {128}, ...);
...

// kernel
__kernel void func(__global int *A, __global int *B, __global int *C)
{
    int index = get_global_id(0);
    C[index] = A[index] + B[index];
}

那當你有不等長的資料時

如同你在寫一般的 for 一樣有很多方法可以去分配每一圈所做的事

(例如我推文提的「幾號之後／第幾圈之後別做那邊的事」)

OpenCL 只是把這部份的迴圈內容給抽出來成了 kernel 以平行化而已

把「每個 kernel」想成「for 的每次迴圈」一些概念就很容易轉化

我上面所提的「工作分配」等等的就只是在這底下的概念而已

--
 ◢ ˊ_▂▃▄▂_ˋ. ◣　　　　　 　　　　▅▅　▅▅　ι●╮　　 █▄▄▄▄▄   
▍./◤_▂▃▄▂_◥ \'▊ 　　ＨＡＲＵＨＩ █████　＜■┘  　 ▄▄▄▄▄▄▄ 
▎⊿ ◤◤◥█◥◥█Δ   　　ＩＳＭ　 　　By-gamejye　￠|\  　　▌▌▌▌▌▄▌▌
▏ζ(▏●‵◥′●▊)Ψ ▏　　 　　　 　　　　█　　　 ⊿Δ 　　▄▄▄  ▄▄▄▄
█/|▊ 〃  、  〃▋ |\ ▎　　　　　 　 　ハルヒ主義　　　  　　█▄▄▄█▄▄  
◥◥|◣   ‵′  ◢/'◢◢S.O.S 世界を大いに盛り上げるための涼宮ハルヒの団　　　 

--
※ 發信站: 批踢踢實業坊(ptt.cc), 來自: 140.112.30.32
※ 文章網址: https://www.ptt.cc/bbs/C_and_CPP/M.1503394860.A.B47.html
※ 編輯: LPH66 (140.112.30.32), 08/22/2017 17:44:31
推 hardman1110: 感謝 所以opencl只是幫你切出平行化的"kernal"
→ hardman1110: 裡面根本沒綁定array index 只是因為平行特性可以拿
→ hardman1110: work item的編號當作array index >> 是這樣嗎?
推 VictorTom: 推:)
→ LPH66: 是的, 這也是我第一句推文「index 是 kernel 的編號」之意
推 ax61316: 推！講解的很清楚