質問編集履歴
9
文章を少し変えました。
test
CHANGED
|
File without changes
|
test
CHANGED
|
@@ -18,7 +18,7 @@
|
|
|
18
18
|
|
|
19
19
|
data[index*M+index2] += 1.0;
|
|
20
20
|
|
|
21
|
-
のようにアクセスしたい。
|
|
21
|
+
のようにアクセスしたいと思っています。
|
|
22
22
|
|
|
23
23
|
|
|
24
24
|
|
|
@@ -30,7 +30,7 @@
|
|
|
30
30
|
|
|
31
31
|
を行うと常にindexとindex2に同じ値が格納されてしまいます。
|
|
32
32
|
|
|
33
|
-
ほとんどの部分が+1されず終わっています
|
|
33
|
+
ほとんどの部分が+1されず終わっています。
|
|
34
34
|
|
|
35
35
|
```
|
|
36
36
|
|
|
@@ -40,18 +40,10 @@
|
|
|
40
40
|
|
|
41
41
|
//
|
|
42
42
|
|
|
43
|
-
// main.cpp
|
|
43
|
+
// main3.cpp
|
|
44
|
-
|
|
45
|
-
// OpenCLHello
|
|
46
44
|
|
|
47
45
|
//
|
|
48
46
|
|
|
49
|
-
// Created by Satoyama Shogo on 2017/01/04.
|
|
50
|
-
|
|
51
|
-
// Copyright © 2017年 Satoyama Shogo. All rights reserved.
|
|
52
|
-
|
|
53
|
-
//
|
|
54
|
-
|
|
55
47
|
#include <iostream>
|
|
56
48
|
|
|
57
49
|
#include <vector>
|
8
補足情報を更新しました。
test
CHANGED
|
File without changes
|
test
CHANGED
|
@@ -368,7 +368,7 @@
|
|
|
368
368
|
|
|
369
369
|
###補足情報(言語/FW/ツール等のバージョンなど)
|
|
370
370
|
|
|
371
|
-
Mac
|
|
371
|
+
MacBook Pro 13インチ、OSはel capitanです。
|
|
372
372
|
|
|
373
373
|
|
|
374
374
|
|
7
main3\.cppのclEnqueueNDRangeKernelの第６パラメータをNULLにしました。
test
CHANGED
|
File without changes
|
test
CHANGED
|
@@ -40,10 +40,18 @@
|
|
|
40
40
|
|
|
41
41
|
//
|
|
42
42
|
|
|
43
|
-
// main
|
|
43
|
+
// main.cpp
|
|
44
|
+
|
|
45
|
+
// OpenCLHello
|
|
44
46
|
|
|
45
47
|
//
|
|
46
48
|
|
|
49
|
+
// Created by Satoyama Shogo on 2017/01/04.
|
|
50
|
+
|
|
51
|
+
// Copyright © 2017年 Satoyama Shogo. All rights reserved.
|
|
52
|
+
|
|
53
|
+
//
|
|
54
|
+
|
|
47
55
|
#include <iostream>
|
|
48
56
|
|
|
49
57
|
#include <vector>
|
|
@@ -254,7 +262,7 @@
|
|
|
254
262
|
|
|
255
263
|
local[1] = 1;
|
|
256
264
|
|
|
257
|
-
EC(clEnqueueNDRangeKernel(q, kernel, 2, offset, global,
|
|
265
|
+
EC(clEnqueueNDRangeKernel(q, kernel, 2, offset, global, NULL, 0, nullptr, nullptr), "clEnqueueNDRangeKernel");
|
|
258
266
|
|
|
259
267
|
|
|
260
268
|
|
|
@@ -318,6 +326,8 @@
|
|
|
318
326
|
|
|
319
327
|
|
|
320
328
|
|
|
329
|
+
|
|
330
|
+
|
|
321
331
|
//ここからカーネル部分
|
|
322
332
|
|
|
323
333
|
//kernel2.cl
|
6
kernel2\.clにget_work_dim\(\)を付け足しました。
test
CHANGED
|
File without changes
|
test
CHANGED
|
@@ -332,12 +332,18 @@
|
|
|
332
332
|
|
|
333
333
|
int index2 = get_global_id(1);
|
|
334
334
|
|
|
335
|
+
int dim = get_work_dim();
|
|
336
|
+
|
|
337
|
+
printf("get_work_dim = %d\n",dim);
|
|
338
|
+
|
|
335
339
|
printf("index = %d , index2 = %d \n",index,index2);
|
|
336
340
|
|
|
337
341
|
data[index*n+index2] += 1.0f;
|
|
338
342
|
|
|
339
343
|
}
|
|
340
344
|
|
|
345
|
+
|
|
346
|
+
|
|
341
347
|
```C++
|
|
342
348
|
|
|
343
349
|
```
|
5
Terminal上での実行環境を修正しました。
test
CHANGED
|
File without changes
|
test
CHANGED
|
@@ -358,7 +358,7 @@
|
|
|
358
358
|
|
|
359
359
|
Terminal上で
|
|
360
360
|
|
|
361
|
-
/System/Library/Frameworks/OpenCL.framework/Libraries/openclc -c -o kernel
|
|
361
|
+
/System/Library/Frameworks/OpenCL.framework/Libraries/openclc -c -o kernel2.cl.gpu_32.bc -arch gpu_32 -emit-llvm kernel2.cl
|
|
362
362
|
|
|
363
363
|
g++ -O3 -std=c++11 -framework opencl main3.cpp -o test
|
|
364
364
|
|
4
該当のソースコードのところを修正しました。
test
CHANGED
|
File without changes
|
test
CHANGED
|
@@ -38,310 +38,308 @@
|
|
|
38
38
|
|
|
39
39
|
###該当のソースコード
|
|
40
40
|
|
|
41
|
+
//
|
|
42
|
+
|
|
43
|
+
// main3.cpp
|
|
44
|
+
|
|
45
|
+
//
|
|
46
|
+
|
|
47
|
+
#include <iostream>
|
|
48
|
+
|
|
49
|
+
#include <vector>
|
|
50
|
+
|
|
51
|
+
#include <OpenCL/opencl.h>
|
|
52
|
+
|
|
53
|
+
#include <numeric>
|
|
54
|
+
|
|
55
|
+
|
|
56
|
+
|
|
57
|
+
#define PLATFORM_MAX 4
|
|
58
|
+
|
|
59
|
+
#define DEVICE_MAX 4
|
|
60
|
+
|
|
61
|
+
|
|
62
|
+
|
|
63
|
+
|
|
64
|
+
|
|
65
|
+
void EC(cl_int result, const char *title)
|
|
66
|
+
|
|
67
|
+
{
|
|
68
|
+
|
|
69
|
+
if (result != CL_SUCCESS) {
|
|
70
|
+
|
|
71
|
+
std::cout << "Error: " << title << "(" << result << ")\n";
|
|
72
|
+
|
|
73
|
+
}
|
|
74
|
+
|
|
75
|
+
}
|
|
76
|
+
|
|
77
|
+
|
|
78
|
+
|
|
79
|
+
|
|
80
|
+
|
|
81
|
+
cl_int err = CL_SUCCESS;
|
|
82
|
+
|
|
83
|
+
void EC2(const char *title)
|
|
84
|
+
|
|
85
|
+
{
|
|
86
|
+
|
|
87
|
+
if (err != CL_SUCCESS) {
|
|
88
|
+
|
|
89
|
+
std::cout << "Error: " << title << "(" << err << ")\n";
|
|
90
|
+
|
|
91
|
+
}
|
|
92
|
+
|
|
93
|
+
err = CL_SUCCESS;
|
|
94
|
+
|
|
95
|
+
}
|
|
96
|
+
|
|
97
|
+
|
|
98
|
+
|
|
99
|
+
|
|
100
|
+
|
|
101
|
+
int main(int argc, const char * argv[])
|
|
102
|
+
|
|
103
|
+
{
|
|
104
|
+
|
|
105
|
+
// プラットフォーム一覧を取得
|
|
106
|
+
|
|
107
|
+
cl_platform_id platforms[PLATFORM_MAX];
|
|
108
|
+
|
|
109
|
+
cl_uint platformCount;
|
|
110
|
+
|
|
111
|
+
EC(clGetPlatformIDs(PLATFORM_MAX, platforms, &platformCount), "clGetPlatformIDs");
|
|
112
|
+
|
|
113
|
+
if (platformCount == 0) {
|
|
114
|
+
|
|
115
|
+
std::cerr << "No platform.\n";
|
|
116
|
+
|
|
117
|
+
return EXIT_FAILURE;
|
|
118
|
+
|
|
119
|
+
}
|
|
120
|
+
|
|
121
|
+
|
|
122
|
+
|
|
123
|
+
// 見つかったプラットフォームの情報を印字
|
|
124
|
+
|
|
125
|
+
for (int i = 0; i < platformCount; i++) {
|
|
126
|
+
|
|
127
|
+
char vendor[100] = {0};
|
|
128
|
+
|
|
129
|
+
char version[100] = {0};
|
|
130
|
+
|
|
131
|
+
EC(clGetPlatformInfo(platforms[i], CL_PLATFORM_VENDOR, sizeof(vendor), vendor, nullptr), "clGetPlatformInfo");
|
|
132
|
+
|
|
133
|
+
EC(clGetPlatformInfo(platforms[i], CL_PLATFORM_VERSION, sizeof(version), version, nullptr), "clGetPlatformInfo");
|
|
134
|
+
|
|
135
|
+
std::cout << "Platform id: " << platforms[i] << ", Vendor: " << vendor << ", Version: " << version << "\n";
|
|
136
|
+
|
|
137
|
+
}
|
|
138
|
+
|
|
139
|
+
|
|
140
|
+
|
|
141
|
+
// デバイス一覧を取得
|
|
142
|
+
|
|
143
|
+
cl_device_id devices[DEVICE_MAX];
|
|
144
|
+
|
|
145
|
+
cl_uint deviceCount;
|
|
146
|
+
|
|
147
|
+
EC(clGetDeviceIDs(platforms[0], CL_DEVICE_TYPE_GPU, DEVICE_MAX, devices, &deviceCount), "clGetDeviceIDs");
|
|
148
|
+
|
|
149
|
+
if (deviceCount == 0) {
|
|
150
|
+
|
|
151
|
+
std::cerr << "No device.\n";
|
|
152
|
+
|
|
153
|
+
return EXIT_FAILURE;
|
|
154
|
+
|
|
155
|
+
}
|
|
156
|
+
|
|
157
|
+
|
|
158
|
+
|
|
159
|
+
// 見つかったデバイスの情報を印字
|
|
160
|
+
|
|
161
|
+
std::cout << deviceCount << " device(s) found.\n";
|
|
162
|
+
|
|
163
|
+
for (int i = 0; i < deviceCount; i++) {
|
|
164
|
+
|
|
165
|
+
char name[100] = {0};
|
|
166
|
+
|
|
167
|
+
size_t len;
|
|
168
|
+
|
|
169
|
+
EC(clGetDeviceInfo(devices[i], CL_DEVICE_NAME, sizeof(name), name, &len), "clGetDeviceInfo");
|
|
170
|
+
|
|
171
|
+
std::cout << "Device id: " << i << ", Name: " << name << "\n";
|
|
172
|
+
|
|
173
|
+
}
|
|
174
|
+
|
|
175
|
+
|
|
176
|
+
|
|
177
|
+
// コンテキストの作成
|
|
178
|
+
|
|
179
|
+
cl_context ctx = clCreateContext(nullptr, 1, devices, nullptr, nullptr, &err);
|
|
180
|
+
|
|
181
|
+
EC2("clCreateContext");
|
|
182
|
+
|
|
183
|
+
|
|
184
|
+
|
|
185
|
+
// コンパイル済みclプログラムの読み込み
|
|
186
|
+
|
|
187
|
+
const char* bitcode_path = "kernel2.cl.gpu_32.bc";
|
|
188
|
+
|
|
189
|
+
size_t len = strlen(bitcode_path);
|
|
190
|
+
|
|
191
|
+
cl_program program = clCreateProgramWithBinary(ctx, 1, devices, &len, (const unsigned char**)&bitcode_path, nullptr, &err);
|
|
192
|
+
|
|
193
|
+
EC2("clCreateProgramWithBinary");
|
|
194
|
+
|
|
195
|
+
|
|
196
|
+
|
|
197
|
+
// プログラムのビルド
|
|
198
|
+
|
|
199
|
+
EC(clBuildProgram(program, 1, devices, nullptr, nullptr, nullptr), "clBuildProgram");
|
|
200
|
+
|
|
201
|
+
|
|
202
|
+
|
|
203
|
+
// カーネルの作成
|
|
204
|
+
|
|
205
|
+
cl_kernel kernel = clCreateKernel(program, "addone", &err);
|
|
206
|
+
|
|
207
|
+
EC2("clCreateKernel");
|
|
208
|
+
|
|
209
|
+
|
|
210
|
+
|
|
211
|
+
// データを用意
|
|
212
|
+
|
|
213
|
+
int n = 10;
|
|
214
|
+
|
|
215
|
+
std::vector<float> data(n*n,0.0f);
|
|
216
|
+
|
|
217
|
+
// デバイスメモリを確保しつつデータをコピー
|
|
218
|
+
|
|
219
|
+
cl_mem device_mem = clCreateBuffer(ctx, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, sizeof(float) * n*n, data.data(), &err);
|
|
220
|
+
|
|
221
|
+
EC2("clCreateBuffer");
|
|
222
|
+
|
|
223
|
+
|
|
224
|
+
|
|
225
|
+
// カーネルの引数をセット
|
|
226
|
+
|
|
227
|
+
EC(clSetKernelArg(kernel, 0, sizeof(cl_mem), &device_mem), "clSetKernelArg");
|
|
228
|
+
|
|
229
|
+
EC(clSetKernelArg(kernel, 1, sizeof(int), &n), "clSetKernelArg");
|
|
230
|
+
|
|
231
|
+
|
|
232
|
+
|
|
233
|
+
// コマンドキューの作成
|
|
234
|
+
|
|
235
|
+
cl_command_queue q = clCreateCommandQueue(ctx, devices[0], 0, &err);
|
|
236
|
+
|
|
237
|
+
EC2("clCreateCommandQueue");
|
|
238
|
+
|
|
239
|
+
|
|
240
|
+
|
|
241
|
+
// カーネルの実行
|
|
242
|
+
|
|
243
|
+
size_t global[2],local[2],offset[2];
|
|
244
|
+
|
|
245
|
+
offset[0] = 0;
|
|
246
|
+
|
|
247
|
+
offset[0] = 0;
|
|
248
|
+
|
|
249
|
+
global[0] = n;
|
|
250
|
+
|
|
251
|
+
global[1] = n;
|
|
252
|
+
|
|
253
|
+
local[0] = 1;
|
|
254
|
+
|
|
255
|
+
local[1] = 1;
|
|
256
|
+
|
|
257
|
+
EC(clEnqueueNDRangeKernel(q, kernel, 2, offset, global, local, 0, nullptr, nullptr), "clEnqueueNDRangeKernel");
|
|
258
|
+
|
|
259
|
+
|
|
260
|
+
|
|
261
|
+
// 結果を読み込み
|
|
262
|
+
|
|
263
|
+
EC(clEnqueueReadBuffer(q, device_mem, CL_TRUE, 0, sizeof(float) * n*n, data.data(), 0, nullptr, nullptr), "clEnqueueReadBuffer");
|
|
264
|
+
|
|
265
|
+
|
|
266
|
+
|
|
267
|
+
// 結果の印字
|
|
268
|
+
|
|
269
|
+
for (int i = 0; i < n*n; i++) {
|
|
270
|
+
|
|
271
|
+
std::cout << data[i] << ", ";
|
|
272
|
+
|
|
273
|
+
}
|
|
274
|
+
|
|
275
|
+
std::cout << "\n";
|
|
276
|
+
|
|
277
|
+
float total = std::accumulate(data.begin(),data.end(),0.0);
|
|
278
|
+
|
|
279
|
+
std::cout << total << std::endl;
|
|
280
|
+
|
|
281
|
+
// コマンドキューの解放
|
|
282
|
+
|
|
283
|
+
EC(clReleaseCommandQueue(q), "clReleaseCommandQueue");
|
|
284
|
+
|
|
285
|
+
|
|
286
|
+
|
|
287
|
+
// デバイスメモリを解放
|
|
288
|
+
|
|
289
|
+
EC(clReleaseMemObject(device_mem), "clReleaseMemObject");
|
|
290
|
+
|
|
291
|
+
|
|
292
|
+
|
|
293
|
+
// カーネルの解放
|
|
294
|
+
|
|
295
|
+
EC(clReleaseKernel(kernel), "clReleaseKernel");
|
|
296
|
+
|
|
297
|
+
|
|
298
|
+
|
|
299
|
+
// プログラムの解放
|
|
300
|
+
|
|
301
|
+
EC(clReleaseProgram(program), "clReleaseProgram");
|
|
302
|
+
|
|
303
|
+
|
|
304
|
+
|
|
305
|
+
// コンテキストの解放
|
|
306
|
+
|
|
307
|
+
EC(clReleaseContext(ctx), "clReleaseContext");
|
|
308
|
+
|
|
309
|
+
|
|
310
|
+
|
|
311
|
+
std::cout << "Done.\n";
|
|
312
|
+
|
|
313
|
+
return EXIT_SUCCESS;
|
|
314
|
+
|
|
315
|
+
}
|
|
316
|
+
|
|
317
|
+
|
|
318
|
+
|
|
319
|
+
|
|
320
|
+
|
|
321
|
+
//ここからカーネル部分
|
|
322
|
+
|
|
323
|
+
//kernel2.cl
|
|
324
|
+
|
|
325
|
+
__kernel
|
|
326
|
+
|
|
327
|
+
void addone(__global float* data,const int n)
|
|
328
|
+
|
|
329
|
+
{
|
|
330
|
+
|
|
331
|
+
int index = get_global_id(0);
|
|
332
|
+
|
|
333
|
+
int index2 = get_global_id(1);
|
|
334
|
+
|
|
335
|
+
printf("index = %d , index2 = %d \n",index,index2);
|
|
336
|
+
|
|
337
|
+
data[index*n+index2] += 1.0f;
|
|
338
|
+
|
|
339
|
+
}
|
|
340
|
+
|
|
41
341
|
```C++
|
|
42
342
|
|
|
43
|
-
//
|
|
44
|
-
|
|
45
|
-
// main3.cpp
|
|
46
|
-
|
|
47
|
-
//
|
|
48
|
-
|
|
49
|
-
#include <iostream>
|
|
50
|
-
|
|
51
|
-
#include <vector>
|
|
52
|
-
|
|
53
|
-
#include <OpenCL/opencl.h>
|
|
54
|
-
|
|
55
|
-
#include <numeric>
|
|
56
|
-
|
|
57
|
-
|
|
58
|
-
|
|
59
|
-
#define PLATFORM_MAX 4
|
|
60
|
-
|
|
61
|
-
#define DEVICE_MAX 4
|
|
62
|
-
|
|
63
|
-
|
|
64
|
-
|
|
65
|
-
|
|
66
|
-
|
|
67
|
-
void EC(cl_int result, const char *title)
|
|
68
|
-
|
|
69
|
-
{
|
|
70
|
-
|
|
71
|
-
if (result != CL_SUCCESS) {
|
|
72
|
-
|
|
73
|
-
std::cout << "Error: " << title << "(" << result << ")\n";
|
|
74
|
-
|
|
75
|
-
}
|
|
76
|
-
|
|
77
|
-
}
|
|
78
|
-
|
|
79
|
-
|
|
80
|
-
|
|
81
|
-
|
|
82
|
-
|
|
83
|
-
cl_int err = CL_SUCCESS;
|
|
84
|
-
|
|
85
|
-
void EC2(const char *title)
|
|
86
|
-
|
|
87
|
-
{
|
|
88
|
-
|
|
89
|
-
if (err != CL_SUCCESS) {
|
|
90
|
-
|
|
91
|
-
std::cout << "Error: " << title << "(" << err << ")\n";
|
|
92
|
-
|
|
93
|
-
}
|
|
94
|
-
|
|
95
|
-
err = CL_SUCCESS;
|
|
96
|
-
|
|
97
|
-
}
|
|
98
|
-
|
|
99
|
-
|
|
100
|
-
|
|
101
|
-
|
|
102
|
-
|
|
103
|
-
int main(int argc, const char * argv[])
|
|
104
|
-
|
|
105
|
-
{
|
|
106
|
-
|
|
107
|
-
// プラットフォーム一覧を取得
|
|
108
|
-
|
|
109
|
-
cl_platform_id platforms[PLATFORM_MAX];
|
|
110
|
-
|
|
111
|
-
cl_uint platformCount;
|
|
112
|
-
|
|
113
|
-
EC(clGetPlatformIDs(PLATFORM_MAX, platforms, &platformCount), "clGetPlatformIDs");
|
|
114
|
-
|
|
115
|
-
if (platformCount == 0) {
|
|
116
|
-
|
|
117
|
-
std::cerr << "No platform.\n";
|
|
118
|
-
|
|
119
|
-
return EXIT_FAILURE;
|
|
120
|
-
|
|
121
|
-
}
|
|
122
|
-
|
|
123
|
-
|
|
124
|
-
|
|
125
|
-
// 見つかったプラットフォームの情報を印字
|
|
126
|
-
|
|
127
|
-
for (int i = 0; i < platformCount; i++) {
|
|
128
|
-
|
|
129
|
-
char vendor[100] = {0};
|
|
130
|
-
|
|
131
|
-
char version[100] = {0};
|
|
132
|
-
|
|
133
|
-
EC(clGetPlatformInfo(platforms[i], CL_PLATFORM_VENDOR, sizeof(vendor), vendor, nullptr), "clGetPlatformInfo");
|
|
134
|
-
|
|
135
|
-
EC(clGetPlatformInfo(platforms[i], CL_PLATFORM_VERSION, sizeof(version), version, nullptr), "clGetPlatformInfo");
|
|
136
|
-
|
|
137
|
-
std::cout << "Platform id: " << platforms[i] << ", Vendor: " << vendor << ", Version: " << version << "\n";
|
|
138
|
-
|
|
139
|
-
}
|
|
140
|
-
|
|
141
|
-
|
|
142
|
-
|
|
143
|
-
// デバイス一覧を取得
|
|
144
|
-
|
|
145
|
-
cl_device_id devices[DEVICE_MAX];
|
|
146
|
-
|
|
147
|
-
cl_uint deviceCount;
|
|
148
|
-
|
|
149
|
-
EC(clGetDeviceIDs(platforms[0], CL_DEVICE_TYPE_GPU, DEVICE_MAX, devices, &deviceCount), "clGetDeviceIDs");
|
|
150
|
-
|
|
151
|
-
if (deviceCount == 0) {
|
|
152
|
-
|
|
153
|
-
std::cerr << "No device.\n";
|
|
154
|
-
|
|
155
|
-
return EXIT_FAILURE;
|
|
156
|
-
|
|
157
|
-
}
|
|
158
|
-
|
|
159
|
-
|
|
160
|
-
|
|
161
|
-
// 見つかったデバイスの情報を印字
|
|
162
|
-
|
|
163
|
-
std::cout << deviceCount << " device(s) found.\n";
|
|
164
|
-
|
|
165
|
-
for (int i = 0; i < deviceCount; i++) {
|
|
166
|
-
|
|
167
|
-
char name[100] = {0};
|
|
168
|
-
|
|
169
|
-
size_t len;
|
|
170
|
-
|
|
171
|
-
EC(clGetDeviceInfo(devices[i], CL_DEVICE_NAME, sizeof(name), name, &len), "clGetDeviceInfo");
|
|
172
|
-
|
|
173
|
-
std::cout << "Device id: " << i << ", Name: " << name << "\n";
|
|
174
|
-
|
|
175
|
-
}
|
|
176
|
-
|
|
177
|
-
|
|
178
|
-
|
|
179
|
-
// コンテキストの作成
|
|
180
|
-
|
|
181
|
-
cl_context ctx = clCreateContext(nullptr, 1, devices, nullptr, nullptr, &err);
|
|
182
|
-
|
|
183
|
-
EC2("clCreateContext");
|
|
184
|
-
|
|
185
|
-
|
|
186
|
-
|
|
187
|
-
// コンパイル済みclプログラムの読み込み
|
|
188
|
-
|
|
189
|
-
const char* bitcode_path = "kernel2.cl.gpu_32.bc";
|
|
190
|
-
|
|
191
|
-
size_t len = strlen(bitcode_path);
|
|
192
|
-
|
|
193
|
-
cl_program program = clCreateProgramWithBinary(ctx, 1, devices, &len, (const unsigned char**)&bitcode_path, nullptr, &err);
|
|
194
|
-
|
|
195
|
-
EC2("clCreateProgramWithBinary");
|
|
196
|
-
|
|
197
|
-
|
|
198
|
-
|
|
199
|
-
// プログラムのビルド
|
|
200
|
-
|
|
201
|
-
EC(clBuildProgram(program, 1, devices, nullptr, nullptr, nullptr), "clBuildProgram");
|
|
202
|
-
|
|
203
|
-
|
|
204
|
-
|
|
205
|
-
// カーネルの作成
|
|
206
|
-
|
|
207
|
-
cl_kernel kernel = clCreateKernel(program, "addone", &err);
|
|
208
|
-
|
|
209
|
-
EC2("clCreateKernel");
|
|
210
|
-
|
|
211
|
-
|
|
212
|
-
|
|
213
|
-
// データを用意
|
|
214
|
-
|
|
215
|
-
int n = 10;
|
|
216
|
-
|
|
217
|
-
std::vector<float> data(n*n,0.0f);
|
|
218
|
-
|
|
219
|
-
// デバイスメモリを確保しつつデータをコピー
|
|
220
|
-
|
|
221
|
-
cl_mem device_mem = clCreateBuffer(ctx, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, sizeof(float) * n*n, data.data(), &err);
|
|
222
|
-
|
|
223
|
-
EC2("clCreateBuffer");
|
|
224
|
-
|
|
225
|
-
|
|
226
|
-
|
|
227
|
-
// カーネルの引数をセット
|
|
228
|
-
|
|
229
|
-
EC(clSetKernelArg(kernel, 0, sizeof(cl_mem), &device_mem), "clSetKernelArg");
|
|
230
|
-
|
|
231
|
-
EC(clSetKernelArg(kernel, 1, sizeof(int), &n), "clSetKernelArg");
|
|
232
|
-
|
|
233
|
-
|
|
234
|
-
|
|
235
|
-
// コマンドキューの作成
|
|
236
|
-
|
|
237
|
-
cl_command_queue q = clCreateCommandQueue(ctx, devices[0], 0, &err);
|
|
238
|
-
|
|
239
|
-
EC2("clCreateCommandQueue");
|
|
240
|
-
|
|
241
|
-
|
|
242
|
-
|
|
243
|
-
// カーネルの実行
|
|
244
|
-
|
|
245
|
-
size_t global[2],local[2],offset[2];
|
|
246
|
-
|
|
247
|
-
offset[0] = 0;
|
|
248
|
-
|
|
249
|
-
offset[0] = 0;
|
|
250
|
-
|
|
251
|
-
global[0] = n;
|
|
252
|
-
|
|
253
|
-
global[1] = n;
|
|
254
|
-
|
|
255
|
-
local[0] = 1;
|
|
256
|
-
|
|
257
|
-
local[1] = 1;
|
|
258
|
-
|
|
259
|
-
EC(clEnqueueNDRangeKernel(q, kernel, 2, offset, global, local, 0, nullptr, nullptr), "clEnqueueNDRangeKernel");
|
|
260
|
-
|
|
261
|
-
|
|
262
|
-
|
|
263
|
-
// 結果を読み込み
|
|
264
|
-
|
|
265
|
-
EC(clEnqueueReadBuffer(q, device_mem, CL_TRUE, 0, sizeof(float) * n*n, data.data(), 0, nullptr, nullptr), "clEnqueueReadBuffer");
|
|
266
|
-
|
|
267
|
-
|
|
268
|
-
|
|
269
|
-
// 結果の印字
|
|
270
|
-
|
|
271
|
-
for (int i = 0; i < n*n; i++) {
|
|
272
|
-
|
|
273
|
-
std::cout << data[i] << ", ";
|
|
274
|
-
|
|
275
|
-
}
|
|
276
|
-
|
|
277
|
-
std::cout << "\n";
|
|
278
|
-
|
|
279
|
-
float total = std::accumulate(data.begin(),data.end(),0.0);
|
|
280
|
-
|
|
281
|
-
std::cout << total << std::endl;
|
|
282
|
-
|
|
283
|
-
// コマンドキューの解放
|
|
284
|
-
|
|
285
|
-
EC(clReleaseCommandQueue(q), "clReleaseCommandQueue");
|
|
286
|
-
|
|
287
|
-
|
|
288
|
-
|
|
289
|
-
// デバイスメモリを解放
|
|
290
|
-
|
|
291
|
-
EC(clReleaseMemObject(device_mem), "clReleaseMemObject");
|
|
292
|
-
|
|
293
|
-
|
|
294
|
-
|
|
295
|
-
// カーネルの解放
|
|
296
|
-
|
|
297
|
-
EC(clReleaseKernel(kernel), "clReleaseKernel");
|
|
298
|
-
|
|
299
|
-
|
|
300
|
-
|
|
301
|
-
// プログラムの解放
|
|
302
|
-
|
|
303
|
-
EC(clReleaseProgram(program), "clReleaseProgram");
|
|
304
|
-
|
|
305
|
-
|
|
306
|
-
|
|
307
|
-
// コンテキストの解放
|
|
308
|
-
|
|
309
|
-
EC(clReleaseContext(ctx), "clReleaseContext");
|
|
310
|
-
|
|
311
|
-
|
|
312
|
-
|
|
313
|
-
std::cout << "Done.\n";
|
|
314
|
-
|
|
315
|
-
return EXIT_SUCCESS;
|
|
316
|
-
|
|
317
|
-
}
|
|
318
|
-
|
|
319
|
-
|
|
320
|
-
|
|
321
|
-
|
|
322
|
-
|
|
323
|
-
//ここからカーネル部分
|
|
324
|
-
|
|
325
|
-
//kernel2.cl
|
|
326
|
-
|
|
327
|
-
__kernel
|
|
328
|
-
|
|
329
|
-
void addone(__global float* data,const int n)
|
|
330
|
-
|
|
331
|
-
{
|
|
332
|
-
|
|
333
|
-
int index = get_global_id(0);
|
|
334
|
-
|
|
335
|
-
int index2 = get_global_id(1);
|
|
336
|
-
|
|
337
|
-
printf("index = %d , index2 = %d \n",index,index2);
|
|
338
|
-
|
|
339
|
-
data[index*n+index2] += 1.0f;
|
|
340
|
-
|
|
341
|
-
}
|
|
342
|
-
|
|
343
|
-
|
|
344
|
-
|
|
345
343
|
```
|
|
346
344
|
|
|
347
345
|
|
3
Terminal上での実行環境を修正しました。
test
CHANGED
|
File without changes
|
test
CHANGED
|
@@ -362,6 +362,8 @@
|
|
|
362
362
|
|
|
363
363
|
/System/Library/Frameworks/OpenCL.framework/Libraries/openclc -c -o kernel3.cl.gpu_32.bc -arch gpu_32 -emit-llvm kernel3.cl
|
|
364
364
|
|
|
365
|
-
g++ -O3 -std=c++11 -framework opencl
|
|
365
|
+
g++ -O3 -std=c++11 -framework opencl main3.cpp -o test
|
|
366
|
+
|
|
367
|
+
./test
|
|
366
368
|
|
|
367
369
|
で実行しました。
|
2
clSetKernelArgのsizeof\(cl_mem\)をsizeof\(int\)に変更しました。
test
CHANGED
|
File without changes
|
test
CHANGED
|
@@ -228,7 +228,7 @@
|
|
|
228
228
|
|
|
229
229
|
EC(clSetKernelArg(kernel, 0, sizeof(cl_mem), &device_mem), "clSetKernelArg");
|
|
230
230
|
|
|
231
|
-
EC(clSetKernelArg(kernel, 1, sizeof(
|
|
231
|
+
EC(clSetKernelArg(kernel, 1, sizeof(int), &n), "clSetKernelArg");
|
|
232
232
|
|
|
233
233
|
|
|
234
234
|
|
1
int index = get_global_id\(1\)をint index = get_global_id\(0\)に修正しました。
test
CHANGED
|
File without changes
|
test
CHANGED
|
@@ -330,7 +330,7 @@
|
|
|
330
330
|
|
|
331
331
|
{
|
|
332
332
|
|
|
333
|
-
int index = get_global_id(
|
|
333
|
+
int index = get_global_id(0);
|
|
334
334
|
|
|
335
335
|
int index2 = get_global_id(1);
|
|
336
336
|
|