前回はとりあえずコードを動かした。 http://tkokamo.hateblo.jp/
前回のコードを見るとカーネルコードの実装以前にclGetPlatformIDsやclGetDeviceIDsという函数が呼ばれていて気になったので、デバイスの情報取得周りについて調べてみた。
基本は前回のコードベースで気になったことを気が済むまで掘り下げる、ということにする。
今回のコード
下のコードがとりあえずまとまっていてよいなと思ったので、とりあえずこいつを動かしてみる。
[takuya@localhost device]$ g++ dev.cpp -lOpenCL
[takuya@localhost device]$ ./a.out
clDeviceQuery Starting...
1 OpenCL Platforms found
CL_PLATFORM_NAME: Intel(R) OpenCL
CL_PLATFORM_VERSION: OpenCL 1.2 LINUX
OpenCL Device Info:
1 devices found supporting OpenCL on: Intel(R) OpenCL
----------------------------------
Device Intel(R) Core(TM) i7-4770K CPU @ 3.50GHz
---------------------------------
CL_DEVICE_NAME: Intel(R) Core(TM) i7-4770K CPU @ 3.50GHz
CL_DEVICE_VENDOR: Intel(R) Corporation
CL_DRIVER_VERSION: 1.2.0.25
CL_DEVICE_TYPE: CL_DEVICE_TYPE_CPU
CL_DEVICE_MAX_COMPUTE_UNITS: 8
CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS: 3
CL_DEVICE_MAX_WORK_ITEM_SIZES: 8192 / 8192 / 8192
CL_DEVICE_MAX_WORK_GROUP_SIZE: 8192
CL_DEVICE_MAX_CLOCK_FREQUENCY: 3500 MHz
CL_DEVICE_ADDRESS_BITS: 64
CL_DEVICE_MAX_MEM_ALLOC_SIZE: 7971 MByte
CL_DEVICE_GLOBAL_MEM_SIZE: 31885 MByte
CL_DEVICE_ERROR_CORRECTION_SUPPORT: no
CL_DEVICE_LOCAL_MEM_TYPE: global
CL_DEVICE_LOCAL_MEM_SIZE: 32 KByte
CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE: 128 KByte
CL_DEVICE_QUEUE_PROPERTIES: CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE
CL_DEVICE_QUEUE_PROPERTIES: CL_QUEUE_PROFILING_ENABLE
CL_DEVICE_IMAGE_SUPPORT: 1
CL_DEVICE_MAX_READ_IMAGE_ARGS: 480
CL_DEVICE_MAX_WRITE_IMAGE_ARGS: 480
CL_DEVICE_IMAGE <dim> 2D_MAX_WIDTH 16384
2D_MAX_HEIGHT 16384
3D_MAX_WIDTH 2048
3D_MAX_HEIGHT 2048
3D_MAX_DEPTH 2048
CL_DEVICE_PREFERRED_VECTOR_WIDTH_<t> CHAR 1, SHORT 1, INT 1, FLOAT 1, DOUBLE 1
clDeviceQuery, Platform Name = Intel(R) OpenCL, Platform Version = OpenCL 1.2 LINUX, NumDevs = 1, Device = Intel(R) Core(TM) i7-4770K CPU @ 3.50GHz
System Info:
Local Time/Date = 23:22:49, 04/04/2019
CPU Name: Intel(R) Core(TM) i7-4770K CPU @ 3.50GHz
# of CPU processors: 8
Linux version 3.10.0-514.16.1.el7.x86_64 (builder@kbuilder.dev.centos.org) (gcc version 4.8.5 20150623 (Red Hat 4.8.5-11) (GCC) ) #1 SMP Wed Apr 12 15:04:24 UTC 2017
TEST PASSED
いくつか???となるものがあるので、函数の中身を実際に追っていきたいと思う。
関数の中身
mainから順に見て行く。マニュアルは手元に、、、
main()
165 cl_int ciErrNum = clGetPlatformIDs(0, NULL, &num_platforms);
マニュアルには"Obtain the list of platforms available." だけ書かれていてぱっとしないのだが恐らくopenCLの実行環境?のIDを取ることが出来るんだろう。。。
(上では第1引数0で第2引数NULLなので、num_platformsに存在するplatform数を入れるみたい)
182 ciErrNum = clGetPlatformIDs (num_platforms, clPlatformIDs, NULL); 183 for(cl_uint i = 0; i < num_platforms; ++i) { // ここからplatform数情報表示する
実際にIDを取ってきている部分。
160 char cBuffer[1024]; ... 184 ciErrNum = clGetPlatformInfo (clPlatformIDs[i], CL_PLATFORM_NAME, 1024, &cBuffer, NULL); // CL_PLATFORM_NAME: Intel(R) OpenCL
この関数は第1引数で指定されたIDのplatformについて第2引数で指定された値を取得する。
ここではplatform名を取得している。
198 ciErrNum = clGetPlatformInfo (clSelectedPlatformID, CL_PLATFORM_VERSION, sizeof(cBuffer), cBuffer, NULL); // CL_PLATFORM_VERSION: OpenCL 1.2 LINUX
platformのバージョン
210 // Get and log OpenCL device info 211 cl_uint ciDeviceCount; 212 cl_device_id *devices; 213 printf("OpenCL Device Info:\n\n"); 214 ciErrNum = clGetDeviceIDs (clSelectedPlatformID, CL_DEVICE_TYPE_ALL, 0, NULL, &ciDeviceCount);
これはclGetPlatformIDs()のデバイスバージョン。で、ここではデバイス数を取得している。
235 ciErrNum = clGetDeviceIDs (clSelectedPlatformID, CL_DEVICE_TYPE_ALL, ciDeviceCount, devices, &ciDeviceCount);
ここでデバイスIDを取得して、、、
237 for(unsigned int i = 0; i < ciDeviceCount; ++i ) { 238 printf(" ----------------------------------\n"); 239 clGetDeviceInfo(devices[i], CL_DEVICE_NAME, sizeof(cBuffer), &cBuffer, NULL); 240 printf(" Device %s\n", cBuffer); // Device Intel(R) Core(TM) i7-4770K CPU @ 3.50GHz 241 printf(" ---------------------------------\n"); 242 clPrintDevInfo(devices[i]); 243 sProfileString += ", Device = "; 244 sProfileString += cBuffer; 245 }
ここでデバイスの情報を取得/表示している。clPrintDevInfo()はコード内に定義されているので次はこの函数を見てみる。
clPrintDevInfo
ここまできたらなんとなく分かるものが多いので、そういうものははしょる。
21 clGetDeviceInfo(device, CL_DEVICE_NAME, sizeof(device_string), &device_string, NULL); // CL_DEVICE_NAME: Intel(R) Core(TM) i7-4770K CPU @ 3.50GHz 25 clGetDeviceInfo(device, CL_DEVICE_VENDOR, sizeof(device_string), &device_string, NULL); // CL_DEVICE_VENDOR: Intel(R) Corporation 29 clGetDeviceInfo(device, CL_DRIVER_VERSION, sizeof(device_string), &device_string, NULL); // CL_DRIVER_VERSION: 1.2.0.25
ドライバーって何のこと?と思ったのでマニュアルを見てみると、
"OpenCL software driver version string in the form major_number.minor_number."と書かれている。
結局OpenCLのバージョン?
34 clGetDeviceInfo(device, CL_DEVICE_TYPE, sizeof(type), &type, NULL); 35 if( type & CL_DEVICE_TYPE_CPU ) 36 printf(" CL_DEVICE_TYPE:\t\t\t%s\n", "CL_DEVICE_TYPE_CPU"); // CL_DEVICE_TYPE: CL_DEVICE_TYPE_CPU 46 clGetDeviceInfo(device, CL_DEVICE_MAX_COMPUTE_UNITS, sizeof(compute_units), &compute_units, NULL); 47 printf(" CL_DEVICE_MAX_COMPUTE_UNITS:\t\t%u\n", compute_units); // CL_DEVICE_MAX_COMPUTE_UNITS: 8
まぁスレッド数でしょうと思いつつも説明を見てみる。
"The number of parallel compute units on the OpenCL device. A work-group executes on a single compute unit. The minimum value is 1."
GPUとかだとどうなるんだろうか?(ソースコード配布されているgit上でGPUの情報があるが、これは2だった、、、)
51 clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS, sizeof(workitem_dims), &workitem _dims, NULL); // CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS: 3
"Maximum dimensions that specify the global and local work-item IDs used by the data parallel execution model. (Refer to clEnqueueNDRangeKernel). The minimum value is 3 for devices that are not of type CL_DEVICE_TYPE_CUSTOM."
なんだろう、、、CUDAでいうgrid, block, threadみたいなものなんだろうか?
56 clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(workitem_size), &workitem_size , NULL); // CL_DEVICE_MAX_WORK_ITEM_SIZES: 8192 / 8192 / 8192
各DIMENTIONで指定できる最大のwork-items、、、らしい。
こういうのはカーネルを動かす時にもっとわかると思う。
61 clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(workgroup_size), &workgroup_si ze, NULL); // CL_DEVICE_MAX_WORK_GROUP_SIZE: 8192
"Maximum number of work-items in a work-group that a device is capable of executing on a single compute unit"
ふむ、これもカーネル動かす時に意識するのかな、、、
clEnqueueNDRangeKernel()が関連しているらしいので、カーネルのコード実行部分のプログラムをみる時に調べてみる。
66 clGetDeviceInfo(device, CL_DEVICE_MAX_CLOCK_FREQUENCY, sizeof(clock_frequency), &clock_frequ ency, NULL); // CL_DEVICE_MAX_CLOCK_FREQUENCY: 3500 MHz 71 clGetDeviceInfo(device, CL_DEVICE_ADDRESS_BITS, sizeof(addr_bits), &addr_bits, NULL); // CL_DEVICE_ADDRESS_BITS: 64 76 clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE, sizeof(max_mem_alloc_size), &max_mem_a lloc_size, NULL); // CL_DEVICE_MAX_MEM_ALLOC_SIZE: 7971 MByte
"Max size of memory object allocation in bytes. The minimum value is max(min(102410241024, 1/4th of CL_DEVICE_GLOBAL_MEM_SIZE), 3210241024) for devices that are not of type CL_DEVICE_TYPE_CUSTOM"
なので 1/4th of CL_DEVICE_GLOBAL_MEM_SIZE(メモリ容量)が取られたんでしょう。
これはKBか?
81 clGetDeviceInfo(device, CL_DEVICE_GLOBAL_MEM_SIZE, sizeof(mem_size), &mem_size, NULL); // CL_DEVICE_GLOBAL_MEM_SIZE: 31885 MByte
メモリ容量
86 clGetDeviceInfo(device, CL_DEVICE_ERROR_CORRECTION_SUPPORT, sizeof(error_correction_support) , &error_correction_support, NULL); // CL_DEVICE_ERROR_CORRECTION_SUPPORT: no
メモリがECCをサポートしているか、これはdmpidecodeでみることができる。
[takuya@localhost ~]$ sudo dmidecode --type memory
[sudo] password for takuya:
# dmidecode 3.0
# SMBIOS entry point at 0x000f04c0
Found SMBIOS entry point in EFI, reading table from /dev/mem.
SMBIOS 2.7 present.
Handle 0x0042, DMI type 17, 34 bytes
Memory Device
Array Handle: 0x0043
Error Information Handle: Not Provided ★
Total Width: 64 bits ★
Data Width: 64 bits ★
Size: 8192 MB
Form Factor: DIMM
Set: None
Locator: ChannelA-DIMM0
Bank Locator: BANK 0
Type: DDR3
Type Detail: Synchronous
Speed: 1333 MHz
Manufacturer: 0215
Serial Number: 00000000
Asset Tag: 9876543210
Part Number: CMZ32GX3M4A1866C10
Rank: 2
Configured Clock Speed: 1333 MHz
91 clGetDeviceInfo(device, CL_DEVICE_LOCAL_MEM_TYPE, sizeof(local_mem_type), &local_mem_type, N ULL); CL_DEVICE_LOCAL_MEM_TYPE: global
デバイス固有にsramのようなメモリ(キャッシュは対応しない?、、、)を持っている場合はlocal、そうでない場合はglobalらしい。
globalやlocalのモデルは下の記事が詳しい。openCLのモデルについては後日調べるので今回はつっこまない。
https://www.mql5.com/ja/articles/407
95 clGetDeviceInfo(device, CL_DEVICE_LOCAL_MEM_SIZE, sizeof(mem_size), &mem_size, NULL); CL_DEVICE_LOCAL_MEM_SIZE: 32 KByte
ローカルメモリサイズ。ローカル、といっても全てのwork itemからアクセス可能なメモリになるらしい。(上記の記事から)
manualには最低32KBと書かれているので、それが出ているのか、
使っているCPUのL1 DCACHEのサイズが32KBだからそれが出ているのか分からない。
(L1はコアごとだけどコヒーレンシあるので、全work itemからアクセス可能とみなせるだろう、、、多分)
99 clGetDeviceInfo(device, CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE, sizeof(mem_size), &mem_size, NUL L); CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE: 128 KByte
コンスタントバッファとして獲得できる最大サイズ。globalメモリの一部らしい。
CL_DEVICE_QUEUE_PROPERTIES: CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE
CL_DEVICE_QUEUE_PROPERTIES: CL_QUEUE_PROFILING_ENABLE
CL_DEVICE_IMAGE_SUPPORT: 1
CL_DEVICE_MAX_READ_IMAGE_ARGS: 480
CL_DEVICE_MAX_WRITE_IMAGE_ARGS: 480
CL_DEVICE_IMAGE <dim> 2D_MAX_WIDTH 16384
2D_MAX_HEIGHT 16384
3D_MAX_WIDTH 2048
3D_MAX_HEIGHT 2048
3D_MAX_DEPTH 2048
CL_DEVICE_PREFERRED_VECTOR_WIDTH_<t> CHAR 1, SHORT 1, INT 1, FLOAT 1, DOUBLE 1
イメージとは、、、なんだろう、、、とりあえず置いとく。
適当だけどひとまずこんなところか