1 //
2 // Copyright (c) 2017 The Khronos Group Inc.
3 //
4 // Licensed under the Apache License, Version 2.0 (the "License");
5 // you may not use this file except in compliance with the License.
6 // You may obtain a copy of the License at
7 //
8 // http://www.apache.org/licenses/LICENSE-2.0
9 //
10 // Unless required by applicable law or agreed to in writing, software
11 // distributed under the License is distributed on an "AS IS" BASIS,
12 // WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13 // See the License for the specific language governing permissions and
14 // limitations under the License.
15 //
16 #include "action_classes.h"
17
18 #pragma mark -------------------- Base Action Class -------------------------
19
20 const cl_uint BufferSizeReductionFactor = 20;
21
IGetPreferredImageSize2D(cl_device_id device,size_t & outWidth,size_t & outHeight)22 cl_int Action::IGetPreferredImageSize2D(cl_device_id device, size_t &outWidth,
23 size_t &outHeight)
24 {
25 cl_ulong maxAllocSize;
26 size_t maxWidth, maxHeight;
27 cl_int error;
28
29
30 // Get the largest possible buffer we could allocate
31 error = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
32 sizeof(maxAllocSize), &maxAllocSize, NULL);
33 error |= clGetDeviceInfo(device, CL_DEVICE_IMAGE2D_MAX_WIDTH,
34 sizeof(maxWidth), &maxWidth, NULL);
35 error |= clGetDeviceInfo(device, CL_DEVICE_IMAGE2D_MAX_HEIGHT,
36 sizeof(maxHeight), &maxHeight, NULL);
37 test_error(error, "Unable to get device config");
38
39 // Create something of a decent size
40 if (maxWidth * maxHeight * 4 > maxAllocSize / BufferSizeReductionFactor)
41 {
42 float rootSize =
43 sqrtf((float)(maxAllocSize / (BufferSizeReductionFactor * 4)));
44
45 if ((size_t)rootSize > maxWidth)
46 outWidth = maxWidth;
47 else
48 outWidth = (size_t)rootSize;
49 outHeight = (size_t)((maxAllocSize / (BufferSizeReductionFactor * 4))
50 / outWidth);
51 if (outHeight > maxHeight) outHeight = maxHeight;
52 }
53 else
54 {
55 outWidth = maxWidth;
56 outHeight = maxHeight;
57 }
58
59 outWidth /= 2;
60 outHeight /= 2;
61
62 if (outWidth > 2048) outWidth = 2048;
63 if (outHeight > 2048) outHeight = 2048;
64 log_info("\tImage size: %d x %d (%gMB)\n", (int)outWidth, (int)outHeight,
65 (double)((int)outWidth * (int)outHeight * 4) / (1024.0 * 1024.0));
66 return CL_SUCCESS;
67 }
68
IGetPreferredImageSize3D(cl_device_id device,size_t & outWidth,size_t & outHeight,size_t & outDepth)69 cl_int Action::IGetPreferredImageSize3D(cl_device_id device, size_t &outWidth,
70 size_t &outHeight, size_t &outDepth)
71 {
72 cl_ulong maxAllocSize;
73 size_t maxWidth, maxHeight, maxDepth;
74 cl_int error;
75
76
77 // Get the largest possible buffer we could allocate
78 error = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
79 sizeof(maxAllocSize), &maxAllocSize, NULL);
80 error |= clGetDeviceInfo(device, CL_DEVICE_IMAGE3D_MAX_WIDTH,
81 sizeof(maxWidth), &maxWidth, NULL);
82 error |= clGetDeviceInfo(device, CL_DEVICE_IMAGE3D_MAX_HEIGHT,
83 sizeof(maxHeight), &maxHeight, NULL);
84 error |= clGetDeviceInfo(device, CL_DEVICE_IMAGE3D_MAX_DEPTH,
85 sizeof(maxDepth), &maxDepth, NULL);
86 test_error(error, "Unable to get device config");
87
88 // Create something of a decent size
89 if ((cl_ulong)maxWidth * maxHeight * maxDepth
90 > maxAllocSize / (BufferSizeReductionFactor * 4))
91 {
92 float rootSize =
93 cbrtf((float)(maxAllocSize / (BufferSizeReductionFactor * 4)));
94
95 if ((size_t)rootSize > maxWidth)
96 outWidth = maxWidth;
97 else
98 outWidth = (size_t)rootSize;
99 if ((size_t)rootSize > maxHeight)
100 outHeight = maxHeight;
101 else
102 outHeight = (size_t)rootSize;
103 outDepth = (size_t)((maxAllocSize / (BufferSizeReductionFactor * 4))
104 / (outWidth * outHeight));
105 if (outDepth > maxDepth) outDepth = maxDepth;
106 }
107 else
108 {
109 outWidth = maxWidth;
110 outHeight = maxHeight;
111 outDepth = maxDepth;
112 }
113
114 outWidth /= 2;
115 outHeight /= 2;
116 outDepth /= 2;
117
118 if (outWidth > 512) outWidth = 512;
119 if (outHeight > 512) outHeight = 512;
120 if (outDepth > 512) outDepth = 512;
121 log_info("\tImage size: %d x %d x %d (%gMB)\n", (int)outWidth,
122 (int)outHeight, (int)outDepth,
123 (double)((int)outWidth * (int)outHeight * (int)outDepth * 4)
124 / (1024.0 * 1024.0));
125
126 return CL_SUCCESS;
127 }
128
129 #pragma mark -------------------- Execution Sub-Classes -------------------------
130
Setup(cl_device_id device,cl_context context,cl_command_queue queue)131 cl_int NDRangeKernelAction::Setup(cl_device_id device, cl_context context,
132 cl_command_queue queue)
133 {
134 const char *long_kernel[] = {
135 "__kernel void sample_test(__global float *src, __global int *dst)\n"
136 "{\n"
137 " int tid = get_global_id(0);\n"
138 " int i;\n"
139 "\n"
140 " for( i = 0; i < 100000; i++ )\n"
141 " {\n"
142 " dst[tid] = (int)src[tid] * 3;\n"
143 " }\n"
144 "\n"
145 "}\n"
146 };
147
148 size_t threads[1] = { 1000 };
149 int error;
150
151 if (create_single_kernel_helper(context, &mProgram, &mKernel, 1,
152 long_kernel, "sample_test"))
153 {
154 return -1;
155 }
156
157 error = get_max_common_work_group_size(context, mKernel, threads[0],
158 &mLocalThreads[0]);
159 test_error(error, "Unable to get work group size to use");
160
161 mStreams[0] = clCreateBuffer(context, CL_MEM_READ_WRITE,
162 sizeof(cl_float) * 1000, NULL, &error);
163 test_error(error, "Creating test array failed");
164 mStreams[1] = clCreateBuffer(context, CL_MEM_READ_WRITE,
165 sizeof(cl_int) * 1000, NULL, &error);
166 test_error(error, "Creating test array failed");
167
168 /* Set the arguments */
169 error = clSetKernelArg(mKernel, 0, sizeof(mStreams[0]), &mStreams[0]);
170 test_error(error, "Unable to set kernel arguments");
171 error = clSetKernelArg(mKernel, 1, sizeof(mStreams[1]), &mStreams[1]);
172 test_error(error, "Unable to set kernel arguments");
173
174 return CL_SUCCESS;
175 }
176
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)177 cl_int NDRangeKernelAction::Execute(cl_command_queue queue, cl_uint numWaits,
178 cl_event *waits, cl_event *outEvent)
179 {
180 size_t threads[1] = { 1000 };
181 cl_int error =
182 clEnqueueNDRangeKernel(queue, mKernel, 1, NULL, threads, mLocalThreads,
183 numWaits, waits, outEvent);
184 test_error(error, "Unable to execute kernel");
185
186 return CL_SUCCESS;
187 }
188
189 #pragma mark -------------------- Buffer Sub-Classes -------------------------
190
Setup(cl_device_id device,cl_context context,cl_command_queue queue)191 cl_int BufferAction::Setup(cl_device_id device, cl_context context,
192 cl_command_queue queue)
193 {
194 cl_int error;
195 cl_ulong maxAllocSize;
196
197
198 // Get the largest possible buffer we could allocate
199 error = clGetDeviceInfo(device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
200 sizeof(maxAllocSize), &maxAllocSize, NULL);
201
202 // Don't create a buffer quite that big, just so we have some space left
203 // over for other work
204 mSize = (size_t)(maxAllocSize / BufferSizeReductionFactor);
205
206 // Cap at 128M so tests complete in a reasonable amount of time.
207 if (mSize > 128 << 20) mSize = 128 << 20;
208
209 mSize /= 2;
210
211 log_info("\tBuffer size: %gMB\n", (double)mSize / (1024.0 * 1024.0));
212
213 mBuffer = clCreateBuffer(context, CL_MEM_READ_WRITE | CL_MEM_ALLOC_HOST_PTR,
214 mSize, NULL, &error);
215 test_error(error, "Unable to create buffer to test against");
216
217 mOutBuffer = malloc(mSize);
218 if (mOutBuffer == NULL)
219 {
220 log_error("ERROR: Unable to allocate temp buffer (out of memory)\n");
221 return CL_OUT_OF_RESOURCES;
222 }
223
224 return CL_SUCCESS;
225 }
226
Setup(cl_device_id device,cl_context context,cl_command_queue queue)227 cl_int ReadBufferAction::Setup(cl_device_id device, cl_context context,
228 cl_command_queue queue)
229 {
230 return BufferAction::Setup(device, context, queue);
231 }
232
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)233 cl_int ReadBufferAction::Execute(cl_command_queue queue, cl_uint numWaits,
234 cl_event *waits, cl_event *outEvent)
235 {
236 cl_int error = clEnqueueReadBuffer(queue, mBuffer, CL_FALSE, 0, mSize,
237 mOutBuffer, numWaits, waits, outEvent);
238 test_error(error, "Unable to enqueue buffer read");
239
240 return CL_SUCCESS;
241 }
242
Setup(cl_device_id device,cl_context context,cl_command_queue queue)243 cl_int WriteBufferAction::Setup(cl_device_id device, cl_context context,
244 cl_command_queue queue)
245 {
246 return BufferAction::Setup(device, context, queue);
247 }
248
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)249 cl_int WriteBufferAction::Execute(cl_command_queue queue, cl_uint numWaits,
250 cl_event *waits, cl_event *outEvent)
251 {
252 cl_int error = clEnqueueWriteBuffer(queue, mBuffer, CL_FALSE, 0, mSize,
253 mOutBuffer, numWaits, waits, outEvent);
254 test_error(error, "Unable to enqueue buffer write");
255
256 return CL_SUCCESS;
257 }
258
~MapBufferAction()259 MapBufferAction::~MapBufferAction()
260 {
261 if (mQueue)
262 clEnqueueUnmapMemObject(mQueue, mBuffer, mMappedPtr, 0, NULL, NULL);
263 }
264
Setup(cl_device_id device,cl_context context,cl_command_queue queue)265 cl_int MapBufferAction::Setup(cl_device_id device, cl_context context,
266 cl_command_queue queue)
267 {
268 return BufferAction::Setup(device, context, queue);
269 }
270
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)271 cl_int MapBufferAction::Execute(cl_command_queue queue, cl_uint numWaits,
272 cl_event *waits, cl_event *outEvent)
273 {
274 cl_int error;
275 mQueue = queue;
276 mMappedPtr = clEnqueueMapBuffer(queue, mBuffer, CL_FALSE, CL_MAP_READ, 0,
277 mSize, numWaits, waits, outEvent, &error);
278 test_error(error, "Unable to enqueue buffer map");
279
280 return CL_SUCCESS;
281 }
282
Setup(cl_device_id device,cl_context context,cl_command_queue queue)283 cl_int UnmapBufferAction::Setup(cl_device_id device, cl_context context,
284 cl_command_queue queue)
285 {
286 cl_int error = BufferAction::Setup(device, context, queue);
287 if (error != CL_SUCCESS) return error;
288
289 mMappedPtr = clEnqueueMapBuffer(queue, mBuffer, CL_TRUE, CL_MAP_READ, 0,
290 mSize, 0, NULL, NULL, &error);
291 test_error(error, "Unable to enqueue buffer map");
292
293 return CL_SUCCESS;
294 }
295
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)296 cl_int UnmapBufferAction::Execute(cl_command_queue queue, cl_uint numWaits,
297 cl_event *waits, cl_event *outEvent)
298 {
299 cl_int error = clEnqueueUnmapMemObject(queue, mBuffer, mMappedPtr, numWaits,
300 waits, outEvent);
301 test_error(error, "Unable to enqueue buffer unmap");
302
303 return CL_SUCCESS;
304 }
305
306
307 #pragma mark -------------------- Read/Write Image Classes -------------------------
308
Setup(cl_device_id device,cl_context context,cl_command_queue queue)309 cl_int ReadImage2DAction::Setup(cl_device_id device, cl_context context,
310 cl_command_queue queue)
311 {
312 cl_int error;
313
314
315 if ((error = IGetPreferredImageSize2D(device, mWidth, mHeight)))
316 return error;
317
318 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
319 mImage = create_image_2d(context, CL_MEM_READ_ONLY, &format, mWidth,
320 mHeight, 0, NULL, &error);
321
322 test_error(error, "Unable to create image to test against");
323
324 mOutput = malloc(mWidth * mHeight * 4);
325 if (mOutput == NULL)
326 {
327 log_error("ERROR: Unable to allocate buffer: out of memory\n");
328 return CL_OUT_OF_RESOURCES;
329 }
330
331 return CL_SUCCESS;
332 }
333
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)334 cl_int ReadImage2DAction::Execute(cl_command_queue queue, cl_uint numWaits,
335 cl_event *waits, cl_event *outEvent)
336 {
337 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, 1 };
338
339 cl_int error = clEnqueueReadImage(queue, mImage, CL_FALSE, origin, region,
340 0, 0, mOutput, numWaits, waits, outEvent);
341 test_error(error, "Unable to enqueue image read");
342
343 return CL_SUCCESS;
344 }
345
Setup(cl_device_id device,cl_context context,cl_command_queue queue)346 cl_int ReadImage3DAction::Setup(cl_device_id device, cl_context context,
347 cl_command_queue queue)
348 {
349 cl_int error;
350
351
352 if ((error = IGetPreferredImageSize3D(device, mWidth, mHeight, mDepth)))
353 return error;
354
355 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
356 mImage = create_image_3d(context, CL_MEM_READ_ONLY, &format, mWidth,
357 mHeight, mDepth, 0, 0, NULL, &error);
358 test_error(error, "Unable to create image to test against");
359
360 mOutput = malloc(mWidth * mHeight * mDepth * 4);
361 if (mOutput == NULL)
362 {
363 log_error("ERROR: Unable to allocate buffer: out of memory\n");
364 return CL_OUT_OF_RESOURCES;
365 }
366
367 return CL_SUCCESS;
368 }
369
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)370 cl_int ReadImage3DAction::Execute(cl_command_queue queue, cl_uint numWaits,
371 cl_event *waits, cl_event *outEvent)
372 {
373 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, mDepth };
374
375 cl_int error = clEnqueueReadImage(queue, mImage, CL_FALSE, origin, region,
376 0, 0, mOutput, numWaits, waits, outEvent);
377 test_error(error, "Unable to enqueue image read");
378
379 return CL_SUCCESS;
380 }
381
Setup(cl_device_id device,cl_context context,cl_command_queue queue)382 cl_int WriteImage2DAction::Setup(cl_device_id device, cl_context context,
383 cl_command_queue queue)
384 {
385 cl_int error;
386
387
388 if ((error = IGetPreferredImageSize2D(device, mWidth, mHeight)))
389 return error;
390
391 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
392 mImage = create_image_2d(context, CL_MEM_WRITE_ONLY, &format, mWidth,
393 mHeight, 0, NULL, &error);
394 test_error(error, "Unable to create image to test against");
395
396 mOutput = malloc(mWidth * mHeight * 4);
397 if (mOutput == NULL)
398 {
399 log_error("ERROR: Unable to allocate buffer: out of memory\n");
400 return CL_OUT_OF_RESOURCES;
401 }
402
403 return CL_SUCCESS;
404 }
405
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)406 cl_int WriteImage2DAction::Execute(cl_command_queue queue, cl_uint numWaits,
407 cl_event *waits, cl_event *outEvent)
408 {
409 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, 1 };
410
411 cl_int error =
412 clEnqueueWriteImage(queue, mImage, CL_FALSE, origin, region, 0, 0,
413 mOutput, numWaits, waits, outEvent);
414 test_error(error, "Unable to enqueue image write");
415
416 return CL_SUCCESS;
417 }
418
Setup(cl_device_id device,cl_context context,cl_command_queue queue)419 cl_int WriteImage3DAction::Setup(cl_device_id device, cl_context context,
420 cl_command_queue queue)
421 {
422 cl_int error;
423
424
425 if ((error = IGetPreferredImageSize3D(device, mWidth, mHeight, mDepth)))
426 return error;
427
428 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
429 mImage = create_image_3d(context, CL_MEM_READ_ONLY, &format, mWidth,
430 mHeight, mDepth, 0, 0, NULL, &error);
431 test_error(error, "Unable to create image to test against");
432
433 mOutput = malloc(mWidth * mHeight * mDepth * 4);
434 if (mOutput == NULL)
435 {
436 log_error("ERROR: Unable to allocate buffer: out of memory\n");
437 return CL_OUT_OF_RESOURCES;
438 }
439
440 return CL_SUCCESS;
441 }
442
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)443 cl_int WriteImage3DAction::Execute(cl_command_queue queue, cl_uint numWaits,
444 cl_event *waits, cl_event *outEvent)
445 {
446 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, mDepth };
447
448 cl_int error =
449 clEnqueueWriteImage(queue, mImage, CL_FALSE, origin, region, 0, 0,
450 mOutput, numWaits, waits, outEvent);
451 test_error(error, "Unable to enqueue image write");
452
453 return CL_SUCCESS;
454 }
455
456 #pragma mark -------------------- Copy Image Classes -------------------------
457
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)458 cl_int CopyImageAction::Execute(cl_command_queue queue, cl_uint numWaits,
459 cl_event *waits, cl_event *outEvent)
460 {
461 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, mDepth };
462
463 cl_int error =
464 clEnqueueCopyImage(queue, mSrcImage, mDstImage, origin, origin, region,
465 numWaits, waits, outEvent);
466 test_error(error, "Unable to enqueue image copy");
467
468 return CL_SUCCESS;
469 }
470
Setup(cl_device_id device,cl_context context,cl_command_queue queue)471 cl_int CopyImage2Dto2DAction::Setup(cl_device_id device, cl_context context,
472 cl_command_queue queue)
473 {
474 cl_int error;
475
476
477 if ((error = IGetPreferredImageSize2D(device, mWidth, mHeight)))
478 return error;
479
480 mWidth /= 2;
481
482 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
483 mSrcImage = create_image_2d(context, CL_MEM_READ_ONLY, &format, mWidth,
484 mHeight, 0, NULL, &error);
485 test_error(error, "Unable to create image to test against");
486
487 mDstImage = create_image_2d(context, CL_MEM_WRITE_ONLY, &format, mWidth,
488 mHeight, 0, NULL, &error);
489 test_error(error, "Unable to create image to test against");
490
491 mDepth = 1;
492 return CL_SUCCESS;
493 }
494
Setup(cl_device_id device,cl_context context,cl_command_queue queue)495 cl_int CopyImage2Dto3DAction::Setup(cl_device_id device, cl_context context,
496 cl_command_queue queue)
497 {
498 cl_int error;
499
500
501 if ((error = IGetPreferredImageSize3D(device, mWidth, mHeight, mDepth)))
502 return error;
503
504 mDepth /= 2;
505
506 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
507 mSrcImage = create_image_2d(context, CL_MEM_READ_ONLY, &format, mWidth,
508 mHeight, 0, NULL, &error);
509 test_error(error, "Unable to create image to test against");
510
511 mDstImage = create_image_3d(context, CL_MEM_READ_ONLY, &format, mWidth,
512 mHeight, mDepth, 0, 0, NULL, &error);
513 test_error(error, "Unable to create image to test against");
514
515 mDepth = 1;
516 return CL_SUCCESS;
517 }
518
Setup(cl_device_id device,cl_context context,cl_command_queue queue)519 cl_int CopyImage3Dto2DAction::Setup(cl_device_id device, cl_context context,
520 cl_command_queue queue)
521 {
522 cl_int error;
523
524
525 if ((error = IGetPreferredImageSize3D(device, mWidth, mHeight, mDepth)))
526 return error;
527
528 mDepth /= 2;
529
530 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
531 mSrcImage = create_image_3d(context, CL_MEM_READ_ONLY, &format, mWidth,
532 mHeight, mDepth, 0, 0, NULL, &error);
533 test_error(error, "Unable to create image to test against");
534
535 mDstImage = create_image_2d(context, CL_MEM_WRITE_ONLY, &format, mWidth,
536 mHeight, 0, NULL, &error);
537 test_error(error, "Unable to create image to test against");
538
539 mDepth = 1;
540 return CL_SUCCESS;
541 }
542
Setup(cl_device_id device,cl_context context,cl_command_queue queue)543 cl_int CopyImage3Dto3DAction::Setup(cl_device_id device, cl_context context,
544 cl_command_queue queue)
545 {
546 cl_int error;
547
548
549 if ((error = IGetPreferredImageSize3D(device, mWidth, mHeight, mDepth)))
550 return error;
551
552 mDepth /= 2;
553
554 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
555 mSrcImage = create_image_3d(context, CL_MEM_READ_ONLY, &format, mWidth,
556 mHeight, mDepth, 0, 0, NULL, &error);
557 test_error(error, "Unable to create image to test against");
558
559 mDstImage = create_image_3d(context, CL_MEM_READ_ONLY, &format, mWidth,
560 mHeight, mDepth, 0, 0, NULL, &error);
561 test_error(error, "Unable to create image to test against");
562
563 return CL_SUCCESS;
564 }
565
566 #pragma mark -------------------- Copy Image/Buffer Classes -------------------------
567
Setup(cl_device_id device,cl_context context,cl_command_queue queue)568 cl_int Copy2DImageToBufferAction::Setup(cl_device_id device, cl_context context,
569 cl_command_queue queue)
570 {
571 cl_int error;
572
573
574 if ((error = IGetPreferredImageSize2D(device, mWidth, mHeight)))
575 return error;
576
577 mWidth /= 2;
578
579 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
580 mSrcImage = create_image_2d(context, CL_MEM_READ_ONLY, &format, mWidth,
581 mHeight, 0, NULL, &error);
582 test_error(error, "Unable to create image to test against");
583
584 mDstBuffer = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
585 mWidth * mHeight * 4, NULL, &error);
586 test_error(error, "Unable to create buffer to test against");
587
588 return CL_SUCCESS;
589 }
590
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)591 cl_int Copy2DImageToBufferAction::Execute(cl_command_queue queue,
592 cl_uint numWaits, cl_event *waits,
593 cl_event *outEvent)
594 {
595 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, 1 };
596
597 cl_int error =
598 clEnqueueCopyImageToBuffer(queue, mSrcImage, mDstBuffer, origin, region,
599 0, numWaits, waits, outEvent);
600 test_error(error, "Unable to enqueue image to buffer copy");
601
602 return CL_SUCCESS;
603 }
604
Setup(cl_device_id device,cl_context context,cl_command_queue queue)605 cl_int Copy3DImageToBufferAction::Setup(cl_device_id device, cl_context context,
606 cl_command_queue queue)
607 {
608 cl_int error;
609
610
611 if ((error = IGetPreferredImageSize3D(device, mWidth, mHeight, mDepth)))
612 return error;
613
614 mDepth /= 2;
615
616 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
617 mSrcImage = create_image_3d(context, CL_MEM_READ_ONLY, &format, mWidth,
618 mHeight, mDepth, 0, 0, NULL, &error);
619 test_error(error, "Unable to create image to test against");
620
621 mDstBuffer = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
622 mWidth * mHeight * mDepth * 4, NULL, &error);
623 test_error(error, "Unable to create buffer to test against");
624
625 return CL_SUCCESS;
626 }
627
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)628 cl_int Copy3DImageToBufferAction::Execute(cl_command_queue queue,
629 cl_uint numWaits, cl_event *waits,
630 cl_event *outEvent)
631 {
632 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, mDepth };
633
634 cl_int error =
635 clEnqueueCopyImageToBuffer(queue, mSrcImage, mDstBuffer, origin, region,
636 0, numWaits, waits, outEvent);
637 test_error(error, "Unable to enqueue image to buffer copy");
638
639 return CL_SUCCESS;
640 }
641
Setup(cl_device_id device,cl_context context,cl_command_queue queue)642 cl_int CopyBufferTo2DImageAction::Setup(cl_device_id device, cl_context context,
643 cl_command_queue queue)
644 {
645 cl_int error;
646
647
648 if ((error = IGetPreferredImageSize2D(device, mWidth, mHeight)))
649 return error;
650
651 mWidth /= 2;
652
653 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
654
655 mSrcBuffer = clCreateBuffer(context, CL_MEM_READ_ONLY, mWidth * mHeight * 4,
656 NULL, &error);
657 test_error(error, "Unable to create buffer to test against");
658
659 mDstImage = create_image_2d(context, CL_MEM_WRITE_ONLY, &format, mWidth,
660 mHeight, 0, NULL, &error);
661 test_error(error, "Unable to create image to test against");
662
663 return CL_SUCCESS;
664 }
665
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)666 cl_int CopyBufferTo2DImageAction::Execute(cl_command_queue queue,
667 cl_uint numWaits, cl_event *waits,
668 cl_event *outEvent)
669 {
670 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, 1 };
671
672 cl_int error =
673 clEnqueueCopyBufferToImage(queue, mSrcBuffer, mDstImage, 0, origin,
674 region, numWaits, waits, outEvent);
675 test_error(error, "Unable to enqueue buffer to image copy");
676
677 return CL_SUCCESS;
678 }
679
Setup(cl_device_id device,cl_context context,cl_command_queue queue)680 cl_int CopyBufferTo3DImageAction::Setup(cl_device_id device, cl_context context,
681 cl_command_queue queue)
682 {
683 cl_int error;
684
685
686 if ((error = IGetPreferredImageSize3D(device, mWidth, mHeight, mDepth)))
687 return error;
688
689 mDepth /= 2;
690
691 mSrcBuffer = clCreateBuffer(context, CL_MEM_READ_ONLY,
692 mWidth * mHeight * mDepth * 4, NULL, &error);
693 test_error(error, "Unable to create buffer to test against");
694
695 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
696 mDstImage = create_image_3d(context, CL_MEM_READ_ONLY, &format, mWidth,
697 mHeight, mDepth, 0, 0, NULL, &error);
698 test_error(error, "Unable to create image to test against");
699
700 return CL_SUCCESS;
701 }
702
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)703 cl_int CopyBufferTo3DImageAction::Execute(cl_command_queue queue,
704 cl_uint numWaits, cl_event *waits,
705 cl_event *outEvent)
706 {
707 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, mDepth };
708
709 cl_int error =
710 clEnqueueCopyBufferToImage(queue, mSrcBuffer, mDstImage, 0, origin,
711 region, numWaits, waits, outEvent);
712 test_error(error, "Unable to enqueue buffer to image copy");
713
714 return CL_SUCCESS;
715 }
716
717 #pragma mark -------------------- Map Image Class -------------------------
718
~MapImageAction()719 MapImageAction::~MapImageAction()
720 {
721 if (mQueue)
722 clEnqueueUnmapMemObject(mQueue, mImage, mMappedPtr, 0, NULL, NULL);
723 }
724
Setup(cl_device_id device,cl_context context,cl_command_queue queue)725 cl_int MapImageAction::Setup(cl_device_id device, cl_context context,
726 cl_command_queue queue)
727 {
728 cl_int error;
729
730
731 if ((error = IGetPreferredImageSize2D(device, mWidth, mHeight)))
732 return error;
733
734 cl_image_format format = { CL_RGBA, CL_SIGNED_INT8 };
735 mImage = create_image_2d(context, CL_MEM_READ_ONLY | CL_MEM_ALLOC_HOST_PTR,
736 &format, mWidth, mHeight, 0, NULL, &error);
737 test_error(error, "Unable to create image to test against");
738
739 return CL_SUCCESS;
740 }
741
Execute(cl_command_queue queue,cl_uint numWaits,cl_event * waits,cl_event * outEvent)742 cl_int MapImageAction::Execute(cl_command_queue queue, cl_uint numWaits,
743 cl_event *waits, cl_event *outEvent)
744 {
745 cl_int error;
746
747 size_t origin[3] = { 0, 0, 0 }, region[3] = { mWidth, mHeight, 1 };
748 size_t outPitch;
749
750 mQueue = queue;
751 mMappedPtr =
752 clEnqueueMapImage(queue, mImage, CL_FALSE, CL_MAP_READ, origin, region,
753 &outPitch, NULL, numWaits, waits, outEvent, &error);
754 test_error(error, "Unable to enqueue image map");
755
756 return CL_SUCCESS;
757 }
758