[PR]
×
[PR]上記の広告は3ヶ月以上新規記事投稿のないブログに表示されています。新しい記事を書く事で広告が消えます。
プログラミング、3DCGとその他いろいろについて
[PR]上記の広告は3ヶ月以上新規記事投稿のないブログに表示されています。新しい記事を書く事で広告が消えます。
こちらを合わせてお読みください。
今回はOpenCL Cに用意された関数、get_group_id()とget_local_id()について解説します。
以前私の書いた文章から引用します:
今回は、OpenCL C言語に用意された関数、get_group_id()とget_local_id()についてメモします。
get_group_id()
get_local_id()
size_t get_group_id(uint dimIndex) size_t get_local_id(uint dimIndex)
dimIndexは、次元の数です。0はx、1はy、2はzですね。
戻り値はそれぞれのIDを意味します。
どういう事かというと、こういうことです:
GPGPUには、スレッドがたくさんあるのですが、ただだらーっと並べられるのではなく、スレッドは3次元の構造を取ります。
そしてそれがいくつかのグループにまとめられているのです。
get_group_idは、それを実行するスレッドの所属するグループが、何グループ目なのかを表します。
get_local_idは、それを実行するスレッドが、グループの中で何番目のスレッドなのかをあらわします。
上の2つの関数のdimIndexは、このうちXかYかZかを表すのです。
…というわけで、OpenCLNetを使ってこの2つの関数の働きを見てみましょう。
using OpenCLNet; using System.Runtime.InteropServices; struct MyItem { public float GlobalID; public float GroupID; public float LocalID; public override string ToString() { return string.Format("{0}, {1}, {2}", GlobalID, GroupID, LocalID); } } class Program { static void Main() { var platform = OpenCL.GetPlatform(0); var context = platform.CreateDefaultContext(); var commandQueue = context.CreateCommandQueue(platform.QueryDevices(DeviceType.DEFAULT)[0]); var program = context.CreateProgramWithSource( System.IO.File.ReadAllText("myKernelProgram.cl") ); program.Build(); var kernel = program.CreateKernel("myKernelFunction"); const int elementCount = 6; Mem buffer = context.CreateBuffer( MemFlags.READ_WRITE, elementCount * System.Runtime.InteropServices.Marshal.SizeOf(typeof(MyItem)) ); kernel.SetArg(0, buffer); commandQueue.EnqueueNDRangeKernel( kernel, 1, null, new[] { elementCount }, new[] { 2 }); var readBack = new MyItem[elementCount]; ReadBuffer(commandQueue, buffer, readBack); foreach (var item in readBack) { System.Console.WriteLine(item); } kernel.Dispose(); program.Dispose(); commandQueue.Dispose(); context.Dispose(); } static void ReadBuffer<T>(CommandQueue queue, Mem buffer, T[] systemBuffer) where T :struct { GCHandle handle = GCHandle.Alloc(systemBuffer, GCHandleType.Pinned); queue.EnqueueReadBuffer( buffer, true, 0, Marshal.SizeOf(typeof(T)) * systemBuffer.Length, handle.AddrOfPinnedObject() ); handle.Free(); } }
typedef struct MyItem_s { float GlobalID; float GroupID; float LocalID; } MyItem; __kernel void myKernelFunction(__global MyItem* items) { int globalThreadID = get_global_id(0); items[globalThreadID].GlobalID = globalThreadID; items[globalThreadID].GroupID = get_group_id(0); items[globalThreadID].LocalID = get_local_id(0); }
このプログラムは、6スレッド作り、それを1グループ2スレッドでグループ分けしています。
(スレッド0、スレッド1)、(スレッド2、スレッド3)、(スレッド4、スレッド5)
そしてそれぞれのスレッドの、各IDを出力しています。 グローバルID、グループID、ローカルIDを、です。
その結果はこうなります。
0, 0, 0 1, 0, 1 2, 1, 0 3, 1, 1 4, 2, 0 5, 2, 1
つまり、
スレッド0 | スレッド1 | スレッド2 | スレッド3 | スレッド4 | スレッド5 | |
グローバルID | 0 | 1 | 2 | 3 | 4 | 5 |
グループID | 0 | 0 | 1 | 1 | 2 | 2 |
ローカルID | 0 | 1 | 0 | 1 | 0 | 1 |
というわけですね。