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