1 /*
2 * Copyright (C) 2018-2021 Intel Corporation
3 *
4 * SPDX-License-Identifier: MIT
5 *
6 */
7
8 #include "api.h"
9
10 #include "shared/source/aub/aub_center.h"
11 #include "shared/source/built_ins/built_ins.h"
12 #include "shared/source/command_stream/command_stream_receiver.h"
13 #include "shared/source/debug_settings/debug_settings_manager.h"
14 #include "shared/source/execution_environment/root_device_environment.h"
15 #include "shared/source/helpers/aligned_memory.h"
16 #include "shared/source/helpers/get_info.h"
17 #include "shared/source/helpers/hw_info.h"
18 #include "shared/source/helpers/kernel_helpers.h"
19 #include "shared/source/memory_manager/unified_memory_manager.h"
20 #include "shared/source/os_interface/device_factory.h"
21 #include "shared/source/os_interface/os_context.h"
22 #include "shared/source/utilities/api_intercept.h"
23 #include "shared/source/utilities/stackvec.h"
24
25 #include "opencl/source/accelerators/intel_motion_estimation.h"
26 #include "opencl/source/api/additional_extensions.h"
27 #include "opencl/source/built_ins/vme_builtin.h"
28 #include "opencl/source/cl_device/cl_device.h"
29 #include "opencl/source/command_queue/command_queue.h"
30 #include "opencl/source/context/context.h"
31 #include "opencl/source/context/driver_diagnostics.h"
32 #include "opencl/source/device_queue/device_queue.h"
33 #include "opencl/source/event/user_event.h"
34 #include "opencl/source/execution_environment/cl_execution_environment.h"
35 #include "opencl/source/gtpin/gtpin_notify.h"
36 #include "opencl/source/helpers/cl_memory_properties_helpers.h"
37 #include "opencl/source/helpers/cl_validators.h"
38 #include "opencl/source/helpers/get_info_status_mapper.h"
39 #include "opencl/source/helpers/queue_helpers.h"
40 #include "opencl/source/kernel/kernel.h"
41 #include "opencl/source/kernel/kernel_info_cl.h"
42 #include "opencl/source/kernel/multi_device_kernel.h"
43 #include "opencl/source/mem_obj/buffer.h"
44 #include "opencl/source/mem_obj/image.h"
45 #include "opencl/source/mem_obj/mem_obj_helper.h"
46 #include "opencl/source/mem_obj/pipe.h"
47 #include "opencl/source/platform/platform.h"
48 #include "opencl/source/program/program.h"
49 #include "opencl/source/sampler/sampler.h"
50 #include "opencl/source/sharings/sharing_factory.h"
51 #include "opencl/source/tracing/tracing_api.h"
52 #include "opencl/source/tracing/tracing_notify.h"
53 #include "opencl/source/utilities/cl_logger.h"
54
55 #include "CL/cl.h"
56 #include "config.h"
57
58 #include <algorithm>
59 #include <cstring>
60
61 using namespace NEO;
62
clGetPlatformIDs(cl_uint numEntries,cl_platform_id * platforms,cl_uint * numPlatforms)63 cl_int CL_API_CALL clGetPlatformIDs(cl_uint numEntries,
64 cl_platform_id *platforms,
65 cl_uint *numPlatforms) {
66 TRACING_ENTER(clGetPlatformIDs, &numEntries, &platforms, &numPlatforms);
67 cl_int retVal = CL_SUCCESS;
68 API_ENTER(&retVal);
69 DBG_LOG_INPUTS("numEntries", numEntries,
70 "platforms", platforms,
71 "numPlatforms", numPlatforms);
72
73 do {
74 // if platforms is nullptr, we must return the number of valid platforms we
75 // support in the num_platforms variable (if it is non-nullptr)
76 if ((platforms == nullptr) && (numPlatforms == nullptr)) {
77 retVal = CL_INVALID_VALUE;
78 break;
79 }
80
81 // platform != nullptr and num_entries == 0 is defined by spec as invalid
82 if (numEntries == 0 && platforms != nullptr) {
83 retVal = CL_INVALID_VALUE;
84 break;
85 }
86
87 static std::mutex mutex;
88 std::unique_lock<std::mutex> lock(mutex);
89 if (platformsImpl->empty()) {
90 auto executionEnvironment = new ClExecutionEnvironment();
91 executionEnvironment->incRefInternal();
92 auto allDevices = DeviceFactory::createDevices(*executionEnvironment);
93 executionEnvironment->decRefInternal();
94 if (allDevices.empty()) {
95 retVal = CL_OUT_OF_HOST_MEMORY;
96 break;
97 }
98 auto groupedDevices = Platform::groupDevices(std::move(allDevices));
99 for (auto &deviceVector : groupedDevices) {
100
101 auto pPlatform = Platform::createFunc(*executionEnvironment);
102 if (!pPlatform || !pPlatform->initialize(std::move(deviceVector))) {
103 retVal = CL_OUT_OF_HOST_MEMORY;
104 break;
105 }
106 platformsImpl->push_back(std::move(pPlatform));
107 }
108 if (retVal != CL_SUCCESS) {
109 break;
110 }
111 }
112 cl_uint numPlatformsToExpose = std::min(numEntries, static_cast<cl_uint>(platformsImpl->size()));
113 if (numEntries == 0) {
114 numPlatformsToExpose = static_cast<cl_uint>(platformsImpl->size());
115 }
116 if (platforms) {
117 for (auto i = 0u; i < numPlatformsToExpose; i++) {
118 platforms[i] = (*platformsImpl)[i].get();
119 }
120 }
121
122 if (numPlatforms) {
123 *numPlatforms = numPlatformsToExpose;
124 }
125 } while (false);
126 TRACING_EXIT(clGetPlatformIDs, &retVal);
127 return retVal;
128 }
129
clIcdGetPlatformIDsKHR(cl_uint numEntries,cl_platform_id * platforms,cl_uint * numPlatforms)130 CL_API_ENTRY cl_int CL_API_CALL clIcdGetPlatformIDsKHR(cl_uint numEntries,
131 cl_platform_id *platforms,
132 cl_uint *numPlatforms) {
133 cl_int retVal = CL_SUCCESS;
134 API_ENTER(&retVal);
135 DBG_LOG_INPUTS("numEntries", numEntries,
136 "platforms", platforms,
137 "numPlatforms", numPlatforms);
138 retVal = clGetPlatformIDs(numEntries, platforms, numPlatforms);
139 return retVal;
140 }
141
clGetPlatformInfo(cl_platform_id platform,cl_platform_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)142 cl_int CL_API_CALL clGetPlatformInfo(cl_platform_id platform,
143 cl_platform_info paramName,
144 size_t paramValueSize,
145 void *paramValue,
146 size_t *paramValueSizeRet) {
147 TRACING_ENTER(clGetPlatformInfo, &platform, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
148 cl_int retVal = CL_INVALID_PLATFORM;
149 API_ENTER(&retVal);
150 DBG_LOG_INPUTS("platform", platform,
151 "paramName", paramName,
152 "paramValueSize", paramValueSize,
153 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
154 "paramValueSizeRet", paramValueSizeRet);
155 auto pPlatform = castToObject<Platform>(platform);
156 if (pPlatform) {
157 retVal = pPlatform->getInfo(paramName, paramValueSize,
158 paramValue, paramValueSizeRet);
159 }
160 TRACING_EXIT(clGetPlatformInfo, &retVal);
161 return retVal;
162 }
163
clGetDeviceIDs(cl_platform_id platform,cl_device_type deviceType,cl_uint numEntries,cl_device_id * devices,cl_uint * numDevices)164 cl_int CL_API_CALL clGetDeviceIDs(cl_platform_id platform,
165 cl_device_type deviceType,
166 cl_uint numEntries,
167 cl_device_id *devices,
168 cl_uint *numDevices) {
169 TRACING_ENTER(clGetDeviceIDs, &platform, &deviceType, &numEntries, &devices, &numDevices);
170 cl_int retVal = CL_SUCCESS;
171 API_ENTER(&retVal);
172 DBG_LOG_INPUTS("platform", platform,
173 "deviceType", deviceType,
174 "numEntries", numEntries,
175 "devices", devices,
176 "numDevices", numDevices);
177 const cl_device_type validType = CL_DEVICE_TYPE_GPU | CL_DEVICE_TYPE_CPU |
178 CL_DEVICE_TYPE_ACCELERATOR | CL_DEVICE_TYPE_DEFAULT |
179 CL_DEVICE_TYPE_CUSTOM;
180 Platform *pPlatform = nullptr;
181
182 do {
183 /* Check parameter consistency */
184 if (devices == nullptr && numDevices == nullptr) {
185 retVal = CL_INVALID_VALUE;
186 break;
187 }
188
189 if (devices && numEntries == 0) {
190 retVal = CL_INVALID_VALUE;
191 break;
192 }
193
194 if ((deviceType & validType) == 0) {
195 retVal = CL_INVALID_DEVICE_TYPE;
196 break;
197 }
198
199 if (platform != nullptr) {
200 pPlatform = castToObject<Platform>(platform);
201 if (pPlatform == nullptr) {
202 retVal = CL_INVALID_PLATFORM;
203 break;
204 }
205 } else {
206 cl_uint numPlatforms = 0u;
207 retVal = clGetPlatformIDs(0, nullptr, &numPlatforms);
208 if (numPlatforms == 0u) {
209 retVal = CL_DEVICE_NOT_FOUND;
210 break;
211 }
212 pPlatform = (*platformsImpl)[0].get();
213 }
214
215 DEBUG_BREAK_IF(pPlatform->isInitialized() != true);
216 cl_uint numDev = static_cast<cl_uint>(pPlatform->getNumDevices());
217 if (numDev == 0) {
218 retVal = CL_DEVICE_NOT_FOUND;
219 break;
220 }
221
222 if (DebugManager.flags.LimitAmountOfReturnedDevices.get()) {
223 numDev = std::min(static_cast<cl_uint>(DebugManager.flags.LimitAmountOfReturnedDevices.get()), numDev);
224 }
225
226 if (deviceType == CL_DEVICE_TYPE_ALL) {
227 /* According to Spec, set it to all except TYPE_CUSTOM. */
228 deviceType = CL_DEVICE_TYPE_GPU | CL_DEVICE_TYPE_CPU |
229 CL_DEVICE_TYPE_ACCELERATOR | CL_DEVICE_TYPE_DEFAULT;
230 } else if (deviceType == CL_DEVICE_TYPE_DEFAULT) {
231 /* We just set it to GPU now. */
232 deviceType = CL_DEVICE_TYPE_GPU;
233 }
234
235 cl_uint retNum = 0;
236 for (auto platformDeviceIndex = 0u; platformDeviceIndex < numDev; platformDeviceIndex++) {
237
238 ClDevice *device = pPlatform->getClDevice(platformDeviceIndex);
239 UNRECOVERABLE_IF(device == nullptr);
240
241 if (deviceType & device->getDeviceInfo().deviceType) {
242 if (devices) {
243 if (retNum >= numEntries) {
244 break;
245 }
246 devices[retNum] = device;
247 }
248 retNum++;
249 }
250 }
251
252 if (numDevices) {
253 *numDevices = retNum;
254 }
255
256 /* If no suitable device, set a error. */
257 if (retNum == 0)
258 retVal = CL_DEVICE_NOT_FOUND;
259 } while (false);
260 TRACING_EXIT(clGetDeviceIDs, &retVal);
261 return retVal;
262 }
263
clGetDeviceInfo(cl_device_id device,cl_device_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)264 cl_int CL_API_CALL clGetDeviceInfo(cl_device_id device,
265 cl_device_info paramName,
266 size_t paramValueSize,
267 void *paramValue,
268 size_t *paramValueSizeRet) {
269 TRACING_ENTER(clGetDeviceInfo, &device, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
270 cl_int retVal = CL_INVALID_DEVICE;
271 API_ENTER(&retVal);
272 DBG_LOG_INPUTS("clDevice", device, "paramName", paramName, "paramValueSize", paramValueSize, "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize), "paramValueSizeRet", paramValueSizeRet);
273
274 ClDevice *pDevice = castToObject<ClDevice>(device);
275 if (pDevice != nullptr) {
276 retVal = pDevice->getDeviceInfo(paramName, paramValueSize,
277 paramValue, paramValueSizeRet);
278 }
279 TRACING_EXIT(clGetDeviceInfo, &retVal);
280 return retVal;
281 }
282
clCreateSubDevices(cl_device_id inDevice,const cl_device_partition_property * properties,cl_uint numDevices,cl_device_id * outDevices,cl_uint * numDevicesRet)283 cl_int CL_API_CALL clCreateSubDevices(cl_device_id inDevice,
284 const cl_device_partition_property *properties,
285 cl_uint numDevices,
286 cl_device_id *outDevices,
287 cl_uint *numDevicesRet) {
288
289 ClDevice *pInDevice = castToObject<ClDevice>(inDevice);
290 if (pInDevice == nullptr) {
291 return CL_INVALID_DEVICE;
292 }
293 auto subDevicesCount = pInDevice->getNumSubDevices();
294 if (subDevicesCount <= 1) {
295 return CL_DEVICE_PARTITION_FAILED;
296 }
297 if ((properties == nullptr) ||
298 (properties[0] != CL_DEVICE_PARTITION_BY_AFFINITY_DOMAIN) ||
299 ((properties[1] != CL_DEVICE_AFFINITY_DOMAIN_NUMA) && (properties[1] != CL_DEVICE_AFFINITY_DOMAIN_NEXT_PARTITIONABLE)) ||
300 (properties[2] != 0)) {
301 return CL_INVALID_VALUE;
302 }
303
304 if (numDevicesRet != nullptr) {
305 *numDevicesRet = subDevicesCount;
306 }
307
308 if (outDevices == nullptr) {
309 return CL_SUCCESS;
310 }
311
312 if (numDevices < subDevicesCount) {
313 return CL_INVALID_VALUE;
314 }
315
316 for (uint32_t i = 0; i < subDevicesCount; i++) {
317 auto pClDevice = pInDevice->getSubDevice(i);
318 pClDevice->retainApi();
319 outDevices[i] = pClDevice;
320 }
321
322 return CL_SUCCESS;
323 }
324
clRetainDevice(cl_device_id device)325 cl_int CL_API_CALL clRetainDevice(cl_device_id device) {
326 TRACING_ENTER(clRetainDevice, &device);
327 cl_int retVal = CL_INVALID_DEVICE;
328 API_ENTER(&retVal);
329 DBG_LOG_INPUTS("device", device);
330 auto pDevice = castToObject<ClDevice>(device);
331 if (pDevice) {
332 pDevice->retainApi();
333 retVal = CL_SUCCESS;
334 }
335
336 TRACING_EXIT(clRetainDevice, &retVal);
337 return retVal;
338 }
339
clReleaseDevice(cl_device_id device)340 cl_int CL_API_CALL clReleaseDevice(cl_device_id device) {
341 TRACING_ENTER(clReleaseDevice, &device);
342 cl_int retVal = CL_INVALID_DEVICE;
343 API_ENTER(&retVal);
344 DBG_LOG_INPUTS("device", device);
345 auto pDevice = castToObject<ClDevice>(device);
346 if (pDevice) {
347 pDevice->releaseApi();
348 retVal = CL_SUCCESS;
349 }
350
351 TRACING_EXIT(clReleaseDevice, &retVal);
352 return retVal;
353 }
354
clCreateContext(const cl_context_properties * properties,cl_uint numDevices,const cl_device_id * devices,void (CL_CALLBACK * funcNotify)(const char *,const void *,size_t,void *),void * userData,cl_int * errcodeRet)355 cl_context CL_API_CALL clCreateContext(const cl_context_properties *properties,
356 cl_uint numDevices,
357 const cl_device_id *devices,
358 void(CL_CALLBACK *funcNotify)(const char *, const void *,
359 size_t, void *),
360 void *userData,
361 cl_int *errcodeRet) {
362 TRACING_ENTER(clCreateContext, &properties, &numDevices, &devices, &funcNotify, &userData, &errcodeRet);
363
364 cl_int retVal = CL_SUCCESS;
365 cl_context context = nullptr;
366 API_ENTER(&retVal);
367 DBG_LOG_INPUTS("properties", properties, "numDevices", numDevices, "cl_device_id", devices, "funcNotify", funcNotify, "userData", userData);
368
369 do {
370 if (devices == nullptr) {
371 /* Must have device. */
372 retVal = CL_INVALID_VALUE;
373 break;
374 }
375
376 /* validateObjects make sure numDevices != 0. */
377 retVal = validateObjects(DeviceList(numDevices, devices));
378 if (retVal != CL_SUCCESS)
379 break;
380
381 if (funcNotify == nullptr && userData != nullptr) {
382 retVal = CL_INVALID_VALUE;
383 break;
384 }
385 auto pPlatform = Context::getPlatformFromProperties(properties, retVal);
386 if (CL_SUCCESS != retVal) {
387 break;
388 }
389
390 ClDeviceVector allDevs(devices, numDevices);
391 if (!pPlatform) {
392 pPlatform = allDevs[0]->getPlatform();
393 }
394 for (auto &pClDevice : allDevs) {
395 if (pClDevice->getPlatform() != pPlatform) {
396 retVal = CL_INVALID_DEVICE;
397 break;
398 }
399 }
400 if (CL_SUCCESS != retVal) {
401 break;
402 }
403 context = Context::create<Context>(properties, allDevs, funcNotify, userData, retVal);
404 } while (false);
405
406 if (errcodeRet) {
407 *errcodeRet = retVal;
408 }
409 TRACING_EXIT(clCreateContext, &context);
410 return context;
411 }
412
clCreateContextFromType(const cl_context_properties * properties,cl_device_type deviceType,void (CL_CALLBACK * funcNotify)(const char *,const void *,size_t,void *),void * userData,cl_int * errcodeRet)413 cl_context CL_API_CALL clCreateContextFromType(const cl_context_properties *properties,
414 cl_device_type deviceType,
415 void(CL_CALLBACK *funcNotify)(const char *, const void *,
416 size_t, void *),
417 void *userData,
418 cl_int *errcodeRet) {
419 TRACING_ENTER(clCreateContextFromType, &properties, &deviceType, &funcNotify, &userData, &errcodeRet);
420 cl_int retVal = CL_SUCCESS;
421 API_ENTER(&retVal);
422 DBG_LOG_INPUTS("properties", properties, "deviceType", deviceType, "funcNotify", funcNotify, "userData", userData);
423 Context *pContext = nullptr;
424
425 do {
426 if (funcNotify == nullptr && userData != nullptr) {
427 retVal = CL_INVALID_VALUE;
428 break;
429 }
430 auto pPlatform = Context::getPlatformFromProperties(properties, retVal);
431 if (CL_SUCCESS != retVal) {
432 break;
433 }
434 cl_uint numDevices = 0;
435 /* Query the number of device first. */
436 retVal = clGetDeviceIDs(pPlatform, deviceType, 0, nullptr, &numDevices);
437 if (retVal != CL_SUCCESS) {
438 break;
439 }
440
441 DEBUG_BREAK_IF(numDevices <= 0);
442 cl_device_id device = nullptr;
443
444 retVal = clGetDeviceIDs(pPlatform, deviceType, 1, &device, nullptr);
445 DEBUG_BREAK_IF(retVal != CL_SUCCESS);
446
447 ClDeviceVector deviceVector(&device, 1);
448 pContext = Context::create<Context>(properties, deviceVector, funcNotify, userData, retVal);
449 } while (false);
450
451 if (errcodeRet) {
452 *errcodeRet = retVal;
453 }
454 TRACING_EXIT(clCreateContextFromType, (cl_context *)&pContext);
455 return pContext;
456 }
457
clRetainContext(cl_context context)458 cl_int CL_API_CALL clRetainContext(cl_context context) {
459 TRACING_ENTER(clRetainContext, &context);
460 cl_int retVal = CL_SUCCESS;
461 API_ENTER(&retVal);
462 DBG_LOG_INPUTS("context", context);
463 Context *pContext = castToObject<Context>(context);
464 if (pContext) {
465 pContext->retain();
466 TRACING_EXIT(clRetainContext, &retVal);
467 return retVal;
468 }
469 retVal = CL_INVALID_CONTEXT;
470 TRACING_EXIT(clRetainContext, &retVal);
471 return retVal;
472 }
473
clReleaseContext(cl_context context)474 cl_int CL_API_CALL clReleaseContext(cl_context context) {
475 TRACING_ENTER(clReleaseContext, &context);
476 cl_int retVal = CL_SUCCESS;
477 API_ENTER(&retVal);
478 DBG_LOG_INPUTS("context", context);
479 Context *pContext = castToObject<Context>(context);
480 if (pContext) {
481 pContext->release();
482 TRACING_EXIT(clReleaseContext, &retVal);
483 return retVal;
484 }
485 retVal = CL_INVALID_CONTEXT;
486 TRACING_EXIT(clReleaseContext, &retVal);
487 return retVal;
488 }
489
clGetContextInfo(cl_context context,cl_context_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)490 cl_int CL_API_CALL clGetContextInfo(cl_context context,
491 cl_context_info paramName,
492 size_t paramValueSize,
493 void *paramValue,
494 size_t *paramValueSizeRet) {
495 TRACING_ENTER(clGetContextInfo, &context, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
496 auto retVal = CL_INVALID_CONTEXT;
497 API_ENTER(&retVal);
498 DBG_LOG_INPUTS("context", context,
499 "paramName", paramName,
500 "paramValueSize", paramValueSize,
501 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
502 "paramValueSizeRet", paramValueSizeRet);
503 auto pContext = castToObject<Context>(context);
504
505 if (pContext) {
506 retVal = pContext->getInfo(paramName, paramValueSize,
507 paramValue, paramValueSizeRet);
508 }
509
510 TRACING_EXIT(clGetContextInfo, &retVal);
511 return retVal;
512 }
513
clCreateCommandQueue(cl_context context,cl_device_id device,const cl_command_queue_properties properties,cl_int * errcodeRet)514 cl_command_queue CL_API_CALL clCreateCommandQueue(cl_context context,
515 cl_device_id device,
516 const cl_command_queue_properties properties,
517 cl_int *errcodeRet) {
518 TRACING_ENTER(clCreateCommandQueue, &context, &device, (cl_command_queue_properties *)&properties, &errcodeRet);
519 cl_command_queue commandQueue = nullptr;
520 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
521 cl_int retVal = CL_SUCCESS;
522 API_ENTER(&retVal);
523 DBG_LOG_INPUTS("context", context,
524 "device", device,
525 "properties", properties);
526
527 do {
528 if (properties &
529 ~(CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE | CL_QUEUE_PROFILING_ENABLE)) {
530 retVal = CL_INVALID_VALUE;
531 break;
532 }
533
534 Context *pContext = nullptr;
535 ClDevice *pDevice = nullptr;
536
537 retVal = validateObjects(
538 WithCastToInternal(context, &pContext),
539 WithCastToInternal(device, &pDevice));
540
541 if (retVal != CL_SUCCESS) {
542 break;
543 }
544 if (!pContext->isDeviceAssociated(*pDevice)) {
545 retVal = CL_INVALID_DEVICE;
546 break;
547 }
548
549 cl_queue_properties props[] = {
550 CL_QUEUE_PROPERTIES, properties,
551 0};
552
553 commandQueue = CommandQueue::create(pContext,
554 pDevice,
555 props,
556 false,
557 retVal);
558
559 if (pContext->isProvidingPerformanceHints()) {
560 pContext->providePerformanceHint(CL_CONTEXT_DIAGNOSTICS_LEVEL_NEUTRAL_INTEL, DRIVER_CALLS_INTERNAL_CL_FLUSH);
561 if (castToObjectOrAbort<CommandQueue>(commandQueue)->isProfilingEnabled()) {
562 pContext->providePerformanceHint(CL_CONTEXT_DIAGNOSTICS_LEVEL_NEUTRAL_INTEL, PROFILING_ENABLED);
563 if (pDevice->getDeviceInfo().preemptionSupported && pDevice->getHardwareInfo().platform.eProductFamily < IGFX_SKYLAKE) {
564 pContext->providePerformanceHint(CL_CONTEXT_DIAGNOSTICS_LEVEL_NEUTRAL_INTEL, PROFILING_ENABLED_WITH_DISABLED_PREEMPTION);
565 }
566 }
567 }
568 } while (false);
569
570 err.set(retVal);
571 DBG_LOG_INPUTS("commandQueue", commandQueue);
572 TRACING_EXIT(clCreateCommandQueue, &commandQueue);
573 return commandQueue;
574 }
575
clRetainCommandQueue(cl_command_queue commandQueue)576 cl_int CL_API_CALL clRetainCommandQueue(cl_command_queue commandQueue) {
577 TRACING_ENTER(clRetainCommandQueue, &commandQueue);
578 cl_int retVal = CL_INVALID_COMMAND_QUEUE;
579 API_ENTER(&retVal);
580 DBG_LOG_INPUTS("commandQueue", commandQueue);
581 retainQueue<CommandQueue>(commandQueue, retVal);
582 if (retVal == CL_SUCCESS) {
583 TRACING_EXIT(clRetainCommandQueue, &retVal);
584 return retVal;
585 }
586 // if host queue not found - try to query device queue
587 retainQueue<DeviceQueue>(commandQueue, retVal);
588
589 TRACING_EXIT(clRetainCommandQueue, &retVal);
590 return retVal;
591 }
592
clReleaseCommandQueue(cl_command_queue commandQueue)593 cl_int CL_API_CALL clReleaseCommandQueue(cl_command_queue commandQueue) {
594 TRACING_ENTER(clReleaseCommandQueue, &commandQueue);
595 cl_int retVal = CL_INVALID_COMMAND_QUEUE;
596 API_ENTER(&retVal);
597 DBG_LOG_INPUTS("commandQueue", commandQueue);
598
599 releaseQueue<CommandQueue>(commandQueue, retVal);
600 if (retVal == CL_SUCCESS) {
601 TRACING_EXIT(clReleaseCommandQueue, &retVal);
602 return retVal;
603 }
604 // if host queue not found - try to query device queue
605 releaseQueue<DeviceQueue>(commandQueue, retVal);
606
607 TRACING_EXIT(clReleaseCommandQueue, &retVal);
608 return retVal;
609 }
610
clGetCommandQueueInfo(cl_command_queue commandQueue,cl_command_queue_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)611 cl_int CL_API_CALL clGetCommandQueueInfo(cl_command_queue commandQueue,
612 cl_command_queue_info paramName,
613 size_t paramValueSize,
614 void *paramValue,
615 size_t *paramValueSizeRet) {
616 TRACING_ENTER(clGetCommandQueueInfo, &commandQueue, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
617 cl_int retVal = CL_INVALID_COMMAND_QUEUE;
618 API_ENTER(&retVal);
619 DBG_LOG_INPUTS("commandQueue", commandQueue,
620 "paramName", paramName,
621 "paramValueSize", paramValueSize,
622 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
623 "paramValueSizeRet", paramValueSizeRet);
624
625 getQueueInfo<CommandQueue>(commandQueue, paramName, paramValueSize, paramValue, paramValueSizeRet, retVal);
626 // if host queue not found - try to query device queue
627 if (retVal == CL_SUCCESS) {
628 TRACING_EXIT(clGetCommandQueueInfo, &retVal);
629 return retVal;
630 }
631 getQueueInfo<DeviceQueue>(commandQueue, paramName, paramValueSize, paramValue, paramValueSizeRet, retVal);
632
633 TRACING_EXIT(clGetCommandQueueInfo, &retVal);
634 return retVal;
635 }
636
637 // deprecated OpenCL 1.0
clSetCommandQueueProperty(cl_command_queue commandQueue,cl_command_queue_properties properties,cl_bool enable,cl_command_queue_properties * oldProperties)638 cl_int CL_API_CALL clSetCommandQueueProperty(cl_command_queue commandQueue,
639 cl_command_queue_properties properties,
640 cl_bool enable,
641 cl_command_queue_properties *oldProperties) {
642 TRACING_ENTER(clSetCommandQueueProperty, &commandQueue, &properties, &enable, &oldProperties);
643 cl_int retVal = CL_INVALID_VALUE;
644 API_ENTER(&retVal);
645 DBG_LOG_INPUTS("commandQueue", commandQueue,
646 "properties", properties,
647 "enable", enable,
648 "oldProperties", oldProperties);
649 TRACING_EXIT(clSetCommandQueueProperty, &retVal);
650 return retVal;
651 }
652
clCreateBuffer(cl_context context,cl_mem_flags flags,size_t size,void * hostPtr,cl_int * errcodeRet)653 cl_mem CL_API_CALL clCreateBuffer(cl_context context,
654 cl_mem_flags flags,
655 size_t size,
656 void *hostPtr,
657 cl_int *errcodeRet) {
658 TRACING_ENTER(clCreateBuffer, &context, &flags, &size, &hostPtr, &errcodeRet);
659 DBG_LOG_INPUTS("cl_context", context,
660 "cl_mem_flags", flags,
661 "size", size,
662 "hostPtr", NEO::FileLoggerInstance().infoPointerToString(hostPtr, size));
663
664 cl_int retVal = CL_SUCCESS;
665 API_ENTER(&retVal);
666
667 cl_mem_properties *properties = nullptr;
668 cl_mem_flags_intel flagsIntel = 0;
669 cl_mem buffer = BufferFunctions::validateInputAndCreateBuffer(context, properties, flags, flagsIntel, size, hostPtr, retVal);
670
671 ErrorCodeHelper{errcodeRet, retVal};
672 DBG_LOG_INPUTS("buffer", buffer);
673 TRACING_EXIT(clCreateBuffer, &buffer);
674 return buffer;
675 }
676
clCreateBufferWithProperties(cl_context context,const cl_mem_properties * properties,cl_mem_flags flags,size_t size,void * hostPtr,cl_int * errcodeRet)677 cl_mem CL_API_CALL clCreateBufferWithProperties(cl_context context,
678 const cl_mem_properties *properties,
679 cl_mem_flags flags,
680 size_t size,
681 void *hostPtr,
682 cl_int *errcodeRet) {
683 DBG_LOG_INPUTS("cl_context", context,
684 "cl_mem_properties", properties,
685 "cl_mem_flags", flags,
686 "size", size,
687 "hostPtr", NEO::FileLoggerInstance().infoPointerToString(hostPtr, size));
688
689 cl_int retVal = CL_SUCCESS;
690 API_ENTER(&retVal);
691
692 cl_mem_flags_intel flagsIntel = 0;
693 cl_mem buffer = BufferFunctions::validateInputAndCreateBuffer(context, properties, flags, flagsIntel, size, hostPtr, retVal);
694
695 ErrorCodeHelper{errcodeRet, retVal};
696 DBG_LOG_INPUTS("buffer", buffer);
697 return buffer;
698 }
699
clCreateBufferWithPropertiesINTEL(cl_context context,const cl_mem_properties_intel * properties,cl_mem_flags flags,size_t size,void * hostPtr,cl_int * errcodeRet)700 cl_mem CL_API_CALL clCreateBufferWithPropertiesINTEL(cl_context context,
701 const cl_mem_properties_intel *properties,
702 cl_mem_flags flags,
703 size_t size,
704 void *hostPtr,
705 cl_int *errcodeRet) {
706 DBG_LOG_INPUTS("cl_context", context,
707 "cl_mem_properties_intel", properties,
708 "cl_mem_flags", flags,
709 "size", size,
710 "hostPtr", NEO::FileLoggerInstance().infoPointerToString(hostPtr, size));
711
712 cl_int retVal = CL_SUCCESS;
713 API_ENTER(&retVal);
714
715 cl_mem_flags_intel flagsIntel = 0;
716 cl_mem buffer = BufferFunctions::validateInputAndCreateBuffer(context, properties, flags, flagsIntel, size, hostPtr, retVal);
717
718 ErrorCodeHelper{errcodeRet, retVal};
719 DBG_LOG_INPUTS("buffer", buffer);
720 return buffer;
721 }
722
clCreateSubBuffer(cl_mem buffer,cl_mem_flags flags,cl_buffer_create_type bufferCreateType,const void * bufferCreateInfo,cl_int * errcodeRet)723 cl_mem CL_API_CALL clCreateSubBuffer(cl_mem buffer,
724 cl_mem_flags flags,
725 cl_buffer_create_type bufferCreateType,
726 const void *bufferCreateInfo,
727 cl_int *errcodeRet) {
728 TRACING_ENTER(clCreateSubBuffer, &buffer, &flags, &bufferCreateType, &bufferCreateInfo, &errcodeRet);
729 cl_int retVal = CL_SUCCESS;
730 API_ENTER(&retVal);
731 DBG_LOG_INPUTS("buffer", buffer,
732 "flags", flags,
733 "bufferCreateType", bufferCreateType,
734 "bufferCreateInfo", bufferCreateInfo);
735 cl_mem subBuffer = nullptr;
736 Buffer *parentBuffer = castToObject<Buffer>(buffer);
737
738 do {
739 if (parentBuffer == nullptr) {
740 retVal = CL_INVALID_MEM_OBJECT;
741 break;
742 }
743
744 /* Are there some invalid flag bits? */
745 if (!MemObjHelper::checkMemFlagsForSubBuffer(flags)) {
746 retVal = CL_INVALID_VALUE;
747 break;
748 }
749
750 cl_mem_flags parentFlags = parentBuffer->getFlags();
751 cl_mem_flags_intel parentFlagsIntel = parentBuffer->getFlagsIntel();
752
753 if (parentBuffer->isSubBuffer() == true) {
754 retVal = CL_INVALID_MEM_OBJECT;
755 break;
756 }
757
758 /* Check whether flag is valid. */
759 if (((flags & CL_MEM_HOST_READ_ONLY) && (flags & CL_MEM_HOST_NO_ACCESS)) ||
760 ((flags & CL_MEM_HOST_READ_ONLY) && (flags & CL_MEM_HOST_WRITE_ONLY)) ||
761 ((flags & CL_MEM_HOST_WRITE_ONLY) && (flags & CL_MEM_HOST_NO_ACCESS))) {
762 retVal = CL_INVALID_VALUE;
763 break;
764 }
765
766 /* Check whether flag is valid and compatible with parent. */
767 if (flags &&
768 (((parentFlags & CL_MEM_WRITE_ONLY) && (flags & (CL_MEM_READ_WRITE | CL_MEM_READ_ONLY))) ||
769 ((parentFlags & CL_MEM_READ_ONLY) && (flags & (CL_MEM_READ_WRITE | CL_MEM_WRITE_ONLY))) ||
770 ((parentFlags & CL_MEM_HOST_WRITE_ONLY) && (flags & CL_MEM_HOST_READ_ONLY)) ||
771 ((parentFlags & CL_MEM_HOST_READ_ONLY) && (flags & CL_MEM_HOST_WRITE_ONLY)) ||
772 ((parentFlags & CL_MEM_HOST_NO_ACCESS) &&
773 (flags & (CL_MEM_HOST_WRITE_ONLY | CL_MEM_HOST_READ_ONLY))))) {
774 retVal = CL_INVALID_VALUE;
775 break;
776 }
777
778 /* Inherit some flags if we do not set. */
779 if ((flags & (CL_MEM_WRITE_ONLY | CL_MEM_READ_ONLY | CL_MEM_READ_WRITE)) == 0) {
780 flags |= parentFlags & (CL_MEM_WRITE_ONLY | CL_MEM_READ_ONLY | CL_MEM_READ_WRITE);
781 }
782 if ((flags & (CL_MEM_HOST_WRITE_ONLY | CL_MEM_HOST_READ_ONLY | CL_MEM_HOST_NO_ACCESS)) == 0) {
783 flags |= parentFlags & (CL_MEM_HOST_WRITE_ONLY | CL_MEM_HOST_READ_ONLY |
784 CL_MEM_HOST_NO_ACCESS);
785 }
786 flags |= parentFlags & (CL_MEM_USE_HOST_PTR | CL_MEM_ALLOC_HOST_PTR | CL_MEM_COPY_HOST_PTR);
787
788 if (bufferCreateType != CL_BUFFER_CREATE_TYPE_REGION) {
789 retVal = CL_INVALID_VALUE;
790 break;
791 }
792
793 if (bufferCreateInfo == nullptr) {
794 retVal = CL_INVALID_VALUE;
795 break;
796 }
797
798 /* Must have non-zero size. */
799 const cl_buffer_region *region = reinterpret_cast<const cl_buffer_region *>(bufferCreateInfo);
800 if (region->size == 0) {
801 retVal = CL_INVALID_BUFFER_SIZE;
802 break;
803 }
804
805 /* Out of range. */
806 if (region->origin > parentBuffer->getSize() ||
807 region->origin + region->size > parentBuffer->getSize()) {
808 retVal = CL_INVALID_VALUE;
809 break;
810 }
811
812 if (!parentBuffer->isValidSubBufferOffset(region->origin)) {
813 retVal = CL_MISALIGNED_SUB_BUFFER_OFFSET;
814 break;
815 }
816
817 subBuffer = parentBuffer->createSubBuffer(flags, parentFlagsIntel, region, retVal);
818 } while (false);
819
820 if (errcodeRet) {
821 *errcodeRet = retVal;
822 }
823
824 TRACING_EXIT(clCreateSubBuffer, &subBuffer);
825 return subBuffer;
826 }
827
clCreateImage(cl_context context,cl_mem_flags flags,const cl_image_format * imageFormat,const cl_image_desc * imageDesc,void * hostPtr,cl_int * errcodeRet)828 cl_mem CL_API_CALL clCreateImage(cl_context context,
829 cl_mem_flags flags,
830 const cl_image_format *imageFormat,
831 const cl_image_desc *imageDesc,
832 void *hostPtr,
833 cl_int *errcodeRet) {
834 TRACING_ENTER(clCreateImage, &context, &flags, &imageFormat, &imageDesc, &hostPtr, &errcodeRet);
835
836 DBG_LOG_INPUTS("cl_context", context,
837 "cl_mem_flags", flags,
838 "cl_image_format.channel_data_type", imageFormat->image_channel_data_type,
839 "cl_image_format.channel_order", imageFormat->image_channel_order,
840 "cl_image_desc.width", imageDesc->image_width,
841 "cl_image_desc.heigth", imageDesc->image_height,
842 "cl_image_desc.depth", imageDesc->image_depth,
843 "cl_image_desc.type", imageDesc->image_type,
844 "cl_image_desc.array_size", imageDesc->image_array_size,
845 "hostPtr", hostPtr);
846
847 cl_int retVal = CL_SUCCESS;
848 API_ENTER(&retVal);
849
850 cl_mem_properties *properties = nullptr;
851 cl_mem_flags_intel flagsIntel = 0;
852
853 retVal = Image::checkIfDeviceSupportsImages(context);
854
855 cl_mem image = nullptr;
856 if (retVal == CL_SUCCESS) {
857 image = ImageFunctions::validateAndCreateImage(context, properties, flags, flagsIntel, imageFormat, imageDesc, hostPtr, retVal);
858 }
859
860 ErrorCodeHelper{errcodeRet, retVal};
861 DBG_LOG_INPUTS("image", image);
862 TRACING_EXIT(clCreateImage, &image);
863 return image;
864 }
865
clCreateImageWithProperties(cl_context context,const cl_mem_properties * properties,cl_mem_flags flags,const cl_image_format * imageFormat,const cl_image_desc * imageDesc,void * hostPtr,cl_int * errcodeRet)866 cl_mem CL_API_CALL clCreateImageWithProperties(cl_context context,
867 const cl_mem_properties *properties,
868 cl_mem_flags flags,
869 const cl_image_format *imageFormat,
870 const cl_image_desc *imageDesc,
871 void *hostPtr,
872 cl_int *errcodeRet) {
873
874 DBG_LOG_INPUTS("cl_context", context,
875 "cl_mem_properties", properties,
876 "cl_mem_flags", flags,
877 "cl_image_format.channel_data_type", imageFormat->image_channel_data_type,
878 "cl_image_format.channel_order", imageFormat->image_channel_order,
879 "cl_image_desc.width", imageDesc->image_width,
880 "cl_image_desc.heigth", imageDesc->image_height,
881 "cl_image_desc.depth", imageDesc->image_depth,
882 "cl_image_desc.type", imageDesc->image_type,
883 "cl_image_desc.array_size", imageDesc->image_array_size,
884 "hostPtr", hostPtr);
885
886 cl_int retVal = CL_SUCCESS;
887 API_ENTER(&retVal);
888
889 cl_mem_flags_intel flagsIntel = 0;
890
891 retVal = Image::checkIfDeviceSupportsImages(context);
892
893 cl_mem image = nullptr;
894 if (retVal == CL_SUCCESS) {
895 image = ImageFunctions::validateAndCreateImage(context, properties, flags, flagsIntel, imageFormat, imageDesc, hostPtr, retVal);
896 }
897
898 ErrorCodeHelper{errcodeRet, retVal};
899 DBG_LOG_INPUTS("image", image);
900 return image;
901 }
902
clCreateImageWithPropertiesINTEL(cl_context context,const cl_mem_properties_intel * properties,cl_mem_flags flags,const cl_image_format * imageFormat,const cl_image_desc * imageDesc,void * hostPtr,cl_int * errcodeRet)903 cl_mem CL_API_CALL clCreateImageWithPropertiesINTEL(cl_context context,
904 const cl_mem_properties_intel *properties,
905 cl_mem_flags flags,
906 const cl_image_format *imageFormat,
907 const cl_image_desc *imageDesc,
908 void *hostPtr,
909 cl_int *errcodeRet) {
910
911 DBG_LOG_INPUTS("cl_context", context,
912 "cl_mem_properties_intel", properties,
913 "cl_mem_flags", flags,
914 "cl_image_format.channel_data_type", imageFormat->image_channel_data_type,
915 "cl_image_format.channel_order", imageFormat->image_channel_order,
916 "cl_image_desc.width", imageDesc->image_width,
917 "cl_image_desc.heigth", imageDesc->image_height,
918 "cl_image_desc.depth", imageDesc->image_depth,
919 "cl_image_desc.type", imageDesc->image_type,
920 "cl_image_desc.array_size", imageDesc->image_array_size,
921 "hostPtr", hostPtr);
922
923 cl_int retVal = CL_SUCCESS;
924 API_ENTER(&retVal);
925
926 cl_mem_flags_intel flagsIntel = 0;
927 cl_mem image = ImageFunctions::validateAndCreateImage(context, properties, flags, flagsIntel, imageFormat, imageDesc, hostPtr, retVal);
928
929 ErrorCodeHelper{errcodeRet, retVal};
930 DBG_LOG_INPUTS("image", image);
931 return image;
932 }
933
934 // deprecated OpenCL 1.1
clCreateImage2D(cl_context context,cl_mem_flags flags,const cl_image_format * imageFormat,size_t imageWidth,size_t imageHeight,size_t imageRowPitch,void * hostPtr,cl_int * errcodeRet)935 cl_mem CL_API_CALL clCreateImage2D(cl_context context,
936 cl_mem_flags flags,
937 const cl_image_format *imageFormat,
938 size_t imageWidth,
939 size_t imageHeight,
940 size_t imageRowPitch,
941 void *hostPtr,
942 cl_int *errcodeRet) {
943 TRACING_ENTER(clCreateImage2D, &context, &flags, &imageFormat, &imageWidth, &imageHeight, &imageRowPitch, &hostPtr, &errcodeRet);
944
945 DBG_LOG_INPUTS("context", context,
946 "flags", flags,
947 "imageFormat", imageFormat,
948 "imageWidth", imageWidth,
949 "imageHeight", imageHeight,
950 "imageRowPitch", imageRowPitch,
951 "hostPtr", hostPtr);
952
953 cl_int retVal = CL_SUCCESS;
954 API_ENTER(&retVal);
955
956 cl_image_desc imageDesc;
957 memset(&imageDesc, 0, sizeof(cl_image_desc));
958
959 imageDesc.image_height = imageHeight;
960 imageDesc.image_width = imageWidth;
961 imageDesc.image_row_pitch = imageRowPitch;
962 imageDesc.image_type = CL_MEM_OBJECT_IMAGE2D;
963
964 cl_mem_properties *properties = nullptr;
965 cl_mem_flags_intel flagsIntel = 0;
966
967 retVal = Image::checkIfDeviceSupportsImages(context);
968
969 cl_mem image2D = nullptr;
970 if (retVal == CL_SUCCESS) {
971 image2D = ImageFunctions::validateAndCreateImage(context, properties, flags, flagsIntel, imageFormat, &imageDesc, hostPtr, retVal);
972 }
973
974 ErrorCodeHelper{errcodeRet, retVal};
975 DBG_LOG_INPUTS("image 2D", image2D);
976 TRACING_EXIT(clCreateImage2D, &image2D);
977 return image2D;
978 }
979
980 // deprecated OpenCL 1.1
clCreateImage3D(cl_context context,cl_mem_flags flags,const cl_image_format * imageFormat,size_t imageWidth,size_t imageHeight,size_t imageDepth,size_t imageRowPitch,size_t imageSlicePitch,void * hostPtr,cl_int * errcodeRet)981 cl_mem CL_API_CALL clCreateImage3D(cl_context context,
982 cl_mem_flags flags,
983 const cl_image_format *imageFormat,
984 size_t imageWidth,
985 size_t imageHeight,
986 size_t imageDepth,
987 size_t imageRowPitch,
988 size_t imageSlicePitch,
989 void *hostPtr,
990 cl_int *errcodeRet) {
991 TRACING_ENTER(clCreateImage3D, &context, &flags, &imageFormat, &imageWidth, &imageHeight, &imageDepth, &imageRowPitch, &imageSlicePitch, &hostPtr, &errcodeRet);
992
993 DBG_LOG_INPUTS("context", context,
994 "flags", flags,
995 "imageFormat", imageFormat,
996 "imageWidth", imageWidth,
997 "imageHeight", imageHeight,
998 "imageDepth", imageDepth,
999 "imageRowPitch", imageRowPitch,
1000 "imageSlicePitch", imageSlicePitch,
1001 "hostPtr", hostPtr);
1002
1003 cl_int retVal = CL_SUCCESS;
1004 API_ENTER(&retVal);
1005
1006 cl_image_desc imageDesc;
1007 memset(&imageDesc, 0, sizeof(cl_image_desc));
1008
1009 imageDesc.image_depth = imageDepth;
1010 imageDesc.image_height = imageHeight;
1011 imageDesc.image_width = imageWidth;
1012 imageDesc.image_row_pitch = imageRowPitch;
1013 imageDesc.image_slice_pitch = imageSlicePitch;
1014 imageDesc.image_type = CL_MEM_OBJECT_IMAGE3D;
1015
1016 cl_mem_properties *properties = nullptr;
1017 cl_mem_flags_intel intelFlags = 0;
1018
1019 retVal = Image::checkIfDeviceSupportsImages(context);
1020
1021 cl_mem image3D = nullptr;
1022 if (retVal == CL_SUCCESS) {
1023 image3D = ImageFunctions::validateAndCreateImage(context, properties, flags, intelFlags, imageFormat, &imageDesc, hostPtr, retVal);
1024 }
1025
1026 ErrorCodeHelper{errcodeRet, retVal};
1027 DBG_LOG_INPUTS("image 3D", image3D);
1028 TRACING_EXIT(clCreateImage3D, &image3D);
1029 return image3D;
1030 }
1031
clRetainMemObject(cl_mem memobj)1032 cl_int CL_API_CALL clRetainMemObject(cl_mem memobj) {
1033 TRACING_ENTER(clRetainMemObject, &memobj);
1034 cl_int retVal = CL_INVALID_MEM_OBJECT;
1035 API_ENTER(&retVal);
1036 DBG_LOG_INPUTS("memobj", memobj);
1037
1038 auto pMemObj = castToObject<MemObj>(memobj);
1039
1040 if (pMemObj) {
1041 pMemObj->retain();
1042 retVal = CL_SUCCESS;
1043 TRACING_EXIT(clRetainMemObject, &retVal);
1044 return retVal;
1045 }
1046
1047 TRACING_EXIT(clRetainMemObject, &retVal);
1048 return retVal;
1049 }
1050
clReleaseMemObject(cl_mem memobj)1051 cl_int CL_API_CALL clReleaseMemObject(cl_mem memobj) {
1052 TRACING_ENTER(clReleaseMemObject, &memobj);
1053 cl_int retVal = CL_INVALID_MEM_OBJECT;
1054 API_ENTER(&retVal);
1055
1056 DBG_LOG_INPUTS("memobj", memobj);
1057
1058 auto pMemObj = castToObject<MemObj>(memobj);
1059 if (pMemObj) {
1060 pMemObj->release();
1061 retVal = CL_SUCCESS;
1062 TRACING_EXIT(clReleaseMemObject, &retVal);
1063 return retVal;
1064 }
1065
1066 TRACING_EXIT(clReleaseMemObject, &retVal);
1067 return retVal;
1068 }
1069
clGetSupportedImageFormats(cl_context context,cl_mem_flags flags,cl_mem_object_type imageType,cl_uint numEntries,cl_image_format * imageFormats,cl_uint * numImageFormats)1070 cl_int CL_API_CALL clGetSupportedImageFormats(cl_context context,
1071 cl_mem_flags flags,
1072 cl_mem_object_type imageType,
1073 cl_uint numEntries,
1074 cl_image_format *imageFormats,
1075 cl_uint *numImageFormats) {
1076 TRACING_ENTER(clGetSupportedImageFormats, &context, &flags, &imageType, &numEntries, &imageFormats, &numImageFormats);
1077 cl_int retVal = CL_SUCCESS;
1078 API_ENTER(&retVal);
1079 DBG_LOG_INPUTS("context", context,
1080 "flags", flags,
1081 "imageType", imageType,
1082 "numEntries", numEntries,
1083 "imageFormats", imageFormats,
1084 "numImageFormats", numImageFormats);
1085 auto pContext = castToObject<Context>(context);
1086 if (pContext) {
1087 auto pClDevice = pContext->getDevice(0);
1088 if (pClDevice->getHardwareInfo().capabilityTable.supportsImages) {
1089 retVal = pContext->getSupportedImageFormats(&pClDevice->getDevice(), flags, imageType, numEntries,
1090 imageFormats, numImageFormats);
1091 } else {
1092 if (numImageFormats) {
1093 *numImageFormats = 0u;
1094 }
1095 retVal = CL_SUCCESS;
1096 }
1097 } else {
1098 retVal = CL_INVALID_CONTEXT;
1099 }
1100
1101 TRACING_EXIT(clGetSupportedImageFormats, &retVal);
1102 return retVal;
1103 }
1104
clGetMemObjectInfo(cl_mem memobj,cl_mem_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)1105 cl_int CL_API_CALL clGetMemObjectInfo(cl_mem memobj,
1106 cl_mem_info paramName,
1107 size_t paramValueSize,
1108 void *paramValue,
1109 size_t *paramValueSizeRet) {
1110 TRACING_ENTER(clGetMemObjectInfo, &memobj, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
1111 cl_int retVal = CL_SUCCESS;
1112 API_ENTER(&retVal);
1113 DBG_LOG_INPUTS("memobj", memobj,
1114 "paramName", paramName,
1115 "paramValueSize", paramValueSize,
1116 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
1117 "paramValueSizeRet", paramValueSizeRet);
1118 MemObj *pMemObj = nullptr;
1119 retVal = validateObjects(WithCastToInternal(memobj, &pMemObj));
1120 if (CL_SUCCESS != retVal) {
1121 TRACING_EXIT(clGetMemObjectInfo, &retVal);
1122 return retVal;
1123 }
1124
1125 retVal = pMemObj->getMemObjectInfo(paramName, paramValueSize,
1126 paramValue, paramValueSizeRet);
1127 TRACING_EXIT(clGetMemObjectInfo, &retVal);
1128 return retVal;
1129 }
1130
clGetImageInfo(cl_mem image,cl_image_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)1131 cl_int CL_API_CALL clGetImageInfo(cl_mem image,
1132 cl_image_info paramName,
1133 size_t paramValueSize,
1134 void *paramValue,
1135 size_t *paramValueSizeRet) {
1136 TRACING_ENTER(clGetImageInfo, &image, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
1137 cl_int retVal = CL_SUCCESS;
1138 API_ENTER(&retVal);
1139 DBG_LOG_INPUTS("image", image,
1140 "paramName", paramName,
1141 "paramValueSize", paramValueSize,
1142 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
1143 "paramValueSizeRet", paramValueSizeRet);
1144 retVal = validateObjects(image);
1145 if (CL_SUCCESS != retVal) {
1146 TRACING_EXIT(clGetImageInfo, &retVal);
1147 return retVal;
1148 }
1149
1150 auto pImgObj = castToObject<Image>(image);
1151 if (pImgObj == nullptr) {
1152 retVal = CL_INVALID_MEM_OBJECT;
1153 TRACING_EXIT(clGetImageInfo, &retVal);
1154 return retVal;
1155 }
1156
1157 retVal = pImgObj->getImageInfo(paramName, paramValueSize, paramValue, paramValueSizeRet);
1158 TRACING_EXIT(clGetImageInfo, &retVal);
1159 return retVal;
1160 }
1161
clGetImageParamsINTEL(cl_context context,const cl_image_format * imageFormat,const cl_image_desc * imageDesc,size_t * imageRowPitch,size_t * imageSlicePitch)1162 cl_int CL_API_CALL clGetImageParamsINTEL(cl_context context,
1163 const cl_image_format *imageFormat,
1164 const cl_image_desc *imageDesc,
1165 size_t *imageRowPitch,
1166 size_t *imageSlicePitch) {
1167 cl_int retVal = CL_SUCCESS;
1168 API_ENTER(&retVal);
1169 DBG_LOG_INPUTS("context", context,
1170 "imageFormat", imageFormat,
1171 "imageDesc", imageDesc,
1172 "imageRowPitch", imageRowPitch,
1173 "imageSlicePitch", imageSlicePitch);
1174 const ClSurfaceFormatInfo *surfaceFormat = nullptr;
1175 cl_mem_flags memFlags = CL_MEM_READ_ONLY;
1176 retVal = validateObjects(context);
1177 auto pContext = castToObject<Context>(context);
1178
1179 if (CL_SUCCESS == retVal) {
1180 if ((imageFormat == nullptr) || (imageDesc == nullptr) || (imageRowPitch == nullptr) || (imageSlicePitch == nullptr)) {
1181 retVal = CL_INVALID_VALUE;
1182 }
1183 }
1184 if (CL_SUCCESS == retVal) {
1185 retVal = Image::validateImageFormat(imageFormat);
1186 }
1187 if (CL_SUCCESS == retVal) {
1188 auto pClDevice = pContext->getDevice(0);
1189 surfaceFormat = Image::getSurfaceFormatFromTable(memFlags, imageFormat,
1190 pClDevice->getHardwareInfo().capabilityTable.supportsOcl21Features);
1191 retVal = Image::validate(pContext, ClMemoryPropertiesHelper::createMemoryProperties(memFlags, 0, 0, &pClDevice->getDevice()),
1192 surfaceFormat, imageDesc, nullptr);
1193 }
1194 if (CL_SUCCESS == retVal) {
1195 retVal = Image::getImageParams(pContext, memFlags, surfaceFormat, imageDesc, imageRowPitch, imageSlicePitch);
1196 }
1197 return retVal;
1198 }
1199
clSetMemObjectDestructorCallback(cl_mem memobj,void (CL_CALLBACK * funcNotify)(cl_mem,void *),void * userData)1200 cl_int CL_API_CALL clSetMemObjectDestructorCallback(cl_mem memobj,
1201 void(CL_CALLBACK *funcNotify)(cl_mem, void *),
1202 void *userData) {
1203 TRACING_ENTER(clSetMemObjectDestructorCallback, &memobj, &funcNotify, &userData);
1204 cl_int retVal = CL_SUCCESS;
1205 API_ENTER(&retVal);
1206 DBG_LOG_INPUTS("memobj", memobj, "funcNotify", funcNotify, "userData", userData);
1207 retVal = validateObjects(memobj, (void *)funcNotify);
1208
1209 if (CL_SUCCESS != retVal) {
1210 TRACING_EXIT(clSetMemObjectDestructorCallback, &retVal);
1211 return retVal;
1212 }
1213
1214 auto pMemObj = castToObject<MemObj>(memobj);
1215 retVal = pMemObj->setDestructorCallback(funcNotify, userData);
1216 TRACING_EXIT(clSetMemObjectDestructorCallback, &retVal);
1217 return retVal;
1218 }
1219
clCreateSampler(cl_context context,cl_bool normalizedCoords,cl_addressing_mode addressingMode,cl_filter_mode filterMode,cl_int * errcodeRet)1220 cl_sampler CL_API_CALL clCreateSampler(cl_context context,
1221 cl_bool normalizedCoords,
1222 cl_addressing_mode addressingMode,
1223 cl_filter_mode filterMode,
1224 cl_int *errcodeRet) {
1225 TRACING_ENTER(clCreateSampler, &context, &normalizedCoords, &addressingMode, &filterMode, &errcodeRet);
1226 cl_int retVal = CL_SUCCESS;
1227 API_ENTER(&retVal);
1228 DBG_LOG_INPUTS("context", context,
1229 "normalizedCoords", normalizedCoords,
1230 "addressingMode", addressingMode,
1231 "filterMode", filterMode);
1232 retVal = validateObjects(context);
1233 cl_sampler sampler = nullptr;
1234
1235 if (retVal == CL_SUCCESS) {
1236 auto pContext = castToObject<Context>(context);
1237 sampler = Sampler::create(
1238 pContext,
1239 normalizedCoords,
1240 addressingMode,
1241 filterMode,
1242 CL_FILTER_NEAREST,
1243 0.0f,
1244 std::numeric_limits<float>::max(),
1245 retVal);
1246 }
1247
1248 if (errcodeRet) {
1249 *errcodeRet = retVal;
1250 }
1251
1252 TRACING_EXIT(clCreateSampler, &sampler);
1253 return sampler;
1254 }
1255
clRetainSampler(cl_sampler sampler)1256 cl_int CL_API_CALL clRetainSampler(cl_sampler sampler) {
1257 TRACING_ENTER(clRetainSampler, &sampler);
1258 cl_int retVal = CL_SUCCESS;
1259 API_ENTER(&retVal);
1260 DBG_LOG_INPUTS("sampler", sampler);
1261 auto pSampler = castToObject<Sampler>(sampler);
1262 if (pSampler) {
1263 pSampler->retain();
1264 TRACING_EXIT(clRetainSampler, &retVal);
1265 return retVal;
1266 }
1267 retVal = CL_INVALID_SAMPLER;
1268 TRACING_EXIT(clRetainSampler, &retVal);
1269 return retVal;
1270 }
1271
clReleaseSampler(cl_sampler sampler)1272 cl_int CL_API_CALL clReleaseSampler(cl_sampler sampler) {
1273 TRACING_ENTER(clReleaseSampler, &sampler);
1274 cl_int retVal = CL_SUCCESS;
1275 API_ENTER(&retVal);
1276 DBG_LOG_INPUTS("sampler", sampler);
1277 auto pSampler = castToObject<Sampler>(sampler);
1278 if (pSampler) {
1279 pSampler->release();
1280 TRACING_EXIT(clReleaseSampler, &retVal);
1281 return retVal;
1282 }
1283 retVal = CL_INVALID_SAMPLER;
1284 TRACING_EXIT(clReleaseSampler, &retVal);
1285 return retVal;
1286 }
1287
clGetSamplerInfo(cl_sampler sampler,cl_sampler_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)1288 cl_int CL_API_CALL clGetSamplerInfo(cl_sampler sampler,
1289 cl_sampler_info paramName,
1290 size_t paramValueSize,
1291 void *paramValue,
1292 size_t *paramValueSizeRet) {
1293 TRACING_ENTER(clGetSamplerInfo, &sampler, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
1294 cl_int retVal = CL_INVALID_SAMPLER;
1295 API_ENTER(&retVal);
1296 DBG_LOG_INPUTS("sampler", sampler,
1297 "paramName", paramName,
1298 "paramValueSize", paramValueSize,
1299 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
1300 "paramValueSizeRet", paramValueSizeRet);
1301
1302 auto pSampler = castToObject<Sampler>(sampler);
1303
1304 if (pSampler) {
1305 retVal = pSampler->getInfo(paramName, paramValueSize,
1306 paramValue, paramValueSizeRet);
1307 }
1308
1309 TRACING_EXIT(clGetSamplerInfo, &retVal);
1310 return retVal;
1311 }
1312
clCreateProgramWithSource(cl_context context,cl_uint count,const char ** strings,const size_t * lengths,cl_int * errcodeRet)1313 cl_program CL_API_CALL clCreateProgramWithSource(cl_context context,
1314 cl_uint count,
1315 const char **strings,
1316 const size_t *lengths,
1317 cl_int *errcodeRet) {
1318 TRACING_ENTER(clCreateProgramWithSource, &context, &count, &strings, &lengths, &errcodeRet);
1319 cl_int retVal = CL_SUCCESS;
1320 API_ENTER(&retVal);
1321 DBG_LOG_INPUTS("context", context,
1322 "count", count,
1323 "strings", strings,
1324 "lengths", lengths);
1325 Context *pContext = nullptr;
1326 retVal = validateObjects(WithCastToInternal(context, &pContext), count, strings);
1327 cl_program program = nullptr;
1328
1329 if (CL_SUCCESS == retVal) {
1330 program = Program::create(
1331 pContext,
1332 count,
1333 strings,
1334 lengths,
1335 retVal);
1336 }
1337
1338 if (errcodeRet) {
1339 *errcodeRet = retVal;
1340 }
1341
1342 TRACING_EXIT(clCreateProgramWithSource, &program);
1343 return program;
1344 }
1345
clCreateProgramWithBinary(cl_context context,cl_uint numDevices,const cl_device_id * deviceList,const size_t * lengths,const unsigned char ** binaries,cl_int * binaryStatus,cl_int * errcodeRet)1346 cl_program CL_API_CALL clCreateProgramWithBinary(cl_context context,
1347 cl_uint numDevices,
1348 const cl_device_id *deviceList,
1349 const size_t *lengths,
1350 const unsigned char **binaries,
1351 cl_int *binaryStatus,
1352 cl_int *errcodeRet) {
1353 TRACING_ENTER(clCreateProgramWithBinary, &context, &numDevices, &deviceList, &lengths, &binaries, &binaryStatus, &errcodeRet);
1354 cl_int retVal = CL_SUCCESS;
1355 API_ENTER(&retVal);
1356 DBG_LOG_INPUTS("context", context,
1357 "numDevices", numDevices,
1358 "deviceList", deviceList,
1359 "lengths", lengths,
1360 "binaries", binaries,
1361 "binaryStatus", binaryStatus);
1362 Context *pContext = nullptr;
1363 retVal = validateObjects(WithCastToInternal(context, &pContext), deviceList, numDevices, binaries, lengths);
1364 cl_program program = nullptr;
1365 ClDeviceVector deviceVector;
1366
1367 if (retVal == CL_SUCCESS) {
1368 for (auto i = 0u; i < numDevices; i++) {
1369 auto device = castToObject<ClDevice>(deviceList[i]);
1370 if (!device || !pContext->isDeviceAssociated(*device)) {
1371 retVal = CL_INVALID_DEVICE;
1372 break;
1373 }
1374 if (lengths[i] == 0 || binaries[i] == nullptr) {
1375 retVal = CL_INVALID_VALUE;
1376 break;
1377 }
1378 deviceVector.push_back(device);
1379 }
1380 }
1381
1382 NEO::FileLoggerInstance().dumpBinaryProgram(numDevices, lengths, binaries);
1383
1384 if (CL_SUCCESS == retVal) {
1385 program = Program::create(
1386 pContext,
1387 deviceVector,
1388 lengths,
1389 binaries,
1390 binaryStatus,
1391 retVal);
1392 }
1393
1394 if (errcodeRet) {
1395 *errcodeRet = retVal;
1396 }
1397
1398 TRACING_EXIT(clCreateProgramWithBinary, &program);
1399 return program;
1400 }
1401
clCreateProgramWithIL(cl_context context,const void * il,size_t length,cl_int * errcodeRet)1402 cl_program CL_API_CALL clCreateProgramWithIL(cl_context context,
1403 const void *il,
1404 size_t length,
1405 cl_int *errcodeRet) {
1406 TRACING_ENTER(clCreateProgramWithIL, &context, &il, &length, &errcodeRet);
1407 cl_int retVal = CL_SUCCESS;
1408 API_ENTER(&retVal);
1409 DBG_LOG_INPUTS("context", context,
1410 "il", il,
1411 "length", length);
1412
1413 cl_program program = nullptr;
1414 Context *pContext = nullptr;
1415 retVal = validateObjects(WithCastToInternal(context, &pContext), il);
1416 if (retVal == CL_SUCCESS) {
1417 program = ProgramFunctions::createFromIL(
1418 pContext,
1419 il,
1420 length,
1421 retVal);
1422 }
1423
1424 if (errcodeRet != nullptr) {
1425 *errcodeRet = retVal;
1426 }
1427
1428 TRACING_EXIT(clCreateProgramWithIL, &program);
1429 return program;
1430 }
1431
clCreateProgramWithBuiltInKernels(cl_context context,cl_uint numDevices,const cl_device_id * deviceList,const char * kernelNames,cl_int * errcodeRet)1432 cl_program CL_API_CALL clCreateProgramWithBuiltInKernels(cl_context context,
1433 cl_uint numDevices,
1434 const cl_device_id *deviceList,
1435 const char *kernelNames,
1436 cl_int *errcodeRet) {
1437 TRACING_ENTER(clCreateProgramWithBuiltInKernels, &context, &numDevices, &deviceList, &kernelNames, &errcodeRet);
1438 cl_int retVal = CL_SUCCESS;
1439 API_ENTER(&retVal);
1440 DBG_LOG_INPUTS("context", context,
1441 "numDevices", numDevices,
1442 "deviceList", deviceList,
1443 "kernelNames", kernelNames);
1444 cl_program program = nullptr;
1445 Context *pContext = nullptr;
1446
1447 retVal = validateObjects(WithCastToInternal(context, &pContext), numDevices,
1448 deviceList, kernelNames, errcodeRet);
1449
1450 if (retVal == CL_SUCCESS) {
1451 ClDeviceVector deviceVector;
1452 for (auto i = 0u; i < numDevices; i++) {
1453 auto device = castToObject<ClDevice>(deviceList[i]);
1454 if (!device || !pContext->isDeviceAssociated(*device)) {
1455 retVal = CL_INVALID_DEVICE;
1456 break;
1457 }
1458 deviceVector.push_back(device);
1459 }
1460 if (retVal == CL_SUCCESS) {
1461
1462 program = Vme::createBuiltInProgram(
1463 *pContext,
1464 deviceVector,
1465 kernelNames,
1466 retVal);
1467 }
1468 }
1469
1470 if (errcodeRet) {
1471 *errcodeRet = retVal;
1472 }
1473
1474 TRACING_EXIT(clCreateProgramWithBuiltInKernels, &program);
1475 return program;
1476 }
1477
clRetainProgram(cl_program program)1478 cl_int CL_API_CALL clRetainProgram(cl_program program) {
1479 TRACING_ENTER(clRetainProgram, &program);
1480 cl_int retVal = CL_SUCCESS;
1481 API_ENTER(&retVal);
1482 DBG_LOG_INPUTS("program", program);
1483 auto pProgram = castToObject<Program>(program);
1484 if (pProgram) {
1485 pProgram->retain();
1486 TRACING_EXIT(clRetainProgram, &retVal);
1487 return retVal;
1488 }
1489 retVal = CL_INVALID_PROGRAM;
1490 TRACING_EXIT(clRetainProgram, &retVal);
1491 return retVal;
1492 }
1493
clReleaseProgram(cl_program program)1494 cl_int CL_API_CALL clReleaseProgram(cl_program program) {
1495 TRACING_ENTER(clReleaseProgram, &program);
1496 cl_int retVal = CL_SUCCESS;
1497 API_ENTER(&retVal);
1498 DBG_LOG_INPUTS("program", program);
1499 auto pProgram = castToObject<Program>(program);
1500 if (pProgram) {
1501 pProgram->release();
1502 TRACING_EXIT(clReleaseProgram, &retVal);
1503 return retVal;
1504 }
1505 retVal = CL_INVALID_PROGRAM;
1506 TRACING_EXIT(clReleaseProgram, &retVal);
1507 return retVal;
1508 }
1509
clBuildProgram(cl_program program,cl_uint numDevices,const cl_device_id * deviceList,const char * options,void (CL_CALLBACK * funcNotify)(cl_program program,void * userData),void * userData)1510 cl_int CL_API_CALL clBuildProgram(cl_program program,
1511 cl_uint numDevices,
1512 const cl_device_id *deviceList,
1513 const char *options,
1514 void(CL_CALLBACK *funcNotify)(cl_program program, void *userData),
1515 void *userData) {
1516 TRACING_ENTER(clBuildProgram, &program, &numDevices, &deviceList, &options, &funcNotify, &userData);
1517 cl_int retVal = CL_INVALID_PROGRAM;
1518 API_ENTER(&retVal);
1519 DBG_LOG_INPUTS("clProgram", program, "numDevices", numDevices, "cl_device_id", deviceList, "options", (options != nullptr) ? options : "", "funcNotify", funcNotify, "userData", userData);
1520 Program *pProgram = nullptr;
1521
1522 retVal = validateObjects(WithCastToInternal(program, &pProgram), Program::isValidCallback(funcNotify, userData));
1523
1524 if (CL_SUCCESS == retVal) {
1525 if (pProgram->isLocked()) {
1526 retVal = CL_INVALID_OPERATION;
1527 }
1528 }
1529
1530 ClDeviceVector deviceVector;
1531 ClDeviceVector *deviceVectorPtr = &deviceVector;
1532
1533 if (CL_SUCCESS == retVal) {
1534 retVal = Program::processInputDevices(deviceVectorPtr, numDevices, deviceList, pProgram->getDevices());
1535 }
1536 if (CL_SUCCESS == retVal) {
1537 retVal = pProgram->build(*deviceVectorPtr, options, clCacheEnabled);
1538 pProgram->invokeCallback(funcNotify, userData);
1539 }
1540
1541 TRACING_EXIT(clBuildProgram, &retVal);
1542 return retVal;
1543 }
1544
clCompileProgram(cl_program program,cl_uint numDevices,const cl_device_id * deviceList,const char * options,cl_uint numInputHeaders,const cl_program * inputHeaders,const char ** headerIncludeNames,void (CL_CALLBACK * funcNotify)(cl_program program,void * userData),void * userData)1545 cl_int CL_API_CALL clCompileProgram(cl_program program,
1546 cl_uint numDevices,
1547 const cl_device_id *deviceList,
1548 const char *options,
1549 cl_uint numInputHeaders,
1550 const cl_program *inputHeaders,
1551 const char **headerIncludeNames,
1552 void(CL_CALLBACK *funcNotify)(cl_program program, void *userData),
1553 void *userData) {
1554 TRACING_ENTER(clCompileProgram, &program, &numDevices, &deviceList, &options, &numInputHeaders, &inputHeaders, &headerIncludeNames, &funcNotify, &userData);
1555 cl_int retVal = CL_INVALID_PROGRAM;
1556 API_ENTER(&retVal);
1557 DBG_LOG_INPUTS("clProgram", program, "numDevices", numDevices, "cl_device_id", deviceList, "options", (options != nullptr) ? options : "", "numInputHeaders", numInputHeaders);
1558
1559 Program *pProgram = nullptr;
1560
1561 retVal = validateObjects(WithCastToInternal(program, &pProgram), Program::isValidCallback(funcNotify, userData));
1562
1563 if (CL_SUCCESS == retVal) {
1564 if (pProgram->isLocked()) {
1565 retVal = CL_INVALID_OPERATION;
1566 }
1567 }
1568
1569 ClDeviceVector deviceVector;
1570 ClDeviceVector *deviceVectorPtr = &deviceVector;
1571
1572 if (CL_SUCCESS == retVal) {
1573 retVal = Program::processInputDevices(deviceVectorPtr, numDevices, deviceList, pProgram->getDevices());
1574 }
1575 if (CL_SUCCESS == retVal) {
1576 retVal = pProgram->compile(*deviceVectorPtr, options,
1577 numInputHeaders, inputHeaders, headerIncludeNames);
1578 pProgram->invokeCallback(funcNotify, userData);
1579 }
1580
1581 TRACING_EXIT(clCompileProgram, &retVal);
1582 return retVal;
1583 }
1584
clLinkProgram(cl_context context,cl_uint numDevices,const cl_device_id * deviceList,const char * options,cl_uint numInputPrograms,const cl_program * inputPrograms,void (CL_CALLBACK * funcNotify)(cl_program program,void * userData),void * userData,cl_int * errcodeRet)1585 cl_program CL_API_CALL clLinkProgram(cl_context context,
1586 cl_uint numDevices,
1587 const cl_device_id *deviceList,
1588 const char *options,
1589 cl_uint numInputPrograms,
1590 const cl_program *inputPrograms,
1591 void(CL_CALLBACK *funcNotify)(cl_program program, void *userData),
1592 void *userData,
1593 cl_int *errcodeRet) {
1594 TRACING_ENTER(clLinkProgram, &context, &numDevices, &deviceList, &options, &numInputPrograms, &inputPrograms, &funcNotify, &userData, &errcodeRet);
1595 cl_int retVal = CL_SUCCESS;
1596 API_ENTER(&retVal);
1597 DBG_LOG_INPUTS("cl_context", context, "numDevices", numDevices, "cl_device_id", deviceList, "options", (options != nullptr) ? options : "", "numInputPrograms", numInputPrograms);
1598
1599 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
1600 Context *pContext = nullptr;
1601 Program *pProgram = nullptr;
1602
1603 retVal = validateObjects(WithCastToInternal(context, &pContext), Program::isValidCallback(funcNotify, userData));
1604
1605 ClDeviceVector deviceVector;
1606 ClDeviceVector *deviceVectorPtr = &deviceVector;
1607 if (CL_SUCCESS == retVal) {
1608 retVal = Program::processInputDevices(deviceVectorPtr, numDevices, deviceList, pContext->getDevices());
1609 }
1610
1611 if (CL_SUCCESS == retVal) {
1612
1613 pProgram = new Program(pContext, false, *deviceVectorPtr);
1614 retVal = pProgram->link(*deviceVectorPtr, options,
1615 numInputPrograms, inputPrograms);
1616 pProgram->invokeCallback(funcNotify, userData);
1617 }
1618
1619 err.set(retVal);
1620
1621 TRACING_EXIT(clLinkProgram, (cl_program *)&pProgram);
1622 return pProgram;
1623 }
1624
clUnloadPlatformCompiler(cl_platform_id platform)1625 cl_int CL_API_CALL clUnloadPlatformCompiler(cl_platform_id platform) {
1626 TRACING_ENTER(clUnloadPlatformCompiler, &platform);
1627 cl_int retVal = CL_SUCCESS;
1628 API_ENTER(&retVal);
1629 DBG_LOG_INPUTS("platform", platform);
1630
1631 retVal = validateObject(platform);
1632
1633 TRACING_EXIT(clUnloadPlatformCompiler, &retVal);
1634 return retVal;
1635 }
1636
clGetProgramInfo(cl_program program,cl_program_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)1637 cl_int CL_API_CALL clGetProgramInfo(cl_program program,
1638 cl_program_info paramName,
1639 size_t paramValueSize,
1640 void *paramValue,
1641 size_t *paramValueSizeRet) {
1642 TRACING_ENTER(clGetProgramInfo, &program, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
1643 cl_int retVal = CL_SUCCESS;
1644 API_ENTER(&retVal);
1645 DBG_LOG_INPUTS("clProgram", program, "paramName", paramName,
1646 "paramValueSize", paramValueSize,
1647 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
1648 "paramValueSizeRet", paramValueSizeRet);
1649 retVal = validateObjects(program);
1650
1651 if (CL_SUCCESS == retVal) {
1652 Program *pProgram = (Program *)(program);
1653
1654 retVal = pProgram->getInfo(
1655 paramName,
1656 paramValueSize,
1657 paramValue,
1658 paramValueSizeRet);
1659 }
1660 TRACING_EXIT(clGetProgramInfo, &retVal);
1661 return retVal;
1662 }
1663
clGetProgramBuildInfo(cl_program program,cl_device_id device,cl_program_build_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)1664 cl_int CL_API_CALL clGetProgramBuildInfo(cl_program program,
1665 cl_device_id device,
1666 cl_program_build_info paramName,
1667 size_t paramValueSize,
1668 void *paramValue,
1669 size_t *paramValueSizeRet) {
1670 TRACING_ENTER(clGetProgramBuildInfo, &program, &device, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
1671 cl_int retVal = CL_SUCCESS;
1672 API_ENTER(&retVal);
1673 DBG_LOG_INPUTS("clProgram", program, "cl_device_id", device,
1674 "paramName", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
1675 "paramValueSize", paramValueSize, "paramValue", paramValue,
1676 "paramValueSizeRet", paramValueSizeRet);
1677 Program *pProgram = nullptr;
1678 ClDevice *pClDevice = nullptr;
1679
1680 retVal = validateObjects(WithCastToInternal(program, &pProgram), WithCastToInternal(device, &pClDevice));
1681
1682 if (CL_SUCCESS == retVal) {
1683 if (!pProgram->isDeviceAssociated(*pClDevice)) {
1684 retVal = CL_INVALID_DEVICE;
1685 }
1686 }
1687 if (CL_SUCCESS == retVal) {
1688 retVal = pProgram->getBuildInfo(
1689 pClDevice,
1690 paramName,
1691 paramValueSize,
1692 paramValue,
1693 paramValueSizeRet);
1694 }
1695 TRACING_EXIT(clGetProgramBuildInfo, &retVal);
1696 return retVal;
1697 }
1698
clCreateKernel(cl_program clProgram,const char * kernelName,cl_int * errcodeRet)1699 cl_kernel CL_API_CALL clCreateKernel(cl_program clProgram,
1700 const char *kernelName,
1701 cl_int *errcodeRet) {
1702 TRACING_ENTER(clCreateKernel, &clProgram, &kernelName, &errcodeRet);
1703 API_ENTER(errcodeRet);
1704 Program *pProgram = nullptr;
1705 cl_kernel kernel = nullptr;
1706 cl_int retVal = CL_SUCCESS;
1707 DBG_LOG_INPUTS("clProgram", clProgram, "kernelName", kernelName);
1708
1709 do {
1710 if (!isValidObject(clProgram) ||
1711 !(pProgram = castToObject<Program>(clProgram))) {
1712 retVal = CL_INVALID_PROGRAM;
1713 break;
1714 }
1715
1716 if (kernelName == nullptr) {
1717 retVal = CL_INVALID_VALUE;
1718 break;
1719 }
1720
1721 if (!pProgram->isBuilt()) {
1722 retVal = CL_INVALID_PROGRAM_EXECUTABLE;
1723 break;
1724 }
1725
1726 bool kernelFound = false;
1727 KernelInfoContainer kernelInfos;
1728 kernelInfos.resize(pProgram->getMaxRootDeviceIndex() + 1);
1729
1730 for (const auto &pClDevice : pProgram->getDevicesInProgram()) {
1731 auto rootDeviceIndex = pClDevice->getRootDeviceIndex();
1732 auto pKernelInfo = pProgram->getKernelInfo(kernelName, rootDeviceIndex);
1733 if (pKernelInfo) {
1734 kernelFound = true;
1735 kernelInfos[rootDeviceIndex] = pKernelInfo;
1736 }
1737 }
1738
1739 if (!kernelFound) {
1740 retVal = CL_INVALID_KERNEL_NAME;
1741 break;
1742 }
1743
1744 kernel = MultiDeviceKernel::create(
1745 pProgram,
1746 kernelInfos,
1747 &retVal);
1748
1749 DBG_LOG_INPUTS("kernel", kernel);
1750 } while (false);
1751
1752 if (errcodeRet) {
1753 *errcodeRet = retVal;
1754 }
1755 gtpinNotifyKernelCreate(kernel);
1756 TRACING_EXIT(clCreateKernel, &kernel);
1757 return kernel;
1758 }
1759
clCreateKernelsInProgram(cl_program clProgram,cl_uint numKernels,cl_kernel * kernels,cl_uint * numKernelsRet)1760 cl_int CL_API_CALL clCreateKernelsInProgram(cl_program clProgram,
1761 cl_uint numKernels,
1762 cl_kernel *kernels,
1763 cl_uint *numKernelsRet) {
1764 TRACING_ENTER(clCreateKernelsInProgram, &clProgram, &numKernels, &kernels, &numKernelsRet);
1765 cl_int retVal = CL_SUCCESS;
1766 API_ENTER(&retVal);
1767 DBG_LOG_INPUTS("clProgram", clProgram,
1768 "numKernels", numKernels,
1769 "kernels", kernels,
1770 "numKernelsRet", numKernelsRet);
1771 auto pProgram = castToObject<Program>(clProgram);
1772 if (pProgram) {
1773 auto numKernelsInProgram = pProgram->getNumKernels();
1774
1775 if (kernels) {
1776 if (numKernels < numKernelsInProgram) {
1777 retVal = CL_INVALID_VALUE;
1778 TRACING_EXIT(clCreateKernelsInProgram, &retVal);
1779 return retVal;
1780 }
1781
1782 for (unsigned int i = 0; i < numKernelsInProgram; ++i) {
1783 KernelInfoContainer kernelInfos;
1784 kernelInfos.resize(pProgram->getMaxRootDeviceIndex() + 1);
1785 for (const auto &pClDevice : pProgram->getDevicesInProgram()) {
1786 auto rootDeviceIndex = pClDevice->getRootDeviceIndex();
1787 auto kernelInfo = pProgram->getKernelInfo(i, rootDeviceIndex);
1788 DEBUG_BREAK_IF(kernelInfo == nullptr);
1789 kernelInfos[rootDeviceIndex] = kernelInfo;
1790 }
1791 kernels[i] = MultiDeviceKernel::create(
1792 pProgram,
1793 kernelInfos,
1794 nullptr);
1795 gtpinNotifyKernelCreate(kernels[i]);
1796 }
1797 }
1798
1799 if (numKernelsRet) {
1800 *numKernelsRet = static_cast<cl_uint>(numKernelsInProgram);
1801 }
1802 TRACING_EXIT(clCreateKernelsInProgram, &retVal);
1803 return retVal;
1804 }
1805 retVal = CL_INVALID_PROGRAM;
1806 TRACING_EXIT(clCreateKernelsInProgram, &retVal);
1807 return retVal;
1808 }
1809
clRetainKernel(cl_kernel kernel)1810 cl_int CL_API_CALL clRetainKernel(cl_kernel kernel) {
1811 TRACING_ENTER(clRetainKernel, &kernel);
1812 cl_int retVal = CL_SUCCESS;
1813 API_ENTER(&retVal);
1814 DBG_LOG_INPUTS("kernel", kernel);
1815 auto pMultiDeviceKernel = castToObject<MultiDeviceKernel>(kernel);
1816 if (pMultiDeviceKernel) {
1817 pMultiDeviceKernel->retain();
1818 TRACING_EXIT(clRetainKernel, &retVal);
1819 return retVal;
1820 }
1821 retVal = CL_INVALID_KERNEL;
1822 TRACING_EXIT(clRetainKernel, &retVal);
1823 return retVal;
1824 }
1825
clReleaseKernel(cl_kernel kernel)1826 cl_int CL_API_CALL clReleaseKernel(cl_kernel kernel) {
1827 TRACING_ENTER(clReleaseKernel, &kernel);
1828 cl_int retVal = CL_SUCCESS;
1829 API_ENTER(&retVal);
1830 DBG_LOG_INPUTS("kernel", kernel);
1831 auto pMultiDeviceKernel = castToObject<MultiDeviceKernel>(kernel);
1832 if (pMultiDeviceKernel) {
1833 pMultiDeviceKernel->release();
1834 TRACING_EXIT(clReleaseKernel, &retVal);
1835 return retVal;
1836 }
1837 retVal = CL_INVALID_KERNEL;
1838 TRACING_EXIT(clReleaseKernel, &retVal);
1839 return retVal;
1840 }
1841
clSetKernelArg(cl_kernel kernel,cl_uint argIndex,size_t argSize,const void * argValue)1842 cl_int CL_API_CALL clSetKernelArg(cl_kernel kernel,
1843 cl_uint argIndex,
1844 size_t argSize,
1845 const void *argValue) {
1846 TRACING_ENTER(clSetKernelArg, &kernel, &argIndex, &argSize, &argValue);
1847 cl_int retVal = CL_SUCCESS;
1848 API_ENTER(&retVal);
1849 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
1850 retVal = validateObject(WithCastToInternal(kernel, &pMultiDeviceKernel));
1851 DBG_LOG_INPUTS("kernel", kernel, "argIndex", argIndex,
1852 "argSize", argSize, "argValue", NEO::FileLoggerInstance().infoPointerToString(argValue, argSize));
1853 do {
1854 if (retVal != CL_SUCCESS) {
1855 break;
1856 }
1857 if (pMultiDeviceKernel->getKernelArguments().size() <= argIndex) {
1858 retVal = CL_INVALID_ARG_INDEX;
1859 break;
1860 }
1861 retVal = pMultiDeviceKernel->checkCorrectImageAccessQualifier(argIndex, argSize, argValue);
1862 if (retVal != CL_SUCCESS) {
1863 pMultiDeviceKernel->unsetArg(argIndex);
1864 break;
1865 }
1866 retVal = pMultiDeviceKernel->setArg(
1867 argIndex,
1868 argSize,
1869 argValue);
1870 break;
1871
1872 } while (false);
1873 TRACING_EXIT(clSetKernelArg, &retVal);
1874 return retVal;
1875 }
1876
clGetKernelInfo(cl_kernel kernel,cl_kernel_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)1877 cl_int CL_API_CALL clGetKernelInfo(cl_kernel kernel,
1878 cl_kernel_info paramName,
1879 size_t paramValueSize,
1880 void *paramValue,
1881 size_t *paramValueSizeRet) {
1882 TRACING_ENTER(clGetKernelInfo, &kernel, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
1883 cl_int retVal = CL_SUCCESS;
1884 API_ENTER(&retVal);
1885 DBG_LOG_INPUTS("kernel", kernel, "paramName", paramName,
1886 "paramValueSize", paramValueSize,
1887 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
1888 "paramValueSizeRet", paramValueSizeRet);
1889 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
1890 retVal = validateObject(WithCastToInternal(kernel, &pMultiDeviceKernel));
1891 if (retVal == CL_SUCCESS) {
1892 retVal = pMultiDeviceKernel->getInfo(
1893 paramName,
1894 paramValueSize,
1895 paramValue,
1896 paramValueSizeRet);
1897 }
1898 TRACING_EXIT(clGetKernelInfo, &retVal);
1899 return retVal;
1900 }
1901
clGetKernelArgInfo(cl_kernel kernel,cl_uint argIndx,cl_kernel_arg_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)1902 cl_int CL_API_CALL clGetKernelArgInfo(cl_kernel kernel,
1903 cl_uint argIndx,
1904 cl_kernel_arg_info paramName,
1905 size_t paramValueSize,
1906 void *paramValue,
1907 size_t *paramValueSizeRet) {
1908 TRACING_ENTER(clGetKernelArgInfo, &kernel, &argIndx, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
1909 cl_int retVal = CL_SUCCESS;
1910 API_ENTER(&retVal);
1911
1912 DBG_LOG_INPUTS("kernel", kernel,
1913 "argIndx", argIndx,
1914 "paramName", paramName,
1915 "paramValueSize", paramValueSize,
1916 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
1917 "paramValueSizeRet", paramValueSizeRet);
1918
1919 auto pMultiDeviceKernel = castToObject<MultiDeviceKernel>(kernel);
1920 retVal = pMultiDeviceKernel
1921 ? pMultiDeviceKernel->getArgInfo(
1922 argIndx,
1923 paramName,
1924 paramValueSize,
1925 paramValue,
1926 paramValueSizeRet)
1927 : CL_INVALID_KERNEL;
1928 TRACING_EXIT(clGetKernelArgInfo, &retVal);
1929 return retVal;
1930 }
1931
clGetKernelWorkGroupInfo(cl_kernel kernel,cl_device_id device,cl_kernel_work_group_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)1932 cl_int CL_API_CALL clGetKernelWorkGroupInfo(cl_kernel kernel,
1933 cl_device_id device,
1934 cl_kernel_work_group_info paramName,
1935 size_t paramValueSize,
1936 void *paramValue,
1937 size_t *paramValueSizeRet) {
1938 TRACING_ENTER(clGetKernelWorkGroupInfo, &kernel, &device, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
1939 cl_int retVal = CL_SUCCESS;
1940 API_ENTER(&retVal);
1941
1942 DBG_LOG_INPUTS("kernel", kernel,
1943 "device", device,
1944 "paramName", paramName,
1945 "paramValueSize", paramValueSize,
1946 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
1947 "paramValueSizeRet", paramValueSizeRet);
1948
1949 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
1950 retVal = validateObjects(WithCastToInternal(kernel, &pMultiDeviceKernel));
1951
1952 ClDevice *pClDevice = nullptr;
1953 if (CL_SUCCESS == retVal) {
1954 if (pMultiDeviceKernel->getDevices().size() == 1u && !device) {
1955 pClDevice = pMultiDeviceKernel->getDevices()[0];
1956 } else {
1957 retVal = validateObjects(WithCastToInternal(device, &pClDevice));
1958 }
1959 }
1960 if (CL_SUCCESS == retVal) {
1961 auto pKernel = pMultiDeviceKernel->getKernel(pClDevice->getRootDeviceIndex());
1962 retVal = pKernel->getWorkGroupInfo(
1963 paramName,
1964 paramValueSize,
1965 paramValue,
1966 paramValueSizeRet);
1967 }
1968 TRACING_EXIT(clGetKernelWorkGroupInfo, &retVal);
1969 return retVal;
1970 }
1971
clWaitForEvents(cl_uint numEvents,const cl_event * eventList)1972 cl_int CL_API_CALL clWaitForEvents(cl_uint numEvents,
1973 const cl_event *eventList) {
1974 TRACING_ENTER(clWaitForEvents, &numEvents, &eventList);
1975
1976 auto retVal = CL_SUCCESS;
1977 API_ENTER(&retVal);
1978 DBG_LOG_INPUTS("eventList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventList), numEvents));
1979
1980 for (unsigned int i = 0; i < numEvents && retVal == CL_SUCCESS; i++)
1981 retVal = validateObjects(eventList[i]);
1982
1983 if (retVal != CL_SUCCESS) {
1984 TRACING_EXIT(clWaitForEvents, &retVal);
1985 return retVal;
1986 }
1987
1988 retVal = Event::waitForEvents(numEvents, eventList);
1989 TRACING_EXIT(clWaitForEvents, &retVal);
1990 return retVal;
1991 }
1992
clGetEventInfo(cl_event event,cl_event_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)1993 cl_int CL_API_CALL clGetEventInfo(cl_event event,
1994 cl_event_info paramName,
1995 size_t paramValueSize,
1996 void *paramValue,
1997 size_t *paramValueSizeRet) {
1998 TRACING_ENTER(clGetEventInfo, &event, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
1999 auto retVal = CL_SUCCESS;
2000 API_ENTER(&retVal);
2001
2002 DBG_LOG_INPUTS("event", event,
2003 "paramName", paramName,
2004 "paramValueSize", paramValueSize,
2005 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
2006 "paramValueSizeRet", paramValueSizeRet);
2007
2008 Event *neoEvent = castToObject<Event>(event);
2009 if (neoEvent == nullptr) {
2010 retVal = CL_INVALID_EVENT;
2011 TRACING_EXIT(clGetEventInfo, &retVal);
2012 return retVal;
2013 }
2014
2015 GetInfoHelper info(paramValue, paramValueSize, paramValueSizeRet);
2016 auto flushEvents = true;
2017
2018 switch (paramName) {
2019 default: {
2020 retVal = CL_INVALID_VALUE;
2021 TRACING_EXIT(clGetEventInfo, &retVal);
2022 return retVal;
2023 }
2024 // From OCL spec :
2025 // "Return the command-queue associated with event. For user event objects,"
2026 // a nullptr value is returned."
2027 case CL_EVENT_COMMAND_QUEUE: {
2028 if (neoEvent->isUserEvent()) {
2029 retVal = changeGetInfoStatusToCLResultType(info.set<cl_command_queue>(nullptr));
2030 TRACING_EXIT(clGetEventInfo, &retVal);
2031 return retVal;
2032 }
2033 retVal = changeGetInfoStatusToCLResultType(info.set<cl_command_queue>(neoEvent->getCommandQueue()));
2034 TRACING_EXIT(clGetEventInfo, &retVal);
2035 return retVal;
2036 }
2037 case CL_EVENT_CONTEXT:
2038 retVal = changeGetInfoStatusToCLResultType(info.set<cl_context>(neoEvent->getContext()));
2039 TRACING_EXIT(clGetEventInfo, &retVal);
2040 return retVal;
2041 case CL_EVENT_COMMAND_TYPE:
2042 retVal = changeGetInfoStatusToCLResultType(info.set<cl_command_type>(neoEvent->getCommandType()));
2043 TRACING_EXIT(clGetEventInfo, &retVal);
2044 return retVal;
2045 case CL_EVENT_COMMAND_EXECUTION_STATUS:
2046 if (DebugManager.flags.SkipFlushingEventsOnGetStatusCalls.get()) {
2047 flushEvents = false;
2048 }
2049 if (flushEvents) {
2050 neoEvent->tryFlushEvent();
2051 }
2052
2053 if (neoEvent->isUserEvent()) {
2054 auto executionStatus = neoEvent->peekExecutionStatus();
2055 //Spec requires initial state to be queued
2056 //our current design relies heavily on SUBMITTED status which directly corresponds
2057 //to command being able to be submitted, to overcome this we set initial status to queued
2058 //and we override the value stored with the value required by the spec.
2059 if (executionStatus == CL_QUEUED) {
2060 executionStatus = CL_SUBMITTED;
2061 }
2062 retVal = changeGetInfoStatusToCLResultType(info.set<cl_int>(executionStatus));
2063 TRACING_EXIT(clGetEventInfo, &retVal);
2064 return retVal;
2065 }
2066
2067 retVal = changeGetInfoStatusToCLResultType(info.set<cl_int>(neoEvent->updateEventAndReturnCurrentStatus()));
2068 TRACING_EXIT(clGetEventInfo, &retVal);
2069 return retVal;
2070 case CL_EVENT_REFERENCE_COUNT:
2071 retVal = changeGetInfoStatusToCLResultType(info.set<cl_uint>(neoEvent->getReference()));
2072 TRACING_EXIT(clGetEventInfo, &retVal);
2073 return retVal;
2074 }
2075 }
2076
clCreateUserEvent(cl_context context,cl_int * errcodeRet)2077 cl_event CL_API_CALL clCreateUserEvent(cl_context context,
2078 cl_int *errcodeRet) {
2079 TRACING_ENTER(clCreateUserEvent, &context, &errcodeRet);
2080 API_ENTER(errcodeRet);
2081
2082 DBG_LOG_INPUTS("context", context);
2083
2084 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
2085
2086 Context *ctx = castToObject<Context>(context);
2087 if (ctx == nullptr) {
2088 err.set(CL_INVALID_CONTEXT);
2089 cl_event retVal = nullptr;
2090 TRACING_EXIT(clCreateUserEvent, &retVal);
2091 return retVal;
2092 }
2093
2094 Event *userEvent = new UserEvent(ctx);
2095 cl_event userClEvent = userEvent;
2096 DBG_LOG_INPUTS("cl_event", userClEvent, "UserEvent", userEvent);
2097
2098 TRACING_EXIT(clCreateUserEvent, &userClEvent);
2099 return userClEvent;
2100 }
2101
clRetainEvent(cl_event event)2102 cl_int CL_API_CALL clRetainEvent(cl_event event) {
2103 TRACING_ENTER(clRetainEvent, &event);
2104 auto retVal = CL_SUCCESS;
2105 API_ENTER(&retVal);
2106
2107 auto pEvent = castToObject<Event>(event);
2108 DBG_LOG_INPUTS("cl_event", event, "Event", pEvent);
2109
2110 if (pEvent) {
2111 pEvent->retain();
2112 TRACING_EXIT(clRetainEvent, &retVal);
2113 return retVal;
2114 }
2115 retVal = CL_INVALID_EVENT;
2116 TRACING_EXIT(clRetainEvent, &retVal);
2117 return retVal;
2118 }
2119
clReleaseEvent(cl_event event)2120 cl_int CL_API_CALL clReleaseEvent(cl_event event) {
2121 TRACING_ENTER(clReleaseEvent, &event);
2122 auto retVal = CL_SUCCESS;
2123 API_ENTER(&retVal);
2124 auto pEvent = castToObject<Event>(event);
2125 DBG_LOG_INPUTS("cl_event", event, "Event", pEvent);
2126
2127 if (pEvent) {
2128 pEvent->release();
2129 TRACING_EXIT(clReleaseEvent, &retVal);
2130 return retVal;
2131 }
2132 retVal = CL_INVALID_EVENT;
2133 TRACING_EXIT(clReleaseEvent, &retVal);
2134 return retVal;
2135 }
2136
clSetUserEventStatus(cl_event event,cl_int executionStatus)2137 cl_int CL_API_CALL clSetUserEventStatus(cl_event event,
2138 cl_int executionStatus) {
2139 TRACING_ENTER(clSetUserEventStatus, &event, &executionStatus);
2140 auto retVal = CL_SUCCESS;
2141 API_ENTER(&retVal);
2142 auto userEvent = castToObject<UserEvent>(event);
2143 DBG_LOG_INPUTS("cl_event", event, "executionStatus", executionStatus, "UserEvent", userEvent);
2144
2145 if (userEvent == nullptr) {
2146 retVal = CL_INVALID_EVENT;
2147 TRACING_EXIT(clSetUserEventStatus, &retVal);
2148 return retVal;
2149 }
2150
2151 if (executionStatus > CL_COMPLETE) {
2152 retVal = CL_INVALID_VALUE;
2153 TRACING_EXIT(clSetUserEventStatus, &retVal);
2154 return retVal;
2155 }
2156
2157 if (!userEvent->isInitialEventStatus()) {
2158 retVal = CL_INVALID_OPERATION;
2159 TRACING_EXIT(clSetUserEventStatus, &retVal);
2160 return retVal;
2161 }
2162
2163 userEvent->setStatus(executionStatus);
2164 TRACING_EXIT(clSetUserEventStatus, &retVal);
2165 return retVal;
2166 }
2167
clSetEventCallback(cl_event event,cl_int commandExecCallbackType,void (CL_CALLBACK * funcNotify)(cl_event,cl_int,void *),void * userData)2168 cl_int CL_API_CALL clSetEventCallback(cl_event event,
2169 cl_int commandExecCallbackType,
2170 void(CL_CALLBACK *funcNotify)(cl_event, cl_int, void *),
2171 void *userData) {
2172 TRACING_ENTER(clSetEventCallback, &event, &commandExecCallbackType, &funcNotify, &userData);
2173 auto retVal = CL_SUCCESS;
2174 API_ENTER(&retVal);
2175 auto eventObject = castToObject<Event>(event);
2176 DBG_LOG_INPUTS("cl_event", event, "commandExecCallbackType", commandExecCallbackType, "Event", eventObject);
2177
2178 if (eventObject == nullptr) {
2179 retVal = CL_INVALID_EVENT;
2180 TRACING_EXIT(clSetEventCallback, &retVal);
2181 return retVal;
2182 }
2183 switch (commandExecCallbackType) {
2184 case CL_COMPLETE:
2185 case CL_SUBMITTED:
2186 case CL_RUNNING:
2187 break;
2188 default: {
2189 retVal = CL_INVALID_VALUE;
2190 TRACING_EXIT(clSetEventCallback, &retVal);
2191 return retVal;
2192 }
2193 }
2194 if (funcNotify == nullptr) {
2195 retVal = CL_INVALID_VALUE;
2196 TRACING_EXIT(clSetEventCallback, &retVal);
2197 return retVal;
2198 }
2199
2200 eventObject->tryFlushEvent();
2201 eventObject->addCallback(funcNotify, commandExecCallbackType, userData);
2202 TRACING_EXIT(clSetEventCallback, &retVal);
2203 return retVal;
2204 }
2205
clGetEventProfilingInfo(cl_event event,cl_profiling_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)2206 cl_int CL_API_CALL clGetEventProfilingInfo(cl_event event,
2207 cl_profiling_info paramName,
2208 size_t paramValueSize,
2209 void *paramValue,
2210 size_t *paramValueSizeRet) {
2211 TRACING_ENTER(clGetEventProfilingInfo, &event, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
2212 auto retVal = CL_SUCCESS;
2213 API_ENTER(&retVal);
2214 DBG_LOG_INPUTS("event", event,
2215 "paramName", paramName,
2216 "paramValueSize", paramValueSize,
2217 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
2218 "paramValueSizeRet", paramValueSizeRet);
2219 auto eventObject = castToObject<Event>(event);
2220
2221 if (eventObject == nullptr) {
2222 retVal = CL_INVALID_EVENT;
2223 TRACING_EXIT(clGetEventProfilingInfo, &retVal);
2224 return retVal;
2225 }
2226
2227 retVal = eventObject->getEventProfilingInfo(paramName,
2228 paramValueSize,
2229 paramValue,
2230 paramValueSizeRet);
2231 TRACING_EXIT(clGetEventProfilingInfo, &retVal);
2232 return retVal;
2233 }
2234
clFlush(cl_command_queue commandQueue)2235 cl_int CL_API_CALL clFlush(cl_command_queue commandQueue) {
2236 TRACING_ENTER(clFlush, &commandQueue);
2237 cl_int retVal = CL_SUCCESS;
2238 API_ENTER(&retVal);
2239 DBG_LOG_INPUTS("commandQueue", commandQueue);
2240 auto pCommandQueue = castToObject<CommandQueue>(commandQueue);
2241
2242 retVal = pCommandQueue
2243 ? pCommandQueue->flush()
2244 : CL_INVALID_COMMAND_QUEUE;
2245 TRACING_EXIT(clFlush, &retVal);
2246 return retVal;
2247 }
2248
clFinish(cl_command_queue commandQueue)2249 cl_int CL_API_CALL clFinish(cl_command_queue commandQueue) {
2250 TRACING_ENTER(clFinish, &commandQueue);
2251 cl_int retVal = CL_SUCCESS;
2252 API_ENTER(&retVal);
2253 DBG_LOG_INPUTS("commandQueue", commandQueue);
2254 auto pCommandQueue = castToObject<CommandQueue>(commandQueue);
2255
2256 retVal = pCommandQueue
2257 ? pCommandQueue->finish()
2258 : CL_INVALID_COMMAND_QUEUE;
2259 TRACING_EXIT(clFinish, &retVal);
2260 return retVal;
2261 }
2262
clEnqueueReadBuffer(cl_command_queue commandQueue,cl_mem buffer,cl_bool blockingRead,size_t offset,size_t cb,void * ptr,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2263 cl_int CL_API_CALL clEnqueueReadBuffer(cl_command_queue commandQueue,
2264 cl_mem buffer,
2265 cl_bool blockingRead,
2266 size_t offset,
2267 size_t cb,
2268 void *ptr,
2269 cl_uint numEventsInWaitList,
2270 const cl_event *eventWaitList,
2271 cl_event *event) {
2272 TRACING_ENTER(clEnqueueReadBuffer, &commandQueue, &buffer, &blockingRead, &offset, &cb, &ptr, &numEventsInWaitList, &eventWaitList, &event);
2273 CommandQueue *pCommandQueue = nullptr;
2274 Buffer *pBuffer = nullptr;
2275
2276 auto retVal = validateObjects(
2277 WithCastToInternal(commandQueue, &pCommandQueue),
2278 WithCastToInternal(buffer, &pBuffer),
2279 ptr);
2280
2281 API_ENTER(&retVal);
2282
2283 DBG_LOG_INPUTS("commandQueue", commandQueue, "buffer", buffer, "blockingRead", blockingRead,
2284 "offset", offset, "cb", cb, "ptr", ptr,
2285 "numEventsInWaitList", numEventsInWaitList,
2286 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2287 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2288
2289 if (CL_SUCCESS == retVal) {
2290
2291 if (pBuffer->readMemObjFlagsInvalid()) {
2292 retVal = CL_INVALID_OPERATION;
2293 TRACING_EXIT(clEnqueueReadBuffer, &retVal);
2294 return retVal;
2295 }
2296
2297 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
2298 retVal = CL_INVALID_OPERATION;
2299 TRACING_EXIT(clEnqueueReadBuffer, &retVal);
2300 return retVal;
2301 }
2302
2303 retVal = pCommandQueue->enqueueReadBuffer(
2304 pBuffer,
2305 blockingRead,
2306 offset,
2307 cb,
2308 ptr,
2309 nullptr,
2310 numEventsInWaitList,
2311 eventWaitList,
2312 event);
2313 }
2314
2315 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
2316 TRACING_EXIT(clEnqueueReadBuffer, &retVal);
2317 return retVal;
2318 }
2319
clEnqueueReadBufferRect(cl_command_queue commandQueue,cl_mem buffer,cl_bool blockingRead,const size_t * bufferOrigin,const size_t * hostOrigin,const size_t * region,size_t bufferRowPitch,size_t bufferSlicePitch,size_t hostRowPitch,size_t hostSlicePitch,void * ptr,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2320 cl_int CL_API_CALL clEnqueueReadBufferRect(cl_command_queue commandQueue,
2321 cl_mem buffer,
2322 cl_bool blockingRead,
2323 const size_t *bufferOrigin,
2324 const size_t *hostOrigin,
2325 const size_t *region,
2326 size_t bufferRowPitch,
2327 size_t bufferSlicePitch,
2328 size_t hostRowPitch,
2329 size_t hostSlicePitch,
2330 void *ptr,
2331 cl_uint numEventsInWaitList,
2332 const cl_event *eventWaitList,
2333 cl_event *event) {
2334 TRACING_ENTER(clEnqueueReadBufferRect, &commandQueue, &buffer, &blockingRead, &bufferOrigin, &hostOrigin, ®ion, &bufferRowPitch, &bufferSlicePitch, &hostRowPitch, &hostSlicePitch, &ptr, &numEventsInWaitList, &eventWaitList, &event);
2335 cl_int retVal = CL_SUCCESS;
2336 API_ENTER(&retVal);
2337 DBG_LOG_INPUTS("commandQueue", commandQueue,
2338 "buffer", buffer,
2339 "blockingRead", blockingRead,
2340 "bufferOrigin[0]", NEO::FileLoggerInstance().getInput(bufferOrigin, 0),
2341 "bufferOrigin[1]", NEO::FileLoggerInstance().getInput(bufferOrigin, 1),
2342 "bufferOrigin[2]", NEO::FileLoggerInstance().getInput(bufferOrigin, 2),
2343 "hostOrigin[0]", NEO::FileLoggerInstance().getInput(hostOrigin, 0),
2344 "hostOrigin[1]", NEO::FileLoggerInstance().getInput(hostOrigin, 1),
2345 "hostOrigin[2]", NEO::FileLoggerInstance().getInput(hostOrigin, 2),
2346 "region[0]", NEO::FileLoggerInstance().getInput(region, 0),
2347 "region[1]", NEO::FileLoggerInstance().getInput(region, 1),
2348 "region[2]", NEO::FileLoggerInstance().getInput(region, 2),
2349 "bufferRowPitch", bufferRowPitch,
2350 "bufferSlicePitch", bufferSlicePitch,
2351 "hostRowPitch", hostRowPitch,
2352 "hostSlicePitch", hostSlicePitch,
2353 "ptr", ptr,
2354 "numEventsInWaitList", numEventsInWaitList,
2355 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2356 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2357
2358 CommandQueue *pCommandQueue = nullptr;
2359 Buffer *pBuffer = nullptr;
2360
2361 retVal = validateObjects(
2362 WithCastToInternal(commandQueue, &pCommandQueue),
2363 WithCastToInternal(buffer, &pBuffer),
2364 ptr);
2365
2366 if (CL_SUCCESS != retVal) {
2367 TRACING_EXIT(clEnqueueReadBufferRect, &retVal);
2368 return retVal;
2369 }
2370
2371 if (pBuffer->readMemObjFlagsInvalid()) {
2372 retVal = CL_INVALID_OPERATION;
2373 TRACING_EXIT(clEnqueueReadBufferRect, &retVal);
2374 return retVal;
2375 }
2376
2377 if (pBuffer->bufferRectPitchSet(bufferOrigin,
2378 region,
2379 bufferRowPitch,
2380 bufferSlicePitch,
2381 hostRowPitch,
2382 hostSlicePitch,
2383 true) == false) {
2384 retVal = CL_INVALID_VALUE;
2385 TRACING_EXIT(clEnqueueReadBufferRect, &retVal);
2386 return retVal;
2387 }
2388
2389 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_BUFFER_RECT_INTEL, numEventsInWaitList, eventWaitList, event)) {
2390 retVal = CL_INVALID_OPERATION;
2391 TRACING_EXIT(clEnqueueReadBufferRect, &retVal);
2392 return retVal;
2393 }
2394
2395 retVal = pCommandQueue->enqueueReadBufferRect(
2396 pBuffer,
2397 blockingRead,
2398 bufferOrigin,
2399 hostOrigin,
2400 region,
2401 bufferRowPitch,
2402 bufferSlicePitch,
2403 hostRowPitch,
2404 hostSlicePitch,
2405 ptr,
2406 numEventsInWaitList,
2407 eventWaitList,
2408 event);
2409 TRACING_EXIT(clEnqueueReadBufferRect, &retVal);
2410 return retVal;
2411 }
2412
clEnqueueWriteBuffer(cl_command_queue commandQueue,cl_mem buffer,cl_bool blockingWrite,size_t offset,size_t cb,const void * ptr,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2413 cl_int CL_API_CALL clEnqueueWriteBuffer(cl_command_queue commandQueue,
2414 cl_mem buffer,
2415 cl_bool blockingWrite,
2416 size_t offset,
2417 size_t cb,
2418 const void *ptr,
2419 cl_uint numEventsInWaitList,
2420 const cl_event *eventWaitList,
2421 cl_event *event) {
2422 TRACING_ENTER(clEnqueueWriteBuffer, &commandQueue, &buffer, &blockingWrite, &offset, &cb, &ptr, &numEventsInWaitList, &eventWaitList, &event);
2423 cl_int retVal = CL_SUCCESS;
2424 API_ENTER(&retVal);
2425
2426 DBG_LOG_INPUTS("commandQueue", commandQueue, "buffer", buffer, "blockingWrite", blockingWrite,
2427 "offset", offset, "cb", cb, "ptr", ptr,
2428 "numEventsInWaitList", numEventsInWaitList,
2429 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2430 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2431
2432 CommandQueue *pCommandQueue = nullptr;
2433 Buffer *pBuffer = nullptr;
2434
2435 retVal = validateObjects(
2436 WithCastToInternal(commandQueue, &pCommandQueue),
2437 WithCastToInternal(buffer, &pBuffer),
2438 ptr);
2439
2440 if (CL_SUCCESS == retVal) {
2441
2442 if (pBuffer->writeMemObjFlagsInvalid()) {
2443 retVal = CL_INVALID_OPERATION;
2444 TRACING_EXIT(clEnqueueWriteBuffer, &retVal);
2445 return retVal;
2446 }
2447
2448 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
2449 retVal = CL_INVALID_OPERATION;
2450 TRACING_EXIT(clEnqueueWriteBuffer, &retVal);
2451 return retVal;
2452 }
2453
2454 retVal = pCommandQueue->enqueueWriteBuffer(
2455 pBuffer,
2456 blockingWrite,
2457 offset,
2458 cb,
2459 ptr,
2460 nullptr,
2461 numEventsInWaitList,
2462 eventWaitList,
2463 event);
2464 }
2465
2466 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
2467 TRACING_EXIT(clEnqueueWriteBuffer, &retVal);
2468 return retVal;
2469 }
2470
clEnqueueWriteBufferRect(cl_command_queue commandQueue,cl_mem buffer,cl_bool blockingWrite,const size_t * bufferOrigin,const size_t * hostOrigin,const size_t * region,size_t bufferRowPitch,size_t bufferSlicePitch,size_t hostRowPitch,size_t hostSlicePitch,const void * ptr,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2471 cl_int CL_API_CALL clEnqueueWriteBufferRect(cl_command_queue commandQueue,
2472 cl_mem buffer,
2473 cl_bool blockingWrite,
2474 const size_t *bufferOrigin,
2475 const size_t *hostOrigin,
2476 const size_t *region,
2477 size_t bufferRowPitch,
2478 size_t bufferSlicePitch,
2479 size_t hostRowPitch,
2480 size_t hostSlicePitch,
2481 const void *ptr,
2482 cl_uint numEventsInWaitList,
2483 const cl_event *eventWaitList,
2484 cl_event *event) {
2485 TRACING_ENTER(clEnqueueWriteBufferRect, &commandQueue, &buffer, &blockingWrite, &bufferOrigin, &hostOrigin, ®ion, &bufferRowPitch, &bufferSlicePitch, &hostRowPitch, &hostSlicePitch, &ptr, &numEventsInWaitList, &eventWaitList, &event);
2486 cl_int retVal = CL_SUCCESS;
2487 API_ENTER(&retVal);
2488
2489 DBG_LOG_INPUTS("commandQueue", commandQueue, "buffer", buffer, "blockingWrite", blockingWrite,
2490 "bufferOrigin[0]", NEO::FileLoggerInstance().getInput(bufferOrigin, 0), "bufferOrigin[1]", NEO::FileLoggerInstance().getInput(bufferOrigin, 1), "bufferOrigin[2]", NEO::FileLoggerInstance().getInput(bufferOrigin, 2),
2491 "hostOrigin[0]", NEO::FileLoggerInstance().getInput(hostOrigin, 0), "hostOrigin[1]", NEO::FileLoggerInstance().getInput(hostOrigin, 1), "hostOrigin[2]", NEO::FileLoggerInstance().getInput(hostOrigin, 2),
2492 "region[0]", NEO::FileLoggerInstance().getInput(region, 0), "region[1]", NEO::FileLoggerInstance().getInput(region, 1), "region[2]", NEO::FileLoggerInstance().getInput(region, 2),
2493 "bufferRowPitch", bufferRowPitch, "bufferSlicePitch", bufferSlicePitch,
2494 "hostRowPitch", hostRowPitch, "hostSlicePitch", hostSlicePitch, "ptr", ptr,
2495 "numEventsInWaitList", numEventsInWaitList,
2496 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2497 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2498
2499 CommandQueue *pCommandQueue = nullptr;
2500 Buffer *pBuffer = nullptr;
2501
2502 retVal = validateObjects(
2503 WithCastToInternal(commandQueue, &pCommandQueue),
2504 WithCastToInternal(buffer, &pBuffer),
2505 ptr);
2506
2507 if (CL_SUCCESS != retVal) {
2508 TRACING_EXIT(clEnqueueWriteBufferRect, &retVal);
2509 return retVal;
2510 }
2511
2512 if (pBuffer->writeMemObjFlagsInvalid()) {
2513 retVal = CL_INVALID_OPERATION;
2514 TRACING_EXIT(clEnqueueWriteBufferRect, &retVal);
2515 return retVal;
2516 }
2517
2518 if (pBuffer->bufferRectPitchSet(bufferOrigin,
2519 region,
2520 bufferRowPitch,
2521 bufferSlicePitch,
2522 hostRowPitch,
2523 hostSlicePitch,
2524 true) == false) {
2525 retVal = CL_INVALID_VALUE;
2526 TRACING_EXIT(clEnqueueWriteBufferRect, &retVal);
2527 return retVal;
2528 }
2529
2530 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_BUFFER_RECT_INTEL, numEventsInWaitList, eventWaitList, event)) {
2531 retVal = CL_INVALID_OPERATION;
2532 TRACING_EXIT(clEnqueueWriteBufferRect, &retVal);
2533 return retVal;
2534 }
2535
2536 retVal = pCommandQueue->enqueueWriteBufferRect(
2537 pBuffer,
2538 blockingWrite,
2539 bufferOrigin,
2540 hostOrigin,
2541 region,
2542 bufferRowPitch,
2543 bufferSlicePitch,
2544 hostRowPitch,
2545 hostSlicePitch,
2546 ptr,
2547 numEventsInWaitList,
2548 eventWaitList,
2549 event);
2550
2551 TRACING_EXIT(clEnqueueWriteBufferRect, &retVal);
2552 return retVal;
2553 }
2554
clEnqueueFillBuffer(cl_command_queue commandQueue,cl_mem buffer,const void * pattern,size_t patternSize,size_t offset,size_t size,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2555 cl_int CL_API_CALL clEnqueueFillBuffer(cl_command_queue commandQueue,
2556 cl_mem buffer,
2557 const void *pattern,
2558 size_t patternSize,
2559 size_t offset,
2560 size_t size,
2561 cl_uint numEventsInWaitList,
2562 const cl_event *eventWaitList,
2563 cl_event *event) {
2564 TRACING_ENTER(clEnqueueFillBuffer, &commandQueue, &buffer, &pattern, &patternSize, &offset, &size, &numEventsInWaitList, &eventWaitList, &event);
2565 cl_int retVal = CL_SUCCESS;
2566 API_ENTER(&retVal);
2567
2568 DBG_LOG_INPUTS("commandQueue", commandQueue, "buffer", buffer,
2569 "pattern", NEO::FileLoggerInstance().infoPointerToString(pattern, patternSize), "patternSize", patternSize,
2570 "offset", offset, "size", size,
2571 "numEventsInWaitList", numEventsInWaitList,
2572 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2573 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2574
2575 CommandQueue *pCommandQueue = nullptr;
2576 Buffer *pBuffer = nullptr;
2577
2578 retVal = validateObjects(
2579 WithCastToInternal(commandQueue, &pCommandQueue),
2580 WithCastToInternal(buffer, &pBuffer),
2581 pattern,
2582 (PatternSize)patternSize,
2583 EventWaitList(numEventsInWaitList, eventWaitList));
2584
2585 if (CL_SUCCESS == retVal) {
2586 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_FILL_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
2587 retVal = CL_INVALID_OPERATION;
2588 TRACING_EXIT(clEnqueueFillBuffer, &retVal);
2589 return retVal;
2590 }
2591
2592 retVal = pCommandQueue->enqueueFillBuffer(
2593 pBuffer,
2594 pattern,
2595 patternSize,
2596 offset,
2597 size,
2598 numEventsInWaitList,
2599 eventWaitList,
2600 event);
2601 }
2602 TRACING_EXIT(clEnqueueFillBuffer, &retVal);
2603 return retVal;
2604 }
2605
clEnqueueCopyBuffer(cl_command_queue commandQueue,cl_mem srcBuffer,cl_mem dstBuffer,size_t srcOffset,size_t dstOffset,size_t cb,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2606 cl_int CL_API_CALL clEnqueueCopyBuffer(cl_command_queue commandQueue,
2607 cl_mem srcBuffer,
2608 cl_mem dstBuffer,
2609 size_t srcOffset,
2610 size_t dstOffset,
2611 size_t cb,
2612 cl_uint numEventsInWaitList,
2613 const cl_event *eventWaitList,
2614 cl_event *event) {
2615 TRACING_ENTER(clEnqueueCopyBuffer, &commandQueue, &srcBuffer, &dstBuffer, &srcOffset, &dstOffset, &cb, &numEventsInWaitList, &eventWaitList, &event);
2616 cl_int retVal = CL_SUCCESS;
2617 API_ENTER(&retVal);
2618
2619 DBG_LOG_INPUTS("commandQueue", commandQueue, "srcBuffer", srcBuffer, "dstBuffer", dstBuffer,
2620 "srcOffset", srcOffset, "dstOffset", dstOffset, "cb", cb,
2621 "numEventsInWaitList", numEventsInWaitList,
2622 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2623 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2624
2625 CommandQueue *pCommandQueue = nullptr;
2626 Buffer *pSrcBuffer = nullptr;
2627 Buffer *pDstBuffer = nullptr;
2628
2629 retVal = validateObjects(
2630 WithCastToInternal(commandQueue, &pCommandQueue),
2631 WithCastToInternal(srcBuffer, &pSrcBuffer),
2632 WithCastToInternal(dstBuffer, &pDstBuffer));
2633
2634 if (CL_SUCCESS == retVal) {
2635 size_t srcSize = pSrcBuffer->getSize();
2636 size_t dstSize = pDstBuffer->getSize();
2637 if (srcOffset + cb > srcSize || dstOffset + cb > dstSize) {
2638 retVal = CL_INVALID_VALUE;
2639 TRACING_EXIT(clEnqueueCopyBuffer, &retVal);
2640 return retVal;
2641 }
2642
2643 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
2644 retVal = CL_INVALID_OPERATION;
2645 TRACING_EXIT(clEnqueueCopyBuffer, &retVal);
2646 return retVal;
2647 }
2648
2649 retVal = pCommandQueue->enqueueCopyBuffer(
2650 pSrcBuffer,
2651 pDstBuffer,
2652 srcOffset,
2653 dstOffset,
2654 cb,
2655 numEventsInWaitList,
2656 eventWaitList,
2657 event);
2658 }
2659 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
2660 TRACING_EXIT(clEnqueueCopyBuffer, &retVal);
2661 return retVal;
2662 }
2663
clEnqueueCopyBufferRect(cl_command_queue commandQueue,cl_mem srcBuffer,cl_mem dstBuffer,const size_t * srcOrigin,const size_t * dstOrigin,const size_t * region,size_t srcRowPitch,size_t srcSlicePitch,size_t dstRowPitch,size_t dstSlicePitch,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2664 cl_int CL_API_CALL clEnqueueCopyBufferRect(cl_command_queue commandQueue,
2665 cl_mem srcBuffer,
2666 cl_mem dstBuffer,
2667 const size_t *srcOrigin,
2668 const size_t *dstOrigin,
2669 const size_t *region,
2670 size_t srcRowPitch,
2671 size_t srcSlicePitch,
2672 size_t dstRowPitch,
2673 size_t dstSlicePitch,
2674 cl_uint numEventsInWaitList,
2675 const cl_event *eventWaitList,
2676 cl_event *event) {
2677 TRACING_ENTER(clEnqueueCopyBufferRect, &commandQueue, &srcBuffer, &dstBuffer, &srcOrigin, &dstOrigin, ®ion, &srcRowPitch, &srcSlicePitch, &dstRowPitch, &dstSlicePitch, &numEventsInWaitList, &eventWaitList, &event);
2678 cl_int retVal = CL_SUCCESS;
2679 API_ENTER(&retVal);
2680
2681 DBG_LOG_INPUTS("commandQueue", commandQueue, "srcBuffer", srcBuffer, "dstBuffer", dstBuffer,
2682 "srcOrigin[0]", NEO::FileLoggerInstance().getInput(srcOrigin, 0), "srcOrigin[1]", NEO::FileLoggerInstance().getInput(srcOrigin, 1), "srcOrigin[2]", NEO::FileLoggerInstance().getInput(srcOrigin, 2),
2683 "dstOrigin[0]", NEO::FileLoggerInstance().getInput(dstOrigin, 0), "dstOrigin[1]", NEO::FileLoggerInstance().getInput(dstOrigin, 1), "dstOrigin[2]", NEO::FileLoggerInstance().getInput(dstOrigin, 2),
2684 "region[0]", NEO::FileLoggerInstance().getInput(region, 0), "region[1]", NEO::FileLoggerInstance().getInput(region, 1), "region[2]", NEO::FileLoggerInstance().getInput(region, 2),
2685 "srcRowPitch", srcRowPitch, "srcSlicePitch", srcSlicePitch,
2686 "dstRowPitch", dstRowPitch, "dstSlicePitch", dstSlicePitch,
2687 "numEventsInWaitList", numEventsInWaitList,
2688 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2689 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2690
2691 CommandQueue *pCommandQueue = nullptr;
2692 Buffer *pSrcBuffer = nullptr;
2693 Buffer *pDstBuffer = nullptr;
2694
2695 retVal = validateObjects(
2696 WithCastToInternal(commandQueue, &pCommandQueue),
2697 WithCastToInternal(srcBuffer, &pSrcBuffer),
2698 WithCastToInternal(dstBuffer, &pDstBuffer));
2699
2700 if (CL_SUCCESS == retVal) {
2701
2702 if (!pSrcBuffer->bufferRectPitchSet(srcOrigin,
2703 region,
2704 srcRowPitch,
2705 srcSlicePitch,
2706 dstRowPitch,
2707 dstSlicePitch,
2708 true) ||
2709 !pDstBuffer->bufferRectPitchSet(dstOrigin,
2710 region,
2711 srcRowPitch,
2712 srcSlicePitch,
2713 dstRowPitch,
2714 dstSlicePitch,
2715 false)) {
2716 retVal = CL_INVALID_VALUE;
2717 TRACING_EXIT(clEnqueueCopyBufferRect, &retVal);
2718 return retVal;
2719 }
2720 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_BUFFER_RECT_INTEL, numEventsInWaitList, eventWaitList, event)) {
2721 retVal = CL_INVALID_OPERATION;
2722 TRACING_EXIT(clEnqueueCopyBufferRect, &retVal);
2723 return retVal;
2724 }
2725
2726 retVal = pCommandQueue->enqueueCopyBufferRect(
2727 pSrcBuffer,
2728 pDstBuffer,
2729 srcOrigin,
2730 dstOrigin,
2731 region,
2732 srcRowPitch,
2733 srcSlicePitch,
2734 dstRowPitch,
2735 dstSlicePitch,
2736 numEventsInWaitList,
2737 eventWaitList,
2738 event);
2739 }
2740 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
2741 TRACING_EXIT(clEnqueueCopyBufferRect, &retVal);
2742 return retVal;
2743 }
2744
clEnqueueReadImage(cl_command_queue commandQueue,cl_mem image,cl_bool blockingRead,const size_t * origin,const size_t * region,size_t rowPitch,size_t slicePitch,void * ptr,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2745 cl_int CL_API_CALL clEnqueueReadImage(cl_command_queue commandQueue,
2746 cl_mem image,
2747 cl_bool blockingRead,
2748 const size_t *origin,
2749 const size_t *region,
2750 size_t rowPitch,
2751 size_t slicePitch,
2752 void *ptr,
2753 cl_uint numEventsInWaitList,
2754 const cl_event *eventWaitList,
2755 cl_event *event) {
2756 TRACING_ENTER(clEnqueueReadImage, &commandQueue, &image, &blockingRead, &origin, ®ion, &rowPitch, &slicePitch, &ptr, &numEventsInWaitList, &eventWaitList, &event);
2757
2758 CommandQueue *pCommandQueue = nullptr;
2759 Image *pImage = nullptr;
2760
2761 auto retVal = validateObjects(
2762 WithCastToInternal(commandQueue, &pCommandQueue),
2763 WithCastToInternal(image, &pImage));
2764
2765 API_ENTER(&retVal);
2766
2767 DBG_LOG_INPUTS("commandQueue", commandQueue, "image", image, "blockingRead", blockingRead,
2768 "origin[0]", NEO::FileLoggerInstance().getInput(origin, 0), "origin[1]", NEO::FileLoggerInstance().getInput(origin, 1), "origin[2]", NEO::FileLoggerInstance().getInput(origin, 2),
2769 "region[0]", NEO::FileLoggerInstance().getInput(region, 0), "region[1]", NEO::FileLoggerInstance().getInput(region, 1), "region[2]", NEO::FileLoggerInstance().getInput(region, 2),
2770 "rowPitch", rowPitch, "slicePitch", slicePitch, "ptr", ptr,
2771 "numEventsInWaitList", numEventsInWaitList,
2772 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2773 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2774
2775 if (CL_SUCCESS == retVal) {
2776
2777 if (pImage->readMemObjFlagsInvalid()) {
2778 retVal = CL_INVALID_OPERATION;
2779 TRACING_EXIT(clEnqueueReadImage, &retVal);
2780 return retVal;
2781 }
2782 if (isPackedYuvImage(&pImage->getImageFormat())) {
2783 retVal = validateYuvOperation(origin, region);
2784 if (retVal != CL_SUCCESS) {
2785 TRACING_EXIT(clEnqueueReadImage, &retVal);
2786 return retVal;
2787 }
2788 }
2789 retVal = Image::validateRegionAndOrigin(origin, region, pImage->getImageDesc());
2790 if (retVal != CL_SUCCESS) {
2791 TRACING_EXIT(clEnqueueReadImage, &retVal);
2792 return retVal;
2793 }
2794
2795 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_IMAGE_INTEL, numEventsInWaitList, eventWaitList, event)) {
2796 retVal = CL_INVALID_OPERATION;
2797 TRACING_EXIT(clEnqueueReadImage, &retVal);
2798 return retVal;
2799 }
2800
2801 retVal = pCommandQueue->enqueueReadImage(
2802 pImage,
2803 blockingRead,
2804 origin,
2805 region,
2806 rowPitch,
2807 slicePitch,
2808 ptr,
2809 nullptr,
2810 numEventsInWaitList,
2811 eventWaitList,
2812 event);
2813 }
2814 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
2815 TRACING_EXIT(clEnqueueReadImage, &retVal);
2816 return retVal;
2817 }
2818
clEnqueueWriteImage(cl_command_queue commandQueue,cl_mem image,cl_bool blockingWrite,const size_t * origin,const size_t * region,size_t inputRowPitch,size_t inputSlicePitch,const void * ptr,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2819 cl_int CL_API_CALL clEnqueueWriteImage(cl_command_queue commandQueue,
2820 cl_mem image,
2821 cl_bool blockingWrite,
2822 const size_t *origin,
2823 const size_t *region,
2824 size_t inputRowPitch,
2825 size_t inputSlicePitch,
2826 const void *ptr,
2827 cl_uint numEventsInWaitList,
2828 const cl_event *eventWaitList,
2829 cl_event *event) {
2830 TRACING_ENTER(clEnqueueWriteImage, &commandQueue, &image, &blockingWrite, &origin, ®ion, &inputRowPitch, &inputSlicePitch, &ptr, &numEventsInWaitList, &eventWaitList, &event);
2831
2832 CommandQueue *pCommandQueue = nullptr;
2833 Image *pImage = nullptr;
2834
2835 auto retVal = validateObjects(
2836 WithCastToInternal(commandQueue, &pCommandQueue),
2837 WithCastToInternal(image, &pImage));
2838
2839 API_ENTER(&retVal);
2840
2841 DBG_LOG_INPUTS("commandQueue", commandQueue, "image", image, "blockingWrite", blockingWrite,
2842 "origin[0]", NEO::FileLoggerInstance().getInput(origin, 0), "origin[1]", NEO::FileLoggerInstance().getInput(origin, 1), "origin[2]", NEO::FileLoggerInstance().getInput(origin, 2),
2843 "region[0]", NEO::FileLoggerInstance().getInput(region, 0), "region[1]", NEO::FileLoggerInstance().getInput(region, 1), "region[2]", NEO::FileLoggerInstance().getInput(region, 2),
2844 "inputRowPitch", inputRowPitch, "inputSlicePitch", inputSlicePitch, "ptr", ptr,
2845 "numEventsInWaitList", numEventsInWaitList,
2846 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2847 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2848
2849 if (CL_SUCCESS == retVal) {
2850 if (pImage->writeMemObjFlagsInvalid()) {
2851 retVal = CL_INVALID_OPERATION;
2852 TRACING_EXIT(clEnqueueWriteImage, &retVal);
2853 return retVal;
2854 }
2855 if (isPackedYuvImage(&pImage->getImageFormat())) {
2856 retVal = validateYuvOperation(origin, region);
2857 if (retVal != CL_SUCCESS) {
2858 TRACING_EXIT(clEnqueueWriteImage, &retVal);
2859 return retVal;
2860 }
2861 }
2862 retVal = Image::validateRegionAndOrigin(origin, region, pImage->getImageDesc());
2863 if (retVal != CL_SUCCESS) {
2864 TRACING_EXIT(clEnqueueWriteImage, &retVal);
2865 return retVal;
2866 }
2867
2868 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_IMAGE_INTEL, numEventsInWaitList, eventWaitList, event)) {
2869 retVal = CL_INVALID_OPERATION;
2870 TRACING_EXIT(clEnqueueWriteImage, &retVal);
2871 return retVal;
2872 }
2873
2874 retVal = pCommandQueue->enqueueWriteImage(
2875 pImage,
2876 blockingWrite,
2877 origin,
2878 region,
2879 inputRowPitch,
2880 inputSlicePitch,
2881 ptr,
2882 nullptr,
2883 numEventsInWaitList,
2884 eventWaitList,
2885 event);
2886 }
2887 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
2888 TRACING_EXIT(clEnqueueWriteImage, &retVal);
2889 return retVal;
2890 }
2891
clEnqueueFillImage(cl_command_queue commandQueue,cl_mem image,const void * fillColor,const size_t * origin,const size_t * region,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2892 cl_int CL_API_CALL clEnqueueFillImage(cl_command_queue commandQueue,
2893 cl_mem image,
2894 const void *fillColor,
2895 const size_t *origin,
2896 const size_t *region,
2897 cl_uint numEventsInWaitList,
2898 const cl_event *eventWaitList,
2899 cl_event *event) {
2900 TRACING_ENTER(clEnqueueFillImage, &commandQueue, &image, &fillColor, &origin, ®ion, &numEventsInWaitList, &eventWaitList, &event);
2901
2902 CommandQueue *pCommandQueue = nullptr;
2903 Image *dstImage = nullptr;
2904
2905 auto retVal = validateObjects(
2906 WithCastToInternal(commandQueue, &pCommandQueue),
2907 WithCastToInternal(image, &dstImage),
2908 fillColor,
2909 EventWaitList(numEventsInWaitList, eventWaitList));
2910
2911 API_ENTER(&retVal);
2912
2913 DBG_LOG_INPUTS("commandQueue", commandQueue, "image", image, "fillColor", fillColor,
2914 "origin[0]", NEO::FileLoggerInstance().getInput(origin, 0), "origin[1]", NEO::FileLoggerInstance().getInput(origin, 1), "origin[2]", NEO::FileLoggerInstance().getInput(origin, 2),
2915 "region[0]", NEO::FileLoggerInstance().getInput(region, 0), "region[1]", NEO::FileLoggerInstance().getInput(region, 1), "region[2]", NEO::FileLoggerInstance().getInput(region, 2),
2916 "numEventsInWaitList", numEventsInWaitList,
2917 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2918 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2919
2920 if (CL_SUCCESS == retVal) {
2921 retVal = Image::validateRegionAndOrigin(origin, region, dstImage->getImageDesc());
2922 if (retVal != CL_SUCCESS) {
2923 TRACING_EXIT(clEnqueueFillImage, &retVal);
2924 return retVal;
2925 }
2926
2927 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_FILL_IMAGE_INTEL, numEventsInWaitList, eventWaitList, event)) {
2928 retVal = CL_INVALID_OPERATION;
2929 TRACING_EXIT(clEnqueueFillImage, &retVal);
2930 return retVal;
2931 }
2932
2933 retVal = pCommandQueue->enqueueFillImage(
2934 dstImage,
2935 fillColor,
2936 origin,
2937 region,
2938 numEventsInWaitList,
2939 eventWaitList,
2940 event);
2941 }
2942 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
2943 TRACING_EXIT(clEnqueueFillImage, &retVal);
2944 return retVal;
2945 }
2946
clEnqueueCopyImage(cl_command_queue commandQueue,cl_mem srcImage,cl_mem dstImage,const size_t * srcOrigin,const size_t * dstOrigin,const size_t * region,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)2947 cl_int CL_API_CALL clEnqueueCopyImage(cl_command_queue commandQueue,
2948 cl_mem srcImage,
2949 cl_mem dstImage,
2950 const size_t *srcOrigin,
2951 const size_t *dstOrigin,
2952 const size_t *region,
2953 cl_uint numEventsInWaitList,
2954 const cl_event *eventWaitList,
2955 cl_event *event) {
2956 TRACING_ENTER(clEnqueueCopyImage, &commandQueue, &srcImage, &dstImage, &srcOrigin, &dstOrigin, ®ion, &numEventsInWaitList, &eventWaitList, &event);
2957
2958 CommandQueue *pCommandQueue = nullptr;
2959 Image *pSrcImage = nullptr;
2960 Image *pDstImage = nullptr;
2961
2962 auto retVal = validateObjects(WithCastToInternal(commandQueue, &pCommandQueue),
2963 WithCastToInternal(srcImage, &pSrcImage),
2964 WithCastToInternal(dstImage, &pDstImage));
2965
2966 API_ENTER(&retVal);
2967
2968 DBG_LOG_INPUTS("commandQueue", commandQueue, "srcImage", srcImage, "dstImage", dstImage,
2969 "srcOrigin[0]", NEO::FileLoggerInstance().getInput(srcOrigin, 0), "srcOrigin[1]", NEO::FileLoggerInstance().getInput(srcOrigin, 1), "srcOrigin[2]", NEO::FileLoggerInstance().getInput(srcOrigin, 2),
2970 "dstOrigin[0]", NEO::FileLoggerInstance().getInput(dstOrigin, 0), "dstOrigin[1]", NEO::FileLoggerInstance().getInput(dstOrigin, 1), "dstOrigin[2]", NEO::FileLoggerInstance().getInput(dstOrigin, 2),
2971 "region[0]", region ? region[0] : 0, "region[1]", region ? region[1] : 0, "region[2]", region ? region[2] : 0,
2972 "numEventsInWaitList", numEventsInWaitList,
2973 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
2974 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
2975
2976 if (CL_SUCCESS == retVal) {
2977 if (memcmp(&pSrcImage->getImageFormat(), &pDstImage->getImageFormat(), sizeof(cl_image_format))) {
2978 retVal = CL_IMAGE_FORMAT_MISMATCH;
2979 TRACING_EXIT(clEnqueueCopyImage, &retVal);
2980 return retVal;
2981 }
2982 if (isPackedYuvImage(&pSrcImage->getImageFormat())) {
2983 retVal = validateYuvOperation(srcOrigin, region);
2984 if (retVal != CL_SUCCESS) {
2985 TRACING_EXIT(clEnqueueCopyImage, &retVal);
2986 return retVal;
2987 }
2988 }
2989 if (isPackedYuvImage(&pDstImage->getImageFormat())) {
2990 retVal = validateYuvOperation(dstOrigin, region);
2991
2992 if (retVal != CL_SUCCESS) {
2993 TRACING_EXIT(clEnqueueCopyImage, &retVal);
2994 return retVal;
2995 }
2996 if (pDstImage->getImageDesc().image_type == CL_MEM_OBJECT_IMAGE2D && dstOrigin[2] != 0) {
2997 retVal = CL_INVALID_VALUE;
2998 TRACING_EXIT(clEnqueueCopyImage, &retVal);
2999 return retVal;
3000 }
3001 }
3002 retVal = Image::validateRegionAndOrigin(srcOrigin, region, pSrcImage->getImageDesc());
3003 if (retVal != CL_SUCCESS) {
3004 TRACING_EXIT(clEnqueueCopyImage, &retVal);
3005 return retVal;
3006 }
3007 retVal = Image::validateRegionAndOrigin(dstOrigin, region, pDstImage->getImageDesc());
3008 if (retVal != CL_SUCCESS) {
3009 TRACING_EXIT(clEnqueueCopyImage, &retVal);
3010 return retVal;
3011 }
3012
3013 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_IMAGE_INTEL, numEventsInWaitList, eventWaitList, event)) {
3014 retVal = CL_INVALID_OPERATION;
3015 TRACING_EXIT(clEnqueueCopyImage, &retVal);
3016 return retVal;
3017 }
3018
3019 retVal = pCommandQueue->enqueueCopyImage(
3020 pSrcImage,
3021 pDstImage,
3022 srcOrigin,
3023 dstOrigin,
3024 region,
3025 numEventsInWaitList,
3026 eventWaitList,
3027 event);
3028 }
3029 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
3030 TRACING_EXIT(clEnqueueCopyImage, &retVal);
3031 return retVal;
3032 }
3033
clEnqueueCopyImageToBuffer(cl_command_queue commandQueue,cl_mem srcImage,cl_mem dstBuffer,const size_t * srcOrigin,const size_t * region,const size_t dstOffset,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)3034 cl_int CL_API_CALL clEnqueueCopyImageToBuffer(cl_command_queue commandQueue,
3035 cl_mem srcImage,
3036 cl_mem dstBuffer,
3037 const size_t *srcOrigin,
3038 const size_t *region,
3039 const size_t dstOffset,
3040 cl_uint numEventsInWaitList,
3041 const cl_event *eventWaitList,
3042 cl_event *event) {
3043 TRACING_ENTER(clEnqueueCopyImageToBuffer, &commandQueue, &srcImage, &dstBuffer, &srcOrigin, ®ion, (size_t *)&dstOffset, &numEventsInWaitList, &eventWaitList, &event);
3044 cl_int retVal = CL_SUCCESS;
3045 API_ENTER(&retVal);
3046
3047 DBG_LOG_INPUTS("commandQueue", commandQueue, "srcImage", srcImage, "dstBuffer", dstBuffer,
3048 "srcOrigin[0]", NEO::FileLoggerInstance().getInput(srcOrigin, 0), "srcOrigin[1]", NEO::FileLoggerInstance().getInput(srcOrigin, 1), "srcOrigin[2]", NEO::FileLoggerInstance().getInput(srcOrigin, 2),
3049 "region[0]", NEO::FileLoggerInstance().getInput(region, 0), "region[1]", NEO::FileLoggerInstance().getInput(region, 1), "region[2]", NEO::FileLoggerInstance().getInput(region, 2),
3050 "dstOffset", dstOffset,
3051 "numEventsInWaitList", numEventsInWaitList,
3052 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3053 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3054
3055 CommandQueue *pCommandQueue = nullptr;
3056 Image *pSrcImage = nullptr;
3057 Buffer *pDstBuffer = nullptr;
3058
3059 retVal = validateObjects(
3060 WithCastToInternal(commandQueue, &pCommandQueue),
3061 WithCastToInternal(srcImage, &pSrcImage),
3062 WithCastToInternal(dstBuffer, &pDstBuffer));
3063
3064 if (CL_SUCCESS == retVal) {
3065 if (isPackedYuvImage(&pSrcImage->getImageFormat())) {
3066 retVal = validateYuvOperation(srcOrigin, region);
3067 if (retVal != CL_SUCCESS) {
3068 TRACING_EXIT(clEnqueueCopyImageToBuffer, &retVal);
3069 return retVal;
3070 }
3071 }
3072 retVal = Image::validateRegionAndOrigin(srcOrigin, region, pSrcImage->getImageDesc());
3073 if (retVal != CL_SUCCESS) {
3074 TRACING_EXIT(clEnqueueCopyImageToBuffer, &retVal);
3075 return retVal;
3076 }
3077
3078 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_IMAGE_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
3079 retVal = CL_INVALID_OPERATION;
3080 TRACING_EXIT(clEnqueueCopyImageToBuffer, &retVal);
3081 return retVal;
3082 }
3083
3084 retVal = pCommandQueue->enqueueCopyImageToBuffer(
3085 pSrcImage,
3086 pDstBuffer,
3087 srcOrigin,
3088 region,
3089 dstOffset,
3090 numEventsInWaitList,
3091 eventWaitList,
3092 event);
3093 }
3094
3095 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
3096 TRACING_EXIT(clEnqueueCopyImageToBuffer, &retVal);
3097 return retVal;
3098 }
3099
clEnqueueCopyBufferToImage(cl_command_queue commandQueue,cl_mem srcBuffer,cl_mem dstImage,size_t srcOffset,const size_t * dstOrigin,const size_t * region,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)3100 cl_int CL_API_CALL clEnqueueCopyBufferToImage(cl_command_queue commandQueue,
3101 cl_mem srcBuffer,
3102 cl_mem dstImage,
3103 size_t srcOffset,
3104 const size_t *dstOrigin,
3105 const size_t *region,
3106 cl_uint numEventsInWaitList,
3107 const cl_event *eventWaitList,
3108 cl_event *event) {
3109 TRACING_ENTER(clEnqueueCopyBufferToImage, &commandQueue, &srcBuffer, &dstImage, &srcOffset, &dstOrigin, ®ion, &numEventsInWaitList, &eventWaitList, &event);
3110 cl_int retVal = CL_SUCCESS;
3111 API_ENTER(&retVal);
3112
3113 DBG_LOG_INPUTS("commandQueue", commandQueue, "srcBuffer", srcBuffer, "dstImage", dstImage, "srcOffset", srcOffset,
3114 "dstOrigin[0]", NEO::FileLoggerInstance().getInput(dstOrigin, 0), "dstOrigin[1]", NEO::FileLoggerInstance().getInput(dstOrigin, 1), "dstOrigin[2]", NEO::FileLoggerInstance().getInput(dstOrigin, 2),
3115 "region[0]", NEO::FileLoggerInstance().getInput(region, 0), "region[1]", NEO::FileLoggerInstance().getInput(region, 1), "region[2]", NEO::FileLoggerInstance().getInput(region, 2),
3116 "numEventsInWaitList", numEventsInWaitList,
3117 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3118 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3119
3120 CommandQueue *pCommandQueue = nullptr;
3121 Buffer *pSrcBuffer = nullptr;
3122 Image *pDstImage = nullptr;
3123
3124 retVal = validateObjects(
3125 WithCastToInternal(commandQueue, &pCommandQueue),
3126 WithCastToInternal(srcBuffer, &pSrcBuffer),
3127 WithCastToInternal(dstImage, &pDstImage));
3128
3129 if (CL_SUCCESS == retVal) {
3130 if (isPackedYuvImage(&pDstImage->getImageFormat())) {
3131 retVal = validateYuvOperation(dstOrigin, region);
3132 if (retVal != CL_SUCCESS) {
3133 TRACING_EXIT(clEnqueueCopyBufferToImage, &retVal);
3134 return retVal;
3135 }
3136 }
3137 retVal = Image::validateRegionAndOrigin(dstOrigin, region, pDstImage->getImageDesc());
3138 if (retVal != CL_SUCCESS) {
3139 TRACING_EXIT(clEnqueueCopyBufferToImage, &retVal);
3140 return retVal;
3141 }
3142
3143 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_BUFFER_IMAGE_INTEL, numEventsInWaitList, eventWaitList, event)) {
3144 retVal = CL_INVALID_OPERATION;
3145 TRACING_EXIT(clEnqueueCopyBufferToImage, &retVal);
3146 return retVal;
3147 }
3148
3149 retVal = pCommandQueue->enqueueCopyBufferToImage(
3150 pSrcBuffer,
3151 pDstImage,
3152 srcOffset,
3153 dstOrigin,
3154 region,
3155 numEventsInWaitList,
3156 eventWaitList,
3157 event);
3158 }
3159
3160 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
3161 TRACING_EXIT(clEnqueueCopyBufferToImage, &retVal);
3162 return retVal;
3163 }
3164
clEnqueueMapBuffer(cl_command_queue commandQueue,cl_mem buffer,cl_bool blockingMap,cl_map_flags mapFlags,size_t offset,size_t cb,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event,cl_int * errcodeRet)3165 void *CL_API_CALL clEnqueueMapBuffer(cl_command_queue commandQueue,
3166 cl_mem buffer,
3167 cl_bool blockingMap,
3168 cl_map_flags mapFlags,
3169 size_t offset,
3170 size_t cb,
3171 cl_uint numEventsInWaitList,
3172 const cl_event *eventWaitList,
3173 cl_event *event,
3174 cl_int *errcodeRet) {
3175 TRACING_ENTER(clEnqueueMapBuffer, &commandQueue, &buffer, &blockingMap, &mapFlags, &offset, &cb, &numEventsInWaitList, &eventWaitList, &event, &errcodeRet);
3176 void *retPtr = nullptr;
3177 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
3178 cl_int retVal;
3179 API_ENTER(&retVal);
3180 DBG_LOG_INPUTS("commandQueue", commandQueue, "buffer", buffer, "blockingMap", blockingMap,
3181 "mapFlags", mapFlags, "offset", offset, "cb", cb,
3182 "numEventsInWaitList", numEventsInWaitList,
3183 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3184 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3185
3186 do {
3187 auto pCommandQueue = castToObject<CommandQueue>(commandQueue);
3188 if (!pCommandQueue) {
3189 retVal = CL_INVALID_COMMAND_QUEUE;
3190 break;
3191 }
3192
3193 auto pBuffer = castToObject<Buffer>(buffer);
3194 if (!pBuffer) {
3195 retVal = CL_INVALID_MEM_OBJECT;
3196 break;
3197 }
3198
3199 if (pBuffer->mapMemObjFlagsInvalid(mapFlags)) {
3200 retVal = CL_INVALID_OPERATION;
3201 break;
3202 }
3203
3204 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_MAP_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
3205 retVal = CL_INVALID_OPERATION;
3206 break;
3207 }
3208
3209 retPtr = pCommandQueue->enqueueMapBuffer(
3210 pBuffer,
3211 blockingMap,
3212 mapFlags,
3213 offset,
3214 cb,
3215 numEventsInWaitList,
3216 eventWaitList,
3217 event,
3218 retVal);
3219
3220 } while (false);
3221
3222 err.set(retVal);
3223 DBG_LOG_INPUTS("retPtr", retPtr, "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
3224
3225 TRACING_EXIT(clEnqueueMapBuffer, &retPtr);
3226 return retPtr;
3227 }
3228
clEnqueueMapImage(cl_command_queue commandQueue,cl_mem image,cl_bool blockingMap,cl_map_flags mapFlags,const size_t * origin,const size_t * region,size_t * imageRowPitch,size_t * imageSlicePitch,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event,cl_int * errcodeRet)3229 void *CL_API_CALL clEnqueueMapImage(cl_command_queue commandQueue,
3230 cl_mem image,
3231 cl_bool blockingMap,
3232 cl_map_flags mapFlags,
3233 const size_t *origin,
3234 const size_t *region,
3235 size_t *imageRowPitch,
3236 size_t *imageSlicePitch,
3237 cl_uint numEventsInWaitList,
3238 const cl_event *eventWaitList,
3239 cl_event *event,
3240 cl_int *errcodeRet) {
3241 TRACING_ENTER(clEnqueueMapImage, &commandQueue, &image, &blockingMap, &mapFlags, &origin, ®ion, &imageRowPitch, &imageSlicePitch, &numEventsInWaitList, &eventWaitList, &event, &errcodeRet);
3242
3243 void *retPtr = nullptr;
3244 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
3245 cl_int retVal;
3246
3247 API_ENTER(&retVal);
3248
3249 DBG_LOG_INPUTS("commandQueue", commandQueue, "image", image,
3250 "blockingMap", blockingMap, "mapFlags", mapFlags,
3251 "origin[0]", NEO::FileLoggerInstance().getInput(origin, 0), "origin[1]", NEO::FileLoggerInstance().getInput(origin, 1),
3252 "origin[2]", NEO::FileLoggerInstance().getInput(origin, 2), "region[0]", NEO::FileLoggerInstance().getInput(region, 0),
3253 "region[1]", NEO::FileLoggerInstance().getInput(region, 1), "region[2]", NEO::FileLoggerInstance().getInput(region, 2),
3254 "imageRowPitch", NEO::FileLoggerInstance().getInput(imageRowPitch, 0),
3255 "imageSlicePitch", NEO::FileLoggerInstance().getInput(imageSlicePitch, 0),
3256 "numEventsInWaitList", numEventsInWaitList,
3257 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3258 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3259
3260 do {
3261 Image *pImage = nullptr;
3262 CommandQueue *pCommandQueue = nullptr;
3263 retVal = validateObjects(
3264 WithCastToInternal(commandQueue, &pCommandQueue),
3265 WithCastToInternal(image, &pImage));
3266
3267 if (retVal != CL_SUCCESS) {
3268 break;
3269 }
3270
3271 if (pImage->mapMemObjFlagsInvalid(mapFlags)) {
3272 retVal = CL_INVALID_OPERATION;
3273 break;
3274 }
3275 if (isPackedYuvImage(&pImage->getImageFormat())) {
3276 retVal = validateYuvOperation(origin, region);
3277 if (retVal != CL_SUCCESS) {
3278 break;
3279 }
3280 }
3281
3282 retVal = Image::validateRegionAndOrigin(origin, region, pImage->getImageDesc());
3283 if (retVal != CL_SUCCESS) {
3284 break;
3285 }
3286
3287 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_MAP_IMAGE_INTEL, numEventsInWaitList, eventWaitList, event)) {
3288 retVal = CL_INVALID_OPERATION;
3289 break;
3290 }
3291
3292 retPtr = pCommandQueue->enqueueMapImage(
3293 pImage,
3294 blockingMap,
3295 mapFlags,
3296 origin,
3297 region,
3298 imageRowPitch,
3299 imageSlicePitch,
3300 numEventsInWaitList,
3301 eventWaitList,
3302 event,
3303 retVal);
3304
3305 } while (false);
3306
3307 err.set(retVal);
3308 DBG_LOG_INPUTS("retPtr", retPtr, "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
3309
3310 TRACING_EXIT(clEnqueueMapImage, &retPtr);
3311 return retPtr;
3312 }
3313
clEnqueueUnmapMemObject(cl_command_queue commandQueue,cl_mem memObj,void * mappedPtr,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)3314 cl_int CL_API_CALL clEnqueueUnmapMemObject(cl_command_queue commandQueue,
3315 cl_mem memObj,
3316 void *mappedPtr,
3317 cl_uint numEventsInWaitList,
3318 const cl_event *eventWaitList,
3319 cl_event *event) {
3320 TRACING_ENTER(clEnqueueUnmapMemObject, &commandQueue, &memObj, &mappedPtr, &numEventsInWaitList, &eventWaitList, &event);
3321
3322 CommandQueue *pCommandQueue = nullptr;
3323 MemObj *pMemObj = nullptr;
3324
3325 cl_int retVal = validateObjects(
3326 WithCastToInternal(commandQueue, &pCommandQueue),
3327 WithCastToInternal(memObj, &pMemObj));
3328
3329 API_ENTER(&retVal);
3330
3331 DBG_LOG_INPUTS("commandQueue", commandQueue,
3332 "memObj", memObj,
3333 "mappedPtr", mappedPtr,
3334 "numEventsInWaitList", numEventsInWaitList,
3335 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3336 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3337
3338 if (retVal == CL_SUCCESS) {
3339 cl_command_queue_capabilities_intel requiredCapability = 0u;
3340 switch (pMemObj->peekClMemObjType()) {
3341 case CL_MEM_OBJECT_BUFFER:
3342 requiredCapability = CL_QUEUE_CAPABILITY_MAP_BUFFER_INTEL;
3343 break;
3344 case CL_MEM_OBJECT_IMAGE2D:
3345 case CL_MEM_OBJECT_IMAGE3D:
3346 case CL_MEM_OBJECT_IMAGE2D_ARRAY:
3347 case CL_MEM_OBJECT_IMAGE1D:
3348 case CL_MEM_OBJECT_IMAGE1D_ARRAY:
3349 case CL_MEM_OBJECT_IMAGE1D_BUFFER:
3350 requiredCapability = CL_QUEUE_CAPABILITY_MAP_IMAGE_INTEL;
3351 break;
3352 default:
3353 retVal = CL_INVALID_MEM_OBJECT;
3354 TRACING_EXIT(clEnqueueUnmapMemObject, &retVal);
3355 return retVal;
3356 }
3357
3358 if (!pCommandQueue->validateCapabilityForOperation(requiredCapability, numEventsInWaitList, eventWaitList, event)) {
3359 retVal = CL_INVALID_OPERATION;
3360 TRACING_EXIT(clEnqueueUnmapMemObject, &retVal);
3361 return retVal;
3362 }
3363
3364 retVal = pCommandQueue->enqueueUnmapMemObject(pMemObj, mappedPtr, numEventsInWaitList, eventWaitList, event);
3365 }
3366
3367 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
3368 TRACING_EXIT(clEnqueueUnmapMemObject, &retVal);
3369 return retVal;
3370 }
3371
clEnqueueMigrateMemObjects(cl_command_queue commandQueue,cl_uint numMemObjects,const cl_mem * memObjects,cl_mem_migration_flags flags,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)3372 cl_int CL_API_CALL clEnqueueMigrateMemObjects(cl_command_queue commandQueue,
3373 cl_uint numMemObjects,
3374 const cl_mem *memObjects,
3375 cl_mem_migration_flags flags,
3376 cl_uint numEventsInWaitList,
3377 const cl_event *eventWaitList,
3378 cl_event *event) {
3379 TRACING_ENTER(clEnqueueMigrateMemObjects, &commandQueue, &numMemObjects, &memObjects, &flags, &numEventsInWaitList, &eventWaitList, &event);
3380 cl_int retVal = CL_SUCCESS;
3381 API_ENTER(&retVal);
3382
3383 DBG_LOG_INPUTS("commandQueue", commandQueue,
3384 "numMemObjects", numMemObjects,
3385 "memObjects", memObjects,
3386 "flags", flags,
3387 "numEventsInWaitList", numEventsInWaitList,
3388 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3389 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3390
3391 CommandQueue *pCommandQueue = nullptr;
3392 retVal = validateObjects(
3393 WithCastToInternal(commandQueue, &pCommandQueue),
3394 EventWaitList(numEventsInWaitList, eventWaitList));
3395
3396 if (CL_SUCCESS != retVal) {
3397 TRACING_EXIT(clEnqueueMigrateMemObjects, &retVal);
3398 return retVal;
3399 }
3400
3401 if (numMemObjects == 0 || memObjects == nullptr) {
3402 retVal = CL_INVALID_VALUE;
3403 TRACING_EXIT(clEnqueueMigrateMemObjects, &retVal);
3404 return retVal;
3405 }
3406
3407 for (unsigned int object = 0; object < numMemObjects; object++) {
3408 auto memObject = castToObject<MemObj>(memObjects[object]);
3409 if (!memObject) {
3410 retVal = CL_INVALID_MEM_OBJECT;
3411 TRACING_EXIT(clEnqueueMigrateMemObjects, &retVal);
3412 return retVal;
3413 }
3414 }
3415
3416 const cl_mem_migration_flags allValidFlags = CL_MIGRATE_MEM_OBJECT_CONTENT_UNDEFINED | CL_MIGRATE_MEM_OBJECT_HOST;
3417
3418 if ((flags & (~allValidFlags)) != 0) {
3419 retVal = CL_INVALID_VALUE;
3420 TRACING_EXIT(clEnqueueMigrateMemObjects, &retVal);
3421 return retVal;
3422 }
3423
3424 retVal = pCommandQueue->enqueueMigrateMemObjects(numMemObjects,
3425 memObjects,
3426 flags,
3427 numEventsInWaitList,
3428 eventWaitList,
3429 event);
3430 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
3431 TRACING_EXIT(clEnqueueMigrateMemObjects, &retVal);
3432 return retVal;
3433 }
3434
clEnqueueNDRangeKernel(cl_command_queue commandQueue,cl_kernel kernel,cl_uint workDim,const size_t * globalWorkOffset,const size_t * globalWorkSize,const size_t * localWorkSize,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)3435 cl_int CL_API_CALL clEnqueueNDRangeKernel(cl_command_queue commandQueue,
3436 cl_kernel kernel,
3437 cl_uint workDim,
3438 const size_t *globalWorkOffset,
3439 const size_t *globalWorkSize,
3440 const size_t *localWorkSize,
3441 cl_uint numEventsInWaitList,
3442 const cl_event *eventWaitList,
3443 cl_event *event) {
3444 TRACING_ENTER(clEnqueueNDRangeKernel, &commandQueue, &kernel, &workDim, &globalWorkOffset, &globalWorkSize, &localWorkSize, &numEventsInWaitList, &eventWaitList, &event);
3445 cl_int retVal = CL_SUCCESS;
3446 API_ENTER(&retVal);
3447 DBG_LOG_INPUTS("commandQueue", commandQueue, "cl_kernel", kernel,
3448 "globalWorkOffset[0]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 0),
3449 "globalWorkOffset[1]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 1),
3450 "globalWorkOffset[2]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 2),
3451 "globalWorkSize", NEO::FileLoggerInstance().getSizes(globalWorkSize, workDim, false),
3452 "localWorkSize", NEO::FileLoggerInstance().getSizes(localWorkSize, workDim, true),
3453 "numEventsInWaitList", numEventsInWaitList,
3454 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3455 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3456
3457 CommandQueue *pCommandQueue = nullptr;
3458 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
3459
3460 retVal = validateObjects(
3461 WithCastToInternal(commandQueue, &pCommandQueue),
3462 WithCastToInternal(kernel, &pMultiDeviceKernel),
3463 EventWaitList(numEventsInWaitList, eventWaitList));
3464
3465 if (CL_SUCCESS != retVal) {
3466 TRACING_EXIT(clEnqueueNDRangeKernel, &retVal);
3467 return retVal;
3468 }
3469
3470 Kernel *pKernel = pMultiDeviceKernel->getKernel(pCommandQueue->getDevice().getRootDeviceIndex());
3471 if ((pKernel->getExecutionType() != KernelExecutionType::Default) ||
3472 pKernel->usesSyncBuffer()) {
3473 retVal = CL_INVALID_KERNEL;
3474 TRACING_EXIT(clEnqueueNDRangeKernel, &retVal);
3475 return retVal;
3476 }
3477
3478 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_KERNEL_INTEL, numEventsInWaitList, eventWaitList, event)) {
3479 retVal = CL_INVALID_OPERATION;
3480 TRACING_EXIT(clEnqueueNDRangeKernel, &retVal);
3481 return retVal;
3482 }
3483
3484 TakeOwnershipWrapper<MultiDeviceKernel> kernelOwnership(*pMultiDeviceKernel, gtpinIsGTPinInitialized());
3485 if (gtpinIsGTPinInitialized()) {
3486 gtpinNotifyKernelSubmit(kernel, pCommandQueue);
3487 }
3488
3489 retVal = pCommandQueue->enqueueKernel(
3490 pKernel,
3491 workDim,
3492 globalWorkOffset,
3493 globalWorkSize,
3494 localWorkSize,
3495 numEventsInWaitList,
3496 eventWaitList,
3497 event);
3498
3499 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
3500 TRACING_EXIT(clEnqueueNDRangeKernel, &retVal);
3501 return retVal;
3502 }
3503
clEnqueueTask(cl_command_queue commandQueue,cl_kernel kernel,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)3504 cl_int CL_API_CALL clEnqueueTask(cl_command_queue commandQueue,
3505 cl_kernel kernel,
3506 cl_uint numEventsInWaitList,
3507 const cl_event *eventWaitList,
3508 cl_event *event) {
3509 TRACING_ENTER(clEnqueueTask, &commandQueue, &kernel, &numEventsInWaitList, &eventWaitList, &event);
3510 cl_int retVal = CL_SUCCESS;
3511 API_ENTER(&retVal);
3512 DBG_LOG_INPUTS("commandQueue", commandQueue, "kernel", kernel,
3513 "numEventsInWaitList", numEventsInWaitList,
3514 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3515 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3516 cl_uint workDim = 3;
3517 size_t *globalWorkOffset = nullptr;
3518 size_t globalWorkSize[3] = {1, 1, 1};
3519 size_t localWorkSize[3] = {1, 1, 1};
3520 retVal = (clEnqueueNDRangeKernel(
3521 commandQueue,
3522 kernel,
3523 workDim,
3524 globalWorkOffset,
3525 globalWorkSize,
3526 localWorkSize,
3527 numEventsInWaitList,
3528 eventWaitList,
3529 event));
3530 TRACING_EXIT(clEnqueueTask, &retVal);
3531 return retVal;
3532 }
3533
clEnqueueNativeKernel(cl_command_queue commandQueue,void (CL_CALLBACK * userFunc)(void *),void * args,size_t cbArgs,cl_uint numMemObjects,const cl_mem * memList,const void ** argsMemLoc,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)3534 cl_int CL_API_CALL clEnqueueNativeKernel(cl_command_queue commandQueue,
3535 void(CL_CALLBACK *userFunc)(void *),
3536 void *args,
3537 size_t cbArgs,
3538 cl_uint numMemObjects,
3539 const cl_mem *memList,
3540 const void **argsMemLoc,
3541 cl_uint numEventsInWaitList,
3542 const cl_event *eventWaitList,
3543 cl_event *event) {
3544 TRACING_ENTER(clEnqueueNativeKernel, &commandQueue, &userFunc, &args, &cbArgs, &numMemObjects, &memList, &argsMemLoc, &numEventsInWaitList, &eventWaitList, &event);
3545 cl_int retVal = CL_OUT_OF_HOST_MEMORY;
3546 API_ENTER(&retVal);
3547 DBG_LOG_INPUTS("commandQueue", commandQueue, "userFunc", userFunc, "args", args,
3548 "cbArgs", cbArgs, "numMemObjects", numMemObjects, "memList", memList, "argsMemLoc", argsMemLoc,
3549 "numEventsInWaitList", numEventsInWaitList,
3550 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3551 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3552
3553 TRACING_EXIT(clEnqueueNativeKernel, &retVal);
3554 return retVal;
3555 }
3556
3557 // deprecated OpenCL 1.1
clEnqueueMarker(cl_command_queue commandQueue,cl_event * event)3558 cl_int CL_API_CALL clEnqueueMarker(cl_command_queue commandQueue,
3559 cl_event *event) {
3560 TRACING_ENTER(clEnqueueMarker, &commandQueue, &event);
3561 cl_int retVal = CL_SUCCESS;
3562 API_ENTER(&retVal);
3563 DBG_LOG_INPUTS("commandQueue", commandQueue, "cl_event", event);
3564
3565 auto pCommandQueue = castToObject<CommandQueue>(commandQueue);
3566 if (pCommandQueue) {
3567 if (!pCommandQueue->validateCapability(CL_QUEUE_CAPABILITY_MARKER_INTEL)) {
3568 retVal = CL_INVALID_OPERATION;
3569 TRACING_EXIT(clEnqueueMarker, &retVal);
3570 return retVal;
3571 }
3572
3573 retVal = pCommandQueue->enqueueMarkerWithWaitList(
3574 0,
3575 nullptr,
3576 event);
3577 TRACING_EXIT(clEnqueueMarker, &retVal);
3578 return retVal;
3579 }
3580 retVal = CL_INVALID_COMMAND_QUEUE;
3581 TRACING_EXIT(clEnqueueMarker, &retVal);
3582 return retVal;
3583 }
3584
3585 // deprecated OpenCL 1.1
clEnqueueWaitForEvents(cl_command_queue commandQueue,cl_uint numEvents,const cl_event * eventList)3586 cl_int CL_API_CALL clEnqueueWaitForEvents(cl_command_queue commandQueue,
3587 cl_uint numEvents,
3588 const cl_event *eventList) {
3589 TRACING_ENTER(clEnqueueWaitForEvents, &commandQueue, &numEvents, &eventList);
3590 cl_int retVal = CL_SUCCESS;
3591 API_ENTER(&retVal);
3592 DBG_LOG_INPUTS("commandQueue", commandQueue, "eventList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventList), numEvents));
3593
3594 auto pCommandQueue = castToObject<CommandQueue>(commandQueue);
3595 if (!pCommandQueue) {
3596 retVal = CL_INVALID_COMMAND_QUEUE;
3597 TRACING_EXIT(clEnqueueWaitForEvents, &retVal);
3598 return retVal;
3599 }
3600 for (unsigned int i = 0; i < numEvents && retVal == CL_SUCCESS; i++) {
3601 retVal = validateObjects(eventList[i]);
3602 }
3603
3604 if (retVal != CL_SUCCESS) {
3605 TRACING_EXIT(clEnqueueWaitForEvents, &retVal);
3606 return retVal;
3607 }
3608
3609 if (!pCommandQueue->validateCapabilitiesForEventWaitList(numEvents, eventList)) {
3610 retVal = CL_INVALID_OPERATION;
3611 TRACING_EXIT(clEnqueueWaitForEvents, &retVal);
3612 return retVal;
3613 }
3614
3615 retVal = Event::waitForEvents(numEvents, eventList);
3616
3617 TRACING_EXIT(clEnqueueWaitForEvents, &retVal);
3618 return retVal;
3619 }
3620
3621 // deprecated OpenCL 1.1
clEnqueueBarrier(cl_command_queue commandQueue)3622 cl_int CL_API_CALL clEnqueueBarrier(cl_command_queue commandQueue) {
3623 TRACING_ENTER(clEnqueueBarrier, &commandQueue);
3624 cl_int retVal = CL_SUCCESS;
3625 API_ENTER(&retVal);
3626 DBG_LOG_INPUTS("commandQueue", commandQueue);
3627 auto pCommandQueue = castToObject<CommandQueue>(commandQueue);
3628 if (pCommandQueue) {
3629 if (!pCommandQueue->validateCapability(CL_QUEUE_CAPABILITY_BARRIER_INTEL)) {
3630 retVal = CL_INVALID_OPERATION;
3631 TRACING_EXIT(clEnqueueBarrier, &retVal);
3632 return retVal;
3633 }
3634
3635 retVal = pCommandQueue->enqueueBarrierWithWaitList(
3636 0,
3637 nullptr,
3638 nullptr);
3639 TRACING_EXIT(clEnqueueBarrier, &retVal);
3640 return retVal;
3641 }
3642 retVal = CL_INVALID_COMMAND_QUEUE;
3643 TRACING_EXIT(clEnqueueBarrier, &retVal);
3644 return retVal;
3645 }
3646
clEnqueueMarkerWithWaitList(cl_command_queue commandQueue,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)3647 cl_int CL_API_CALL clEnqueueMarkerWithWaitList(cl_command_queue commandQueue,
3648 cl_uint numEventsInWaitList,
3649 const cl_event *eventWaitList,
3650 cl_event *event) {
3651 TRACING_ENTER(clEnqueueMarkerWithWaitList, &commandQueue, &numEventsInWaitList, &eventWaitList, &event);
3652 cl_int retVal = CL_SUCCESS;
3653 API_ENTER(&retVal);
3654 DBG_LOG_INPUTS("cl_command_queue", commandQueue,
3655 "numEventsInWaitList", numEventsInWaitList,
3656 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3657 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3658
3659 CommandQueue *pCommandQueue = nullptr;
3660 retVal = validateObjects(
3661 WithCastToInternal(commandQueue, &pCommandQueue),
3662 EventWaitList(numEventsInWaitList, eventWaitList));
3663
3664 if (CL_SUCCESS != retVal) {
3665 TRACING_EXIT(clEnqueueMarkerWithWaitList, &retVal);
3666 return retVal;
3667 }
3668
3669 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_MARKER_INTEL, numEventsInWaitList, eventWaitList, event)) {
3670 retVal = CL_INVALID_OPERATION;
3671 TRACING_EXIT(clEnqueueMarkerWithWaitList, &retVal);
3672 return retVal;
3673 }
3674
3675 retVal = pCommandQueue->enqueueMarkerWithWaitList(
3676 numEventsInWaitList,
3677 eventWaitList,
3678 event);
3679 TRACING_EXIT(clEnqueueMarkerWithWaitList, &retVal);
3680 return retVal;
3681 }
3682
clEnqueueBarrierWithWaitList(cl_command_queue commandQueue,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)3683 cl_int CL_API_CALL clEnqueueBarrierWithWaitList(cl_command_queue commandQueue,
3684 cl_uint numEventsInWaitList,
3685 const cl_event *eventWaitList,
3686 cl_event *event) {
3687 TRACING_ENTER(clEnqueueBarrierWithWaitList, &commandQueue, &numEventsInWaitList, &eventWaitList, &event);
3688 cl_int retVal = CL_SUCCESS;
3689 API_ENTER(&retVal);
3690 DBG_LOG_INPUTS("cl_command_queue", commandQueue,
3691 "numEventsInWaitList", numEventsInWaitList,
3692 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
3693 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
3694
3695 CommandQueue *pCommandQueue = nullptr;
3696
3697 retVal = validateObjects(
3698 WithCastToInternal(commandQueue, &pCommandQueue),
3699 EventWaitList(numEventsInWaitList, eventWaitList));
3700
3701 if (CL_SUCCESS != retVal) {
3702 TRACING_EXIT(clEnqueueBarrierWithWaitList, &retVal);
3703 return retVal;
3704 }
3705
3706 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_BARRIER_INTEL, numEventsInWaitList, eventWaitList, event)) {
3707 retVal = CL_INVALID_OPERATION;
3708 TRACING_EXIT(clEnqueueBarrierWithWaitList, &retVal);
3709 return retVal;
3710 }
3711
3712 retVal = pCommandQueue->enqueueBarrierWithWaitList(
3713 numEventsInWaitList,
3714 eventWaitList,
3715 event);
3716 TRACING_EXIT(clEnqueueBarrierWithWaitList, &retVal);
3717 return retVal;
3718 }
3719
3720 CL_API_ENTRY cl_command_queue CL_API_CALL
clCreatePerfCountersCommandQueueINTEL(cl_context context,cl_device_id device,cl_command_queue_properties properties,cl_uint configuration,cl_int * errcodeRet)3721 clCreatePerfCountersCommandQueueINTEL(
3722 cl_context context,
3723 cl_device_id device,
3724 cl_command_queue_properties properties,
3725 cl_uint configuration,
3726 cl_int *errcodeRet) {
3727 API_ENTER(nullptr);
3728
3729 DBG_LOG_INPUTS("context", context,
3730 "device", device,
3731 "properties", properties,
3732 "configuration", configuration);
3733
3734 cl_command_queue commandQueue = nullptr;
3735 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
3736
3737 ClDevice *pDevice = nullptr;
3738 WithCastToInternal(device, &pDevice);
3739 if (pDevice == nullptr) {
3740 err.set(CL_INVALID_DEVICE);
3741 return commandQueue;
3742 }
3743
3744 if (!pDevice->getHardwareInfo().capabilityTable.instrumentationEnabled) {
3745 err.set(CL_INVALID_DEVICE);
3746 return commandQueue;
3747 }
3748
3749 if ((properties & CL_QUEUE_PROFILING_ENABLE) == 0) {
3750 err.set(CL_INVALID_QUEUE_PROPERTIES);
3751 return commandQueue;
3752 }
3753 if ((properties & CL_QUEUE_ON_DEVICE) != 0) {
3754 err.set(CL_INVALID_QUEUE_PROPERTIES);
3755 return commandQueue;
3756 }
3757 if ((properties & CL_QUEUE_ON_DEVICE_DEFAULT) != 0) {
3758 err.set(CL_INVALID_QUEUE_PROPERTIES);
3759 return commandQueue;
3760 }
3761
3762 if (configuration != 0) {
3763 err.set(CL_INVALID_OPERATION);
3764 return commandQueue;
3765 }
3766
3767 commandQueue = clCreateCommandQueue(context, device, properties, errcodeRet);
3768 if (commandQueue != nullptr) {
3769 auto commandQueueObject = castToObjectOrAbort<CommandQueue>(commandQueue);
3770
3771 if (!commandQueueObject->setPerfCountersEnabled()) {
3772 clReleaseCommandQueue(commandQueue);
3773 commandQueue = nullptr;
3774 err.set(CL_OUT_OF_RESOURCES);
3775 }
3776 }
3777 return commandQueue;
3778 }
3779
3780 CL_API_ENTRY cl_int CL_API_CALL
clSetPerformanceConfigurationINTEL(cl_device_id device,cl_uint count,cl_uint * offsets,cl_uint * values)3781 clSetPerformanceConfigurationINTEL(
3782 cl_device_id device,
3783 cl_uint count,
3784 cl_uint *offsets,
3785 cl_uint *values) {
3786 // Not supported, covered by Metric Library DLL.
3787 return CL_INVALID_OPERATION;
3788 }
3789
clHostMemAllocINTEL(cl_context context,const cl_mem_properties_intel * properties,size_t size,cl_uint alignment,cl_int * errcodeRet)3790 void *clHostMemAllocINTEL(
3791 cl_context context,
3792 const cl_mem_properties_intel *properties,
3793 size_t size,
3794 cl_uint alignment,
3795 cl_int *errcodeRet) {
3796
3797 Context *neoContext = nullptr;
3798
3799 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
3800
3801 auto retVal = validateObjects(WithCastToInternal(context, &neoContext));
3802
3803 if (retVal != CL_SUCCESS) {
3804 err.set(retVal);
3805 return nullptr;
3806 }
3807
3808 SVMAllocsManager::UnifiedMemoryProperties unifiedMemoryProperties(InternalMemoryType::HOST_UNIFIED_MEMORY,
3809 neoContext->getRootDeviceIndices(), neoContext->getDeviceBitfields());
3810 cl_mem_flags flags = 0;
3811 cl_mem_flags_intel flagsIntel = 0;
3812 cl_mem_alloc_flags_intel allocflags = 0;
3813 if (!ClMemoryPropertiesHelper::parseMemoryProperties(properties, unifiedMemoryProperties.allocationFlags, flags, flagsIntel,
3814 allocflags, MemoryPropertiesHelper::ObjType::UNKNOWN,
3815 *neoContext)) {
3816 err.set(CL_INVALID_VALUE);
3817 return nullptr;
3818 }
3819
3820 if (size > neoContext->getDevice(0u)->getSharedDeviceInfo().maxMemAllocSize && !unifiedMemoryProperties.allocationFlags.flags.allowUnrestrictedSize) {
3821 err.set(CL_INVALID_BUFFER_SIZE);
3822 return nullptr;
3823 }
3824
3825 return neoContext->getSVMAllocsManager()->createHostUnifiedMemoryAllocation(size, unifiedMemoryProperties);
3826 }
3827
clDeviceMemAllocINTEL(cl_context context,cl_device_id device,const cl_mem_properties_intel * properties,size_t size,cl_uint alignment,cl_int * errcodeRet)3828 void *clDeviceMemAllocINTEL(
3829 cl_context context,
3830 cl_device_id device,
3831 const cl_mem_properties_intel *properties,
3832 size_t size,
3833 cl_uint alignment,
3834 cl_int *errcodeRet) {
3835 Context *neoContext = nullptr;
3836 ClDevice *neoDevice = nullptr;
3837
3838 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
3839
3840 auto retVal = validateObjects(WithCastToInternal(context, &neoContext), WithCastToInternal(device, &neoDevice));
3841
3842 if (retVal != CL_SUCCESS) {
3843 err.set(retVal);
3844 return nullptr;
3845 }
3846
3847 auto subDeviceBitfields = neoContext->getDeviceBitfields();
3848 subDeviceBitfields[neoDevice->getRootDeviceIndex()] = neoDevice->getDeviceBitfield();
3849
3850 SVMAllocsManager::UnifiedMemoryProperties unifiedMemoryProperties(InternalMemoryType::DEVICE_UNIFIED_MEMORY,
3851 neoContext->getRootDeviceIndices(), subDeviceBitfields);
3852 cl_mem_flags flags = 0;
3853 cl_mem_flags_intel flagsIntel = 0;
3854 cl_mem_alloc_flags_intel allocflags = 0;
3855 if (!ClMemoryPropertiesHelper::parseMemoryProperties(properties, unifiedMemoryProperties.allocationFlags, flags, flagsIntel,
3856 allocflags, MemoryPropertiesHelper::ObjType::UNKNOWN,
3857 *neoContext)) {
3858 err.set(CL_INVALID_VALUE);
3859 return nullptr;
3860 }
3861
3862 if (size > neoDevice->getDevice().getDeviceInfo().maxMemAllocSize &&
3863 !unifiedMemoryProperties.allocationFlags.flags.allowUnrestrictedSize) {
3864 err.set(CL_INVALID_BUFFER_SIZE);
3865 return nullptr;
3866 }
3867
3868 unifiedMemoryProperties.device = &neoDevice->getDevice();
3869
3870 return neoContext->getSVMAllocsManager()->createUnifiedMemoryAllocation(size, unifiedMemoryProperties);
3871 }
3872
clSharedMemAllocINTEL(cl_context context,cl_device_id device,const cl_mem_properties_intel * properties,size_t size,cl_uint alignment,cl_int * errcodeRet)3873 void *clSharedMemAllocINTEL(
3874 cl_context context,
3875 cl_device_id device,
3876 const cl_mem_properties_intel *properties,
3877 size_t size,
3878 cl_uint alignment,
3879 cl_int *errcodeRet) {
3880 Context *neoContext = nullptr;
3881
3882 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
3883
3884 auto retVal = validateObjects(WithCastToInternal(context, &neoContext));
3885
3886 if (retVal != CL_SUCCESS) {
3887 err.set(retVal);
3888 return nullptr;
3889 }
3890
3891 cl_mem_flags flags = 0;
3892 cl_mem_flags_intel flagsIntel = 0;
3893 cl_mem_alloc_flags_intel allocflags = 0;
3894 ClDevice *neoDevice = castToObject<ClDevice>(device);
3895 Device *unifiedMemoryPropertiesDevice = nullptr;
3896 auto subDeviceBitfields = neoContext->getDeviceBitfields();
3897 if (neoDevice) {
3898 if (!neoContext->isDeviceAssociated(*neoDevice)) {
3899 err.set(CL_INVALID_DEVICE);
3900 return nullptr;
3901 }
3902 unifiedMemoryPropertiesDevice = &neoDevice->getDevice();
3903 subDeviceBitfields[neoDevice->getRootDeviceIndex()] = neoDevice->getDeviceBitfield();
3904 } else {
3905 neoDevice = neoContext->getDevice(0);
3906 }
3907 SVMAllocsManager::UnifiedMemoryProperties unifiedMemoryProperties(InternalMemoryType::SHARED_UNIFIED_MEMORY, neoContext->getRootDeviceIndices(), subDeviceBitfields);
3908 unifiedMemoryProperties.device = unifiedMemoryPropertiesDevice;
3909 if (!ClMemoryPropertiesHelper::parseMemoryProperties(properties, unifiedMemoryProperties.allocationFlags, flags, flagsIntel,
3910 allocflags, MemoryPropertiesHelper::ObjType::UNKNOWN,
3911 *neoContext)) {
3912 err.set(CL_INVALID_VALUE);
3913 return nullptr;
3914 }
3915
3916 if (size > neoDevice->getSharedDeviceInfo().maxMemAllocSize && !unifiedMemoryProperties.allocationFlags.flags.allowUnrestrictedSize) {
3917 err.set(CL_INVALID_BUFFER_SIZE);
3918 return nullptr;
3919 }
3920 auto ptr = neoContext->getSVMAllocsManager()->createSharedUnifiedMemoryAllocation(size, unifiedMemoryProperties, neoContext->getSpecialQueue(neoDevice->getRootDeviceIndex()));
3921 if (!ptr) {
3922 err.set(CL_OUT_OF_RESOURCES);
3923 }
3924
3925 return ptr;
3926 }
3927
clMemFreeCommon(cl_context context,const void * ptr,bool blocking)3928 cl_int clMemFreeCommon(cl_context context,
3929 const void *ptr,
3930 bool blocking) {
3931 Context *neoContext = nullptr;
3932 auto retVal = validateObjects(WithCastToInternal(context, &neoContext));
3933
3934 if (retVal != CL_SUCCESS) {
3935 return retVal;
3936 }
3937
3938 if (ptr && !neoContext->getSVMAllocsManager()->freeSVMAlloc(const_cast<void *>(ptr), blocking)) {
3939 return CL_INVALID_VALUE;
3940 }
3941
3942 if (neoContext->getSVMAllocsManager()->getSvmMapOperation(ptr)) {
3943 neoContext->getSVMAllocsManager()->removeSvmMapOperation(ptr);
3944 }
3945
3946 return CL_SUCCESS;
3947 }
3948
clMemFreeINTEL(cl_context context,void * ptr)3949 cl_int clMemFreeINTEL(
3950 cl_context context,
3951 void *ptr) {
3952 return clMemFreeCommon(context, ptr, false);
3953 }
3954
clMemBlockingFreeINTEL(cl_context context,void * ptr)3955 cl_int clMemBlockingFreeINTEL(
3956 cl_context context,
3957 void *ptr) {
3958 return clMemFreeCommon(context, ptr, true);
3959 }
3960
clGetMemAllocInfoINTEL(cl_context context,const void * ptr,cl_mem_info_intel paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)3961 cl_int clGetMemAllocInfoINTEL(
3962 cl_context context,
3963 const void *ptr,
3964 cl_mem_info_intel paramName,
3965 size_t paramValueSize,
3966 void *paramValue,
3967 size_t *paramValueSizeRet) {
3968 Context *pContext = nullptr;
3969 cl_int retVal = CL_SUCCESS;
3970 retVal = validateObject(WithCastToInternal(context, &pContext));
3971
3972 if (!pContext) {
3973 return retVal;
3974 }
3975
3976 auto allocationsManager = pContext->getSVMAllocsManager();
3977 if (!allocationsManager) {
3978 return CL_INVALID_VALUE;
3979 }
3980
3981 GetInfoHelper info(paramValue, paramValueSize, paramValueSizeRet);
3982 auto unifiedMemoryAllocation = allocationsManager->getSVMAlloc(ptr);
3983
3984 switch (paramName) {
3985 case CL_MEM_ALLOC_TYPE_INTEL: {
3986 if (!unifiedMemoryAllocation) {
3987 retVal = changeGetInfoStatusToCLResultType(info.set<cl_int>(CL_MEM_TYPE_UNKNOWN_INTEL));
3988 return retVal;
3989 } else if (unifiedMemoryAllocation->memoryType == InternalMemoryType::HOST_UNIFIED_MEMORY) {
3990 retVal = changeGetInfoStatusToCLResultType(info.set<cl_int>(CL_MEM_TYPE_HOST_INTEL));
3991 return retVal;
3992 } else if (unifiedMemoryAllocation->memoryType == InternalMemoryType::DEVICE_UNIFIED_MEMORY) {
3993 retVal = changeGetInfoStatusToCLResultType(info.set<cl_int>(CL_MEM_TYPE_DEVICE_INTEL));
3994 return retVal;
3995 } else {
3996 retVal = changeGetInfoStatusToCLResultType(info.set<cl_int>(CL_MEM_TYPE_SHARED_INTEL));
3997 return retVal;
3998 }
3999 break;
4000 }
4001 case CL_MEM_ALLOC_BASE_PTR_INTEL: {
4002 if (!unifiedMemoryAllocation) {
4003 return changeGetInfoStatusToCLResultType(info.set<void *>(nullptr));
4004 }
4005 return changeGetInfoStatusToCLResultType(info.set<uint64_t>(unifiedMemoryAllocation->gpuAllocations.getDefaultGraphicsAllocation()->getGpuAddress()));
4006 }
4007 case CL_MEM_ALLOC_SIZE_INTEL: {
4008 if (!unifiedMemoryAllocation) {
4009 return changeGetInfoStatusToCLResultType(info.set<size_t>(0u));
4010 }
4011 return changeGetInfoStatusToCLResultType(info.set<size_t>(unifiedMemoryAllocation->size));
4012 }
4013 case CL_MEM_ALLOC_FLAGS_INTEL: {
4014 if (!unifiedMemoryAllocation) {
4015 return changeGetInfoStatusToCLResultType(info.set<cl_mem_alloc_flags_intel>(0u));
4016 }
4017 return changeGetInfoStatusToCLResultType(info.set<cl_mem_alloc_flags_intel>(unifiedMemoryAllocation->allocationFlagsProperty.allAllocFlags));
4018 }
4019 case CL_MEM_ALLOC_DEVICE_INTEL: {
4020 if (!unifiedMemoryAllocation) {
4021 return changeGetInfoStatusToCLResultType(info.set<cl_device_id>(static_cast<cl_device_id>(nullptr)));
4022 }
4023 auto device = unifiedMemoryAllocation->device ? unifiedMemoryAllocation->device->getSpecializedDevice<ClDevice>() : nullptr;
4024 return changeGetInfoStatusToCLResultType(info.set<cl_device_id>(device));
4025 }
4026
4027 default: {
4028 }
4029 }
4030
4031 return CL_INVALID_VALUE;
4032 }
4033
clSetKernelArgMemPointerINTEL(cl_kernel kernel,cl_uint argIndex,const void * argValue)4034 cl_int clSetKernelArgMemPointerINTEL(
4035 cl_kernel kernel,
4036 cl_uint argIndex,
4037 const void *argValue) {
4038 return clSetKernelArgSVMPointer(kernel, argIndex, argValue);
4039 }
4040
clEnqueueMemsetINTEL(cl_command_queue commandQueue,void * dstPtr,cl_int value,size_t size,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4041 cl_int clEnqueueMemsetINTEL(
4042 cl_command_queue commandQueue,
4043 void *dstPtr,
4044 cl_int value,
4045 size_t size,
4046 cl_uint numEventsInWaitList,
4047 const cl_event *eventWaitList,
4048 cl_event *event) {
4049 auto retVal = clEnqueueSVMMemFill(commandQueue,
4050 dstPtr,
4051 &value,
4052 1u,
4053 size,
4054 numEventsInWaitList,
4055 eventWaitList,
4056 event);
4057 if (retVal == CL_SUCCESS && event) {
4058 auto pEvent = castToObjectOrAbort<Event>(*event);
4059 pEvent->setCmdType(CL_COMMAND_MEMSET_INTEL);
4060 }
4061
4062 return retVal;
4063 }
4064
clEnqueueMemFillINTEL(cl_command_queue commandQueue,void * dstPtr,const void * pattern,size_t patternSize,size_t size,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4065 cl_int clEnqueueMemFillINTEL(
4066 cl_command_queue commandQueue,
4067 void *dstPtr,
4068 const void *pattern,
4069 size_t patternSize,
4070 size_t size,
4071 cl_uint numEventsInWaitList,
4072 const cl_event *eventWaitList,
4073 cl_event *event) {
4074
4075 auto retVal = clEnqueueSVMMemFill(commandQueue,
4076 dstPtr,
4077 pattern,
4078 patternSize,
4079 size,
4080 numEventsInWaitList,
4081 eventWaitList,
4082 event);
4083 if (retVal == CL_SUCCESS && event) {
4084 auto pEvent = castToObjectOrAbort<Event>(*event);
4085 pEvent->setCmdType(CL_COMMAND_MEMFILL_INTEL);
4086 }
4087
4088 return retVal;
4089 }
4090
clEnqueueMemcpyINTEL(cl_command_queue commandQueue,cl_bool blocking,void * dstPtr,const void * srcPtr,size_t size,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4091 cl_int clEnqueueMemcpyINTEL(
4092 cl_command_queue commandQueue,
4093 cl_bool blocking,
4094 void *dstPtr,
4095 const void *srcPtr,
4096 size_t size,
4097 cl_uint numEventsInWaitList,
4098 const cl_event *eventWaitList,
4099 cl_event *event) {
4100 auto retVal = clEnqueueSVMMemcpy(commandQueue,
4101 blocking,
4102 dstPtr,
4103 srcPtr,
4104 size,
4105 numEventsInWaitList,
4106 eventWaitList,
4107 event);
4108 if (retVal == CL_SUCCESS && event) {
4109 auto pEvent = castToObjectOrAbort<Event>(*event);
4110 pEvent->setCmdType(CL_COMMAND_MEMCPY_INTEL);
4111 }
4112
4113 return retVal;
4114 }
4115
clEnqueueMigrateMemINTEL(cl_command_queue commandQueue,const void * ptr,size_t size,cl_mem_migration_flags flags,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4116 cl_int clEnqueueMigrateMemINTEL(
4117 cl_command_queue commandQueue,
4118 const void *ptr,
4119 size_t size,
4120 cl_mem_migration_flags flags,
4121 cl_uint numEventsInWaitList,
4122 const cl_event *eventWaitList,
4123 cl_event *event) {
4124 cl_int retVal = CL_SUCCESS;
4125
4126 CommandQueue *pCommandQueue = nullptr;
4127 retVal = validateObjects(WithCastToInternal(commandQueue, &pCommandQueue), ptr, EventWaitList(numEventsInWaitList, eventWaitList));
4128
4129 if (retVal == CL_SUCCESS) {
4130 pCommandQueue->enqueueMarkerWithWaitList(numEventsInWaitList, eventWaitList, event);
4131
4132 if (event) {
4133 auto pEvent = castToObjectOrAbort<Event>(*event);
4134 pEvent->setCmdType(CL_COMMAND_MIGRATEMEM_INTEL);
4135 }
4136 }
4137
4138 return retVal;
4139 }
4140
clEnqueueMemAdviseINTEL(cl_command_queue commandQueue,const void * ptr,size_t size,cl_mem_advice_intel advice,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4141 cl_int clEnqueueMemAdviseINTEL(
4142 cl_command_queue commandQueue,
4143 const void *ptr,
4144 size_t size,
4145 cl_mem_advice_intel advice,
4146 cl_uint numEventsInWaitList,
4147 const cl_event *eventWaitList,
4148 cl_event *event) {
4149 cl_int retVal = CL_SUCCESS;
4150
4151 CommandQueue *pCommandQueue = nullptr;
4152 retVal = validateObjects(WithCastToInternal(commandQueue, &pCommandQueue), ptr, EventWaitList(numEventsInWaitList, eventWaitList));
4153
4154 if (retVal == CL_SUCCESS) {
4155 pCommandQueue->enqueueMarkerWithWaitList(numEventsInWaitList, eventWaitList, event);
4156
4157 if (event) {
4158 auto pEvent = castToObjectOrAbort<Event>(*event);
4159 pEvent->setCmdType(CL_COMMAND_MEMADVISE_INTEL);
4160 }
4161 }
4162
4163 return retVal;
4164 }
4165
clCreateCommandQueueWithPropertiesKHR(cl_context context,cl_device_id device,const cl_queue_properties_khr * properties,cl_int * errcodeRet)4166 cl_command_queue CL_API_CALL clCreateCommandQueueWithPropertiesKHR(cl_context context,
4167 cl_device_id device,
4168 const cl_queue_properties_khr *properties,
4169 cl_int *errcodeRet) {
4170
4171 API_ENTER(errcodeRet);
4172 DBG_LOG_INPUTS("context", context,
4173 "device", device,
4174 "properties", properties);
4175
4176 return clCreateCommandQueueWithProperties(context, device, properties, errcodeRet);
4177 }
4178
clCreateAcceleratorINTEL(cl_context context,cl_accelerator_type_intel acceleratorType,size_t descriptorSize,const void * descriptor,cl_int * errcodeRet)4179 cl_accelerator_intel CL_API_CALL clCreateAcceleratorINTEL(
4180 cl_context context,
4181 cl_accelerator_type_intel acceleratorType,
4182 size_t descriptorSize,
4183 const void *descriptor,
4184 cl_int *errcodeRet) {
4185 cl_int retVal = CL_SUCCESS;
4186 API_ENTER(&retVal);
4187 DBG_LOG_INPUTS("context", context,
4188 "acceleratorType", acceleratorType,
4189 "descriptorSize", descriptorSize,
4190 "descriptor", NEO::FileLoggerInstance().infoPointerToString(descriptor, descriptorSize));
4191 cl_accelerator_intel accelerator = nullptr;
4192
4193 do {
4194 retVal = validateObjects(context);
4195
4196 if (retVal != CL_SUCCESS) {
4197 retVal = CL_INVALID_CONTEXT;
4198 break;
4199 }
4200
4201 Context *pContext = castToObject<Context>(context);
4202
4203 DEBUG_BREAK_IF(!pContext);
4204
4205 switch (acceleratorType) {
4206 case CL_ACCELERATOR_TYPE_MOTION_ESTIMATION_INTEL:
4207 accelerator = VmeAccelerator::create(
4208 pContext,
4209 acceleratorType,
4210 descriptorSize,
4211 descriptor,
4212 retVal);
4213 break;
4214 default:
4215 retVal = CL_INVALID_ACCELERATOR_TYPE_INTEL;
4216 }
4217
4218 } while (false);
4219
4220 if (errcodeRet) {
4221 *errcodeRet = retVal;
4222 }
4223
4224 return accelerator;
4225 }
4226
clRetainAcceleratorINTEL(cl_accelerator_intel accelerator)4227 cl_int CL_API_CALL clRetainAcceleratorINTEL(
4228 cl_accelerator_intel accelerator) {
4229 cl_int retVal = CL_SUCCESS;
4230 API_ENTER(&retVal);
4231 DBG_LOG_INPUTS("accelerator", accelerator);
4232
4233 IntelAccelerator *pAccelerator = nullptr;
4234
4235 do {
4236 pAccelerator = castToObject<IntelAccelerator>(accelerator);
4237
4238 if (!pAccelerator) {
4239 retVal = CL_INVALID_ACCELERATOR_INTEL;
4240 break;
4241 }
4242
4243 pAccelerator->retain();
4244 } while (false);
4245
4246 return retVal;
4247 }
4248
clGetAcceleratorInfoINTEL(cl_accelerator_intel accelerator,cl_accelerator_info_intel paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)4249 cl_int CL_API_CALL clGetAcceleratorInfoINTEL(
4250 cl_accelerator_intel accelerator,
4251 cl_accelerator_info_intel paramName,
4252 size_t paramValueSize,
4253 void *paramValue,
4254 size_t *paramValueSizeRet) {
4255 cl_int retVal = CL_SUCCESS;
4256 API_ENTER(&retVal);
4257 DBG_LOG_INPUTS("accelerator", accelerator,
4258 "paramName", paramName,
4259 "paramValueSize", paramValueSize,
4260 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
4261 "paramValueSizeRet", paramValueSizeRet);
4262 IntelAccelerator *pAccelerator = nullptr;
4263
4264 do {
4265 pAccelerator = castToObject<IntelAccelerator>(accelerator);
4266
4267 if (!pAccelerator) {
4268 retVal = CL_INVALID_ACCELERATOR_INTEL;
4269 break;
4270 }
4271
4272 retVal = pAccelerator->getInfo(
4273 paramName, paramValueSize, paramValue, paramValueSizeRet);
4274
4275 } while (false);
4276
4277 return retVal;
4278 }
4279
clReleaseAcceleratorINTEL(cl_accelerator_intel accelerator)4280 cl_int CL_API_CALL clReleaseAcceleratorINTEL(
4281 cl_accelerator_intel accelerator) {
4282 cl_int retVal = CL_SUCCESS;
4283 API_ENTER(&retVal);
4284 DBG_LOG_INPUTS("accelerator", accelerator);
4285
4286 IntelAccelerator *pAccelerator = nullptr;
4287
4288 do {
4289 pAccelerator = castToObject<IntelAccelerator>(accelerator);
4290
4291 if (!pAccelerator) {
4292 retVal = CL_INVALID_ACCELERATOR_INTEL;
4293 break;
4294 }
4295
4296 pAccelerator->release();
4297 } while (false);
4298
4299 return retVal;
4300 }
4301
clCreateProgramWithILKHR(cl_context context,const void * il,size_t length,cl_int * errcodeRet)4302 cl_program CL_API_CALL clCreateProgramWithILKHR(cl_context context,
4303 const void *il,
4304 size_t length,
4305 cl_int *errcodeRet) {
4306 cl_int retVal = CL_SUCCESS;
4307 API_ENTER(&retVal);
4308 DBG_LOG_INPUTS("context", context,
4309 "il", NEO::FileLoggerInstance().infoPointerToString(il, length),
4310 "length", length);
4311
4312 cl_program program = nullptr;
4313 Context *pContext = nullptr;
4314 retVal = validateObjects(WithCastToInternal(context, &pContext), il);
4315 if (retVal == CL_SUCCESS) {
4316 program = ProgramFunctions::createFromIL(
4317 pContext,
4318 il,
4319 length,
4320 retVal);
4321 }
4322
4323 if (errcodeRet != nullptr) {
4324 *errcodeRet = retVal;
4325 }
4326
4327 return program;
4328 }
4329
clGetKernelSuggestedLocalWorkSizeKHR(cl_command_queue commandQueue,cl_kernel kernel,cl_uint workDim,const size_t * globalWorkOffset,const size_t * globalWorkSize,size_t * suggestedLocalWorkSize)4330 cl_int CL_API_CALL clGetKernelSuggestedLocalWorkSizeKHR(cl_command_queue commandQueue,
4331 cl_kernel kernel,
4332 cl_uint workDim,
4333 const size_t *globalWorkOffset,
4334 const size_t *globalWorkSize,
4335 size_t *suggestedLocalWorkSize) {
4336 return clGetKernelSuggestedLocalWorkSizeINTEL(commandQueue,
4337 kernel,
4338 workDim,
4339 globalWorkOffset,
4340 globalWorkSize,
4341 suggestedLocalWorkSize);
4342 }
4343
4344 #define RETURN_FUNC_PTR_IF_EXIST(name) \
4345 { \
4346 if (!strcmp(funcName, #name)) { \
4347 void *ret = ((void *)(name)); \
4348 TRACING_EXIT(clGetExtensionFunctionAddress, (void **)&ret); \
4349 return ret; \
4350 } \
4351 }
clGetExtensionFunctionAddress(const char * funcName)4352 void *CL_API_CALL clGetExtensionFunctionAddress(const char *funcName) {
4353 TRACING_ENTER(clGetExtensionFunctionAddress, &funcName);
4354
4355 DBG_LOG_INPUTS("funcName", funcName);
4356 // Support an internal call by the ICD
4357 RETURN_FUNC_PTR_IF_EXIST(clIcdGetPlatformIDsKHR);
4358
4359 //perf counters
4360 RETURN_FUNC_PTR_IF_EXIST(clCreatePerfCountersCommandQueueINTEL);
4361 RETURN_FUNC_PTR_IF_EXIST(clSetPerformanceConfigurationINTEL);
4362 // Support device extensions
4363 RETURN_FUNC_PTR_IF_EXIST(clCreateAcceleratorINTEL);
4364 RETURN_FUNC_PTR_IF_EXIST(clGetAcceleratorInfoINTEL);
4365 RETURN_FUNC_PTR_IF_EXIST(clRetainAcceleratorINTEL);
4366 RETURN_FUNC_PTR_IF_EXIST(clReleaseAcceleratorINTEL);
4367 RETURN_FUNC_PTR_IF_EXIST(clCreateBufferWithPropertiesINTEL);
4368 RETURN_FUNC_PTR_IF_EXIST(clCreateImageWithPropertiesINTEL);
4369 RETURN_FUNC_PTR_IF_EXIST(clAddCommentINTEL);
4370 RETURN_FUNC_PTR_IF_EXIST(clEnqueueVerifyMemoryINTEL);
4371
4372 RETURN_FUNC_PTR_IF_EXIST(clCreateTracingHandleINTEL);
4373 RETURN_FUNC_PTR_IF_EXIST(clSetTracingPointINTEL);
4374 RETURN_FUNC_PTR_IF_EXIST(clDestroyTracingHandleINTEL);
4375 RETURN_FUNC_PTR_IF_EXIST(clEnableTracingINTEL);
4376 RETURN_FUNC_PTR_IF_EXIST(clDisableTracingINTEL);
4377 RETURN_FUNC_PTR_IF_EXIST(clGetTracingStateINTEL);
4378
4379 RETURN_FUNC_PTR_IF_EXIST(clHostMemAllocINTEL);
4380 RETURN_FUNC_PTR_IF_EXIST(clDeviceMemAllocINTEL);
4381 RETURN_FUNC_PTR_IF_EXIST(clSharedMemAllocINTEL);
4382 RETURN_FUNC_PTR_IF_EXIST(clMemFreeINTEL);
4383 RETURN_FUNC_PTR_IF_EXIST(clMemBlockingFreeINTEL);
4384 RETURN_FUNC_PTR_IF_EXIST(clGetMemAllocInfoINTEL);
4385 RETURN_FUNC_PTR_IF_EXIST(clSetKernelArgMemPointerINTEL);
4386 RETURN_FUNC_PTR_IF_EXIST(clEnqueueMemsetINTEL);
4387 RETURN_FUNC_PTR_IF_EXIST(clEnqueueMemFillINTEL);
4388 RETURN_FUNC_PTR_IF_EXIST(clEnqueueMemcpyINTEL);
4389 RETURN_FUNC_PTR_IF_EXIST(clEnqueueMigrateMemINTEL);
4390 RETURN_FUNC_PTR_IF_EXIST(clEnqueueMemAdviseINTEL);
4391 RETURN_FUNC_PTR_IF_EXIST(clGetDeviceFunctionPointerINTEL);
4392 RETURN_FUNC_PTR_IF_EXIST(clGetDeviceGlobalVariablePointerINTEL);
4393 RETURN_FUNC_PTR_IF_EXIST(clGetKernelMaxConcurrentWorkGroupCountINTEL);
4394 RETURN_FUNC_PTR_IF_EXIST(clGetKernelSuggestedLocalWorkSizeINTEL);
4395 RETURN_FUNC_PTR_IF_EXIST(clEnqueueNDCountKernelINTEL);
4396
4397 void *ret = sharingFactory.getExtensionFunctionAddress(funcName);
4398 if (ret != nullptr) {
4399 TRACING_EXIT(clGetExtensionFunctionAddress, &ret);
4400 return ret;
4401 }
4402
4403 // SPIR-V support through the cl_khr_il_program extension
4404 RETURN_FUNC_PTR_IF_EXIST(clCreateProgramWithILKHR);
4405 RETURN_FUNC_PTR_IF_EXIST(clCreateCommandQueueWithPropertiesKHR);
4406
4407 RETURN_FUNC_PTR_IF_EXIST(clSetProgramSpecializationConstant);
4408
4409 RETURN_FUNC_PTR_IF_EXIST(clGetKernelSuggestedLocalWorkSizeKHR);
4410
4411 ret = getAdditionalExtensionFunctionAddress(funcName);
4412 TRACING_EXIT(clGetExtensionFunctionAddress, &ret);
4413 return ret;
4414 }
4415
4416 // OpenCL 1.2
clGetExtensionFunctionAddressForPlatform(cl_platform_id platform,const char * funcName)4417 void *CL_API_CALL clGetExtensionFunctionAddressForPlatform(cl_platform_id platform,
4418 const char *funcName) {
4419 TRACING_ENTER(clGetExtensionFunctionAddressForPlatform, &platform, &funcName);
4420 DBG_LOG_INPUTS("platform", platform, "funcName", funcName);
4421 auto pPlatform = castToObject<Platform>(platform);
4422
4423 if (pPlatform == nullptr) {
4424 void *ret = nullptr;
4425 TRACING_EXIT(clGetExtensionFunctionAddressForPlatform, &ret);
4426 return ret;
4427 }
4428
4429 void *ret = clGetExtensionFunctionAddress(funcName);
4430 TRACING_EXIT(clGetExtensionFunctionAddressForPlatform, &ret);
4431 return ret;
4432 }
4433
clSVMAlloc(cl_context context,cl_svm_mem_flags flags,size_t size,cl_uint alignment)4434 void *CL_API_CALL clSVMAlloc(cl_context context,
4435 cl_svm_mem_flags flags,
4436 size_t size,
4437 cl_uint alignment) {
4438 TRACING_ENTER(clSVMAlloc, &context, &flags, &size, &alignment);
4439 DBG_LOG_INPUTS("context", context,
4440 "flags", flags,
4441 "size", size,
4442 "alignment", alignment);
4443 void *pAlloc = nullptr;
4444 Context *pContext = nullptr;
4445
4446 if (validateObjects(WithCastToInternal(context, &pContext)) != CL_SUCCESS) {
4447 TRACING_EXIT(clSVMAlloc, &pAlloc);
4448 return pAlloc;
4449 }
4450
4451 {
4452 // allow CL_MEM_ALLOW_UNRESTRICTED_SIZE_INTEL with every combination
4453 cl_svm_mem_flags tempFlags = flags & (~CL_MEM_ALLOW_UNRESTRICTED_SIZE_INTEL);
4454
4455 if (tempFlags == 0) {
4456 tempFlags = CL_MEM_READ_WRITE;
4457 }
4458
4459 if (!((tempFlags == CL_MEM_READ_WRITE) ||
4460 (tempFlags == CL_MEM_WRITE_ONLY) ||
4461 (tempFlags == CL_MEM_READ_ONLY) ||
4462 (tempFlags == CL_MEM_SVM_FINE_GRAIN_BUFFER) ||
4463 (tempFlags == (CL_MEM_SVM_FINE_GRAIN_BUFFER | CL_MEM_SVM_ATOMICS)) ||
4464 (tempFlags == (CL_MEM_READ_WRITE | CL_MEM_SVM_FINE_GRAIN_BUFFER)) ||
4465 (tempFlags == (CL_MEM_READ_WRITE | CL_MEM_SVM_FINE_GRAIN_BUFFER | CL_MEM_SVM_ATOMICS)) ||
4466 (tempFlags == (CL_MEM_WRITE_ONLY | CL_MEM_SVM_FINE_GRAIN_BUFFER)) ||
4467 (tempFlags == (CL_MEM_WRITE_ONLY | CL_MEM_SVM_FINE_GRAIN_BUFFER | CL_MEM_SVM_ATOMICS)) ||
4468 (tempFlags == (CL_MEM_READ_ONLY | CL_MEM_SVM_FINE_GRAIN_BUFFER)) ||
4469 (tempFlags == (CL_MEM_READ_ONLY | CL_MEM_SVM_FINE_GRAIN_BUFFER | CL_MEM_SVM_ATOMICS)))) {
4470
4471 TRACING_EXIT(clSVMAlloc, &pAlloc);
4472 return pAlloc;
4473 }
4474 }
4475
4476 auto pDevice = pContext->getDevice(0);
4477 bool allowUnrestrictedSize = (flags & CL_MEM_ALLOW_UNRESTRICTED_SIZE_INTEL) || DebugManager.flags.AllowUnrestrictedSize.get();
4478
4479 if ((size == 0) ||
4480 (!allowUnrestrictedSize && (size > pDevice->getSharedDeviceInfo().maxMemAllocSize))) {
4481 TRACING_EXIT(clSVMAlloc, &pAlloc);
4482 return pAlloc;
4483 }
4484
4485 if ((alignment && (alignment & (alignment - 1))) || (alignment > sizeof(cl_ulong16))) {
4486 TRACING_EXIT(clSVMAlloc, &pAlloc);
4487 return pAlloc;
4488 }
4489
4490 const HardwareInfo &hwInfo = pDevice->getHardwareInfo();
4491 if (!hwInfo.capabilityTable.ftrSvm) {
4492 TRACING_EXIT(clSVMAlloc, &pAlloc);
4493 return pAlloc;
4494 }
4495
4496 if (flags & CL_MEM_SVM_FINE_GRAIN_BUFFER) {
4497 bool supportsFineGrained = hwInfo.capabilityTable.ftrSupportsCoherency;
4498 if (DebugManager.flags.ForceFineGrainedSVMSupport.get() != -1) {
4499 supportsFineGrained = !!DebugManager.flags.ForceFineGrainedSVMSupport.get();
4500 }
4501 if (!supportsFineGrained) {
4502 TRACING_EXIT(clSVMAlloc, &pAlloc);
4503 return pAlloc;
4504 }
4505 }
4506
4507 pAlloc = pContext->getSVMAllocsManager()->createSVMAlloc(size, MemObjHelper::getSvmAllocationProperties(flags), pContext->getRootDeviceIndices(), pContext->getDeviceBitfields());
4508
4509 if (pContext->isProvidingPerformanceHints()) {
4510 pContext->providePerformanceHint(CL_CONTEXT_DIAGNOSTICS_LEVEL_GOOD_INTEL, CL_SVM_ALLOC_MEETS_ALIGNMENT_RESTRICTIONS, pAlloc, size);
4511 }
4512 TRACING_EXIT(clSVMAlloc, &pAlloc);
4513 return pAlloc;
4514 }
4515
clSVMFree(cl_context context,void * svmPointer)4516 void CL_API_CALL clSVMFree(cl_context context,
4517 void *svmPointer) {
4518 TRACING_ENTER(clSVMFree, &context, &svmPointer);
4519 DBG_LOG_INPUTS("context", context,
4520 "svmPointer", svmPointer);
4521
4522 Context *pContext = nullptr;
4523 cl_int retVal = validateObjects(
4524 WithCastToInternal(context, &pContext));
4525
4526 if (retVal != CL_SUCCESS) {
4527 TRACING_EXIT(clSVMFree, nullptr);
4528 return;
4529 }
4530
4531 auto pClDevice = pContext->getDevice(0);
4532 if (!pClDevice->getHardwareInfo().capabilityTable.ftrSvm) {
4533 TRACING_EXIT(clSVMFree, nullptr);
4534 return;
4535 }
4536
4537 pContext->getSVMAllocsManager()->freeSVMAlloc(svmPointer);
4538 TRACING_EXIT(clSVMFree, nullptr);
4539 }
4540
clEnqueueSVMFree(cl_command_queue commandQueue,cl_uint numSvmPointers,void * svmPointers[],void (CL_CALLBACK * pfnFreeFunc)(cl_command_queue queue,cl_uint numSvmPointers,void * svmPointers[],void * userData),void * userData,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4541 cl_int CL_API_CALL clEnqueueSVMFree(cl_command_queue commandQueue,
4542 cl_uint numSvmPointers,
4543 void *svmPointers[],
4544 void(CL_CALLBACK *pfnFreeFunc)(cl_command_queue queue,
4545 cl_uint numSvmPointers,
4546 void *svmPointers[],
4547 void *userData),
4548 void *userData,
4549 cl_uint numEventsInWaitList,
4550 const cl_event *eventWaitList,
4551 cl_event *event) {
4552 TRACING_ENTER(clEnqueueSVMFree, &commandQueue, &numSvmPointers, &svmPointers, &pfnFreeFunc, &userData, &numEventsInWaitList, &eventWaitList, &event);
4553
4554 CommandQueue *pCommandQueue = nullptr;
4555
4556 cl_int retVal = validateObjects(
4557 WithCastToInternal(commandQueue, &pCommandQueue),
4558 EventWaitList(numEventsInWaitList, eventWaitList));
4559
4560 API_ENTER(&retVal);
4561
4562 DBG_LOG_INPUTS("commandQueue", commandQueue,
4563 "numSvmPointers", numSvmPointers,
4564 "svmPointers", svmPointers,
4565 "pfnFreeFunc", pfnFreeFunc,
4566 "userData", userData,
4567 "numEventsInWaitList", numEventsInWaitList,
4568 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
4569 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
4570
4571 if (retVal != CL_SUCCESS) {
4572 TRACING_EXIT(clEnqueueSVMFree, &retVal);
4573 return retVal;
4574 }
4575
4576 auto &device = pCommandQueue->getDevice();
4577 if (!device.getHardwareInfo().capabilityTable.ftrSvm) {
4578 retVal = CL_INVALID_OPERATION;
4579 TRACING_EXIT(clEnqueueSVMFree, &retVal);
4580 return retVal;
4581 }
4582
4583 if (((svmPointers != nullptr) && (numSvmPointers == 0)) ||
4584 ((svmPointers == nullptr) && (numSvmPointers != 0))) {
4585 retVal = CL_INVALID_VALUE;
4586 TRACING_EXIT(clEnqueueSVMFree, &retVal);
4587 return retVal;
4588 }
4589
4590 retVal = pCommandQueue->enqueueSVMFree(
4591 numSvmPointers,
4592 svmPointers,
4593 pfnFreeFunc,
4594 userData,
4595 numEventsInWaitList,
4596 eventWaitList,
4597 event);
4598
4599 TRACING_EXIT(clEnqueueSVMFree, &retVal);
4600 return retVal;
4601 }
4602
clEnqueueSVMMemcpy(cl_command_queue commandQueue,cl_bool blockingCopy,void * dstPtr,const void * srcPtr,size_t size,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4603 cl_int CL_API_CALL clEnqueueSVMMemcpy(cl_command_queue commandQueue,
4604 cl_bool blockingCopy,
4605 void *dstPtr,
4606 const void *srcPtr,
4607 size_t size,
4608 cl_uint numEventsInWaitList,
4609 const cl_event *eventWaitList,
4610 cl_event *event) {
4611 TRACING_ENTER(clEnqueueSVMMemcpy, &commandQueue, &blockingCopy, &dstPtr, &srcPtr, &size, &numEventsInWaitList, &eventWaitList, &event);
4612
4613 CommandQueue *pCommandQueue = nullptr;
4614
4615 cl_int retVal = validateObjects(
4616 WithCastToInternal(commandQueue, &pCommandQueue),
4617 EventWaitList(numEventsInWaitList, eventWaitList));
4618
4619 API_ENTER(&retVal);
4620
4621 DBG_LOG_INPUTS("commandQueue", commandQueue,
4622 "blockingCopy", blockingCopy,
4623 "dstPtr", dstPtr,
4624 "srcPtr", srcPtr,
4625 "size", size,
4626 "numEventsInWaitList", numEventsInWaitList,
4627 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
4628 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
4629
4630 if (retVal != CL_SUCCESS) {
4631 TRACING_EXIT(clEnqueueSVMMemcpy, &retVal);
4632 return retVal;
4633 }
4634
4635 auto &device = pCommandQueue->getDevice();
4636 if (!device.getHardwareInfo().capabilityTable.ftrSvm) {
4637 retVal = CL_INVALID_OPERATION;
4638 TRACING_EXIT(clEnqueueSVMMemcpy, &retVal);
4639 return retVal;
4640 }
4641
4642 if ((dstPtr == nullptr) || (srcPtr == nullptr)) {
4643 retVal = CL_INVALID_VALUE;
4644 TRACING_EXIT(clEnqueueSVMMemcpy, &retVal);
4645 return retVal;
4646 }
4647
4648 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_TRANSFER_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
4649 retVal = CL_INVALID_OPERATION;
4650 TRACING_EXIT(clEnqueueSVMMemcpy, &retVal);
4651 return retVal;
4652 }
4653
4654 retVal = pCommandQueue->enqueueSVMMemcpy(
4655 blockingCopy,
4656 dstPtr,
4657 srcPtr,
4658 size,
4659 numEventsInWaitList,
4660 eventWaitList,
4661 event);
4662
4663 TRACING_EXIT(clEnqueueSVMMemcpy, &retVal);
4664 return retVal;
4665 }
4666
clEnqueueSVMMemFill(cl_command_queue commandQueue,void * svmPtr,const void * pattern,size_t patternSize,size_t size,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4667 cl_int CL_API_CALL clEnqueueSVMMemFill(cl_command_queue commandQueue,
4668 void *svmPtr,
4669 const void *pattern,
4670 size_t patternSize,
4671 size_t size,
4672 cl_uint numEventsInWaitList,
4673 const cl_event *eventWaitList,
4674 cl_event *event) {
4675 TRACING_ENTER(clEnqueueSVMMemFill, &commandQueue, &svmPtr, &pattern, &patternSize, &size, &numEventsInWaitList, &eventWaitList, &event);
4676
4677 CommandQueue *pCommandQueue = nullptr;
4678
4679 cl_int retVal = validateObjects(
4680 WithCastToInternal(commandQueue, &pCommandQueue),
4681 EventWaitList(numEventsInWaitList, eventWaitList));
4682
4683 API_ENTER(&retVal);
4684
4685 DBG_LOG_INPUTS("commandQueue", commandQueue,
4686 "svmPtr", NEO::FileLoggerInstance().infoPointerToString(svmPtr, size),
4687 "pattern", NEO::FileLoggerInstance().infoPointerToString(pattern, patternSize),
4688 "patternSize", patternSize,
4689 "size", size,
4690 "numEventsInWaitList", numEventsInWaitList,
4691 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
4692 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
4693
4694 if (retVal != CL_SUCCESS) {
4695 TRACING_EXIT(clEnqueueSVMMemFill, &retVal);
4696 return retVal;
4697 }
4698
4699 auto &device = pCommandQueue->getDevice();
4700 if (!device.getHardwareInfo().capabilityTable.ftrSvm) {
4701 retVal = CL_INVALID_OPERATION;
4702 TRACING_EXIT(clEnqueueSVMMemFill, &retVal);
4703 return retVal;
4704 }
4705
4706 if ((svmPtr == nullptr) || (size == 0)) {
4707 retVal = CL_INVALID_VALUE;
4708 TRACING_EXIT(clEnqueueSVMMemFill, &retVal);
4709 return retVal;
4710 }
4711
4712 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_FILL_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
4713 retVal = CL_INVALID_OPERATION;
4714 TRACING_EXIT(clEnqueueSVMMemFill, &retVal);
4715 return retVal;
4716 }
4717
4718 retVal = pCommandQueue->enqueueSVMMemFill(
4719 svmPtr,
4720 pattern,
4721 patternSize,
4722 size,
4723 numEventsInWaitList,
4724 eventWaitList,
4725 event);
4726
4727 TRACING_EXIT(clEnqueueSVMMemFill, &retVal);
4728 return retVal;
4729 }
4730
clEnqueueSVMMap(cl_command_queue commandQueue,cl_bool blockingMap,cl_map_flags mapFlags,void * svmPtr,size_t size,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4731 cl_int CL_API_CALL clEnqueueSVMMap(cl_command_queue commandQueue,
4732 cl_bool blockingMap,
4733 cl_map_flags mapFlags,
4734 void *svmPtr,
4735 size_t size,
4736 cl_uint numEventsInWaitList,
4737 const cl_event *eventWaitList,
4738 cl_event *event) {
4739 TRACING_ENTER(clEnqueueSVMMap, &commandQueue, &blockingMap, &mapFlags, &svmPtr, &size, &numEventsInWaitList, &eventWaitList, &event);
4740
4741 CommandQueue *pCommandQueue = nullptr;
4742
4743 cl_int retVal = validateObjects(
4744 WithCastToInternal(commandQueue, &pCommandQueue),
4745 EventWaitList(numEventsInWaitList, eventWaitList));
4746 API_ENTER(&retVal);
4747 DBG_LOG_INPUTS("commandQueue", commandQueue,
4748 "blockingMap", blockingMap,
4749 "mapFlags", mapFlags,
4750 "svmPtr", NEO::FileLoggerInstance().infoPointerToString(svmPtr, size),
4751 "size", size,
4752 "numEventsInWaitList", numEventsInWaitList,
4753 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
4754 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
4755
4756 if (CL_SUCCESS != retVal) {
4757 TRACING_EXIT(clEnqueueSVMMap, &retVal);
4758 return retVal;
4759 }
4760
4761 auto &device = pCommandQueue->getDevice();
4762 if (!device.getHardwareInfo().capabilityTable.ftrSvm) {
4763 retVal = CL_INVALID_OPERATION;
4764 TRACING_EXIT(clEnqueueSVMMap, &retVal);
4765 return retVal;
4766 }
4767
4768 if ((svmPtr == nullptr) || (size == 0)) {
4769 retVal = CL_INVALID_VALUE;
4770 TRACING_EXIT(clEnqueueSVMMap, &retVal);
4771 return retVal;
4772 }
4773
4774 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_MAP_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
4775 retVal = CL_INVALID_OPERATION;
4776 TRACING_EXIT(clEnqueueSVMMap, &retVal);
4777 return retVal;
4778 }
4779
4780 retVal = pCommandQueue->enqueueSVMMap(
4781 blockingMap,
4782 mapFlags,
4783 svmPtr,
4784 size,
4785 numEventsInWaitList,
4786 eventWaitList,
4787 event,
4788 true);
4789
4790 TRACING_EXIT(clEnqueueSVMMap, &retVal);
4791 return retVal;
4792 }
4793
clEnqueueSVMUnmap(cl_command_queue commandQueue,void * svmPtr,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)4794 cl_int CL_API_CALL clEnqueueSVMUnmap(cl_command_queue commandQueue,
4795 void *svmPtr,
4796 cl_uint numEventsInWaitList,
4797 const cl_event *eventWaitList,
4798 cl_event *event) {
4799 TRACING_ENTER(clEnqueueSVMUnmap, &commandQueue, &svmPtr, &numEventsInWaitList, &eventWaitList, &event);
4800
4801 CommandQueue *pCommandQueue = nullptr;
4802
4803 cl_int retVal = validateObjects(
4804 WithCastToInternal(commandQueue, &pCommandQueue),
4805 EventWaitList(numEventsInWaitList, eventWaitList),
4806 svmPtr);
4807
4808 API_ENTER(&retVal);
4809
4810 DBG_LOG_INPUTS("commandQueue", commandQueue,
4811 "svmPtr", svmPtr,
4812 "numEventsInWaitList", numEventsInWaitList,
4813 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
4814 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
4815
4816 if (retVal != CL_SUCCESS) {
4817 TRACING_EXIT(clEnqueueSVMUnmap, &retVal);
4818 return retVal;
4819 }
4820
4821 auto &device = pCommandQueue->getDevice();
4822 if (!device.getHardwareInfo().capabilityTable.ftrSvm) {
4823 retVal = CL_INVALID_OPERATION;
4824 TRACING_EXIT(clEnqueueSVMUnmap, &retVal);
4825 return retVal;
4826 }
4827
4828 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_MAP_BUFFER_INTEL, numEventsInWaitList, eventWaitList, event)) {
4829 retVal = CL_INVALID_OPERATION;
4830 TRACING_EXIT(clEnqueueSVMUnmap, &retVal);
4831 return retVal;
4832 }
4833
4834 retVal = pCommandQueue->enqueueSVMUnmap(
4835 svmPtr,
4836 numEventsInWaitList,
4837 eventWaitList,
4838 event,
4839 true);
4840
4841 TRACING_EXIT(clEnqueueSVMUnmap, &retVal);
4842 return retVal;
4843 }
4844
clSetKernelArgSVMPointer(cl_kernel kernel,cl_uint argIndex,const void * argValue)4845 cl_int CL_API_CALL clSetKernelArgSVMPointer(cl_kernel kernel,
4846 cl_uint argIndex,
4847 const void *argValue) {
4848 TRACING_ENTER(clSetKernelArgSVMPointer, &kernel, &argIndex, &argValue);
4849
4850 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
4851
4852 auto retVal = validateObjects(WithCastToInternal(kernel, &pMultiDeviceKernel));
4853 API_ENTER(&retVal);
4854
4855 DBG_LOG_INPUTS("kernel", kernel, "argIndex", argIndex, "argValue", argValue);
4856
4857 if (CL_SUCCESS != retVal) {
4858 TRACING_EXIT(clSetKernelArgSVMPointer, &retVal);
4859 return retVal;
4860 }
4861
4862 for (const auto &pDevice : pMultiDeviceKernel->getDevices()) {
4863 const HardwareInfo &hwInfo = pDevice->getHardwareInfo();
4864 if (!hwInfo.capabilityTable.ftrSvm) {
4865 retVal = CL_INVALID_OPERATION;
4866 TRACING_EXIT(clSetKernelArgSVMPointer, &retVal);
4867 return retVal;
4868 }
4869 }
4870
4871 if (argIndex >= pMultiDeviceKernel->getKernelArgsNumber()) {
4872 retVal = CL_INVALID_ARG_INDEX;
4873 TRACING_EXIT(clSetKernelArgSVMPointer, &retVal);
4874 return retVal;
4875 }
4876
4877 for (const auto &pDevice : pMultiDeviceKernel->getDevices()) {
4878 auto pKernel = pMultiDeviceKernel->getKernel(pDevice->getRootDeviceIndex());
4879 cl_int kernelArgAddressQualifier = asClKernelArgAddressQualifier(pKernel->getKernelInfo()
4880 .kernelDescriptor.payloadMappings.explicitArgs[argIndex]
4881 .getTraits()
4882 .getAddressQualifier());
4883 if ((kernelArgAddressQualifier != CL_KERNEL_ARG_ADDRESS_GLOBAL) &&
4884 (kernelArgAddressQualifier != CL_KERNEL_ARG_ADDRESS_CONSTANT)) {
4885 retVal = CL_INVALID_ARG_VALUE;
4886 TRACING_EXIT(clSetKernelArgSVMPointer, &retVal);
4887 return retVal;
4888 }
4889 }
4890
4891 MultiGraphicsAllocation *pSvmAllocs = nullptr;
4892 if (argValue != nullptr) {
4893 auto svmManager = pMultiDeviceKernel->getContext().getSVMAllocsManager();
4894 auto svmData = svmManager->getSVMAlloc(argValue);
4895 if (svmData == nullptr) {
4896 for (const auto &pDevice : pMultiDeviceKernel->getDevices()) {
4897 if (!pDevice->areSharedSystemAllocationsAllowed()) {
4898 retVal = CL_INVALID_ARG_VALUE;
4899 TRACING_EXIT(clSetKernelArgSVMPointer, &retVal);
4900 return retVal;
4901 }
4902 }
4903 } else {
4904 pSvmAllocs = &svmData->gpuAllocations;
4905 }
4906 }
4907
4908 retVal = pMultiDeviceKernel->setArgSvmAlloc(argIndex, const_cast<void *>(argValue), pSvmAllocs);
4909 TRACING_EXIT(clSetKernelArgSVMPointer, &retVal);
4910 return retVal;
4911 }
4912
clSetKernelExecInfo(cl_kernel kernel,cl_kernel_exec_info paramName,size_t paramValueSize,const void * paramValue)4913 cl_int CL_API_CALL clSetKernelExecInfo(cl_kernel kernel,
4914 cl_kernel_exec_info paramName,
4915 size_t paramValueSize,
4916 const void *paramValue) {
4917 TRACING_ENTER(clSetKernelExecInfo, &kernel, ¶mName, ¶mValueSize, ¶mValue);
4918
4919 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
4920 auto retVal = validateObjects(WithCastToInternal(kernel, &pMultiDeviceKernel));
4921 API_ENTER(&retVal);
4922
4923 DBG_LOG_INPUTS("kernel", kernel, "paramName", paramName,
4924 "paramValueSize", paramValueSize, "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize));
4925
4926 if (CL_SUCCESS != retVal) {
4927 TRACING_EXIT(clSetKernelExecInfo, &retVal);
4928 return retVal;
4929 }
4930
4931 switch (paramName) {
4932 case CL_KERNEL_EXEC_INFO_SVM_PTRS:
4933 case CL_KERNEL_EXEC_INFO_SVM_FINE_GRAIN_SYSTEM:
4934 for (const auto &pDevice : pMultiDeviceKernel->getDevices()) {
4935 const HardwareInfo &hwInfo = pDevice->getHardwareInfo();
4936 if (!hwInfo.capabilityTable.ftrSvm) {
4937 retVal = CL_INVALID_OPERATION;
4938 TRACING_EXIT(clSetKernelExecInfo, &retVal);
4939 return retVal;
4940 }
4941 }
4942 }
4943
4944 switch (paramName) {
4945 case CL_KERNEL_EXEC_INFO_INDIRECT_DEVICE_ACCESS_INTEL:
4946 case CL_KERNEL_EXEC_INFO_INDIRECT_HOST_ACCESS_INTEL:
4947 case CL_KERNEL_EXEC_INFO_INDIRECT_SHARED_ACCESS_INTEL: {
4948 if (NEO::DebugManager.flags.DisableIndirectAccess.get() != 1 && pMultiDeviceKernel->getHasIndirectAccess() == true) {
4949 auto propertyValue = *reinterpret_cast<const cl_bool *>(paramValue);
4950 pMultiDeviceKernel->setUnifiedMemoryProperty(paramName, propertyValue);
4951 }
4952 } break;
4953
4954 case CL_KERNEL_EXEC_INFO_SVM_PTRS:
4955 case CL_KERNEL_EXEC_INFO_USM_PTRS_INTEL: {
4956 if ((paramValueSize == 0) ||
4957 (paramValueSize % sizeof(void *)) ||
4958 (paramValue == nullptr)) {
4959 retVal = CL_INVALID_VALUE;
4960 TRACING_EXIT(clSetKernelExecInfo, &retVal);
4961 return retVal;
4962 }
4963 size_t numPointers = paramValueSize / sizeof(void *);
4964 size_t *pSvmPtrList = (size_t *)paramValue;
4965
4966 if (paramName == CL_KERNEL_EXEC_INFO_SVM_PTRS) {
4967 pMultiDeviceKernel->clearSvmKernelExecInfo();
4968 } else {
4969 pMultiDeviceKernel->clearUnifiedMemoryExecInfo();
4970 }
4971
4972 for (uint32_t i = 0; i < numPointers; i++) {
4973 auto svmData = pMultiDeviceKernel->getContext().getSVMAllocsManager()->getSVMAlloc((const void *)pSvmPtrList[i]);
4974 if (svmData == nullptr) {
4975 retVal = CL_INVALID_VALUE;
4976 TRACING_EXIT(clSetKernelExecInfo, &retVal);
4977 return retVal;
4978 }
4979 auto &svmAllocs = svmData->gpuAllocations;
4980
4981 if (paramName == CL_KERNEL_EXEC_INFO_SVM_PTRS) {
4982 pMultiDeviceKernel->setSvmKernelExecInfo(svmAllocs);
4983 } else {
4984 pMultiDeviceKernel->setUnifiedMemoryExecInfo(svmAllocs);
4985 }
4986 }
4987 break;
4988 }
4989 case CL_KERNEL_EXEC_INFO_THREAD_ARBITRATION_POLICY_INTEL: {
4990 auto propertyValue = *static_cast<const uint32_t *>(paramValue);
4991 retVal = pMultiDeviceKernel->setKernelThreadArbitrationPolicy(propertyValue);
4992 return retVal;
4993 }
4994 case CL_KERNEL_EXEC_INFO_SVM_FINE_GRAIN_SYSTEM: {
4995 retVal = CL_INVALID_OPERATION;
4996 TRACING_EXIT(clSetKernelExecInfo, &retVal);
4997 return retVal;
4998 }
4999 case CL_KERNEL_EXEC_INFO_KERNEL_TYPE_INTEL: {
5000 if (paramValueSize != sizeof(cl_execution_info_kernel_type_intel) ||
5001 paramValue == nullptr) {
5002 retVal = CL_INVALID_VALUE;
5003 TRACING_EXIT(clSetKernelExecInfo, &retVal);
5004 return retVal;
5005 }
5006 auto kernelType = *static_cast<const cl_execution_info_kernel_type_intel *>(paramValue);
5007 retVal = pMultiDeviceKernel->setKernelExecutionType(kernelType);
5008 TRACING_EXIT(clSetKernelExecInfo, &retVal);
5009 return retVal;
5010 }
5011 default: {
5012 retVal = pMultiDeviceKernel->setAdditionalKernelExecInfoWithParam(paramName, paramValueSize, paramValue);
5013 TRACING_EXIT(clSetKernelExecInfo, &retVal);
5014 return retVal;
5015 }
5016 }
5017
5018 TRACING_EXIT(clSetKernelExecInfo, &retVal);
5019 return retVal;
5020 };
5021
clCreatePipe(cl_context context,cl_mem_flags flags,cl_uint pipePacketSize,cl_uint pipeMaxPackets,const cl_pipe_properties * properties,cl_int * errcodeRet)5022 cl_mem CL_API_CALL clCreatePipe(cl_context context,
5023 cl_mem_flags flags,
5024 cl_uint pipePacketSize,
5025 cl_uint pipeMaxPackets,
5026 const cl_pipe_properties *properties,
5027 cl_int *errcodeRet) {
5028 TRACING_ENTER(clCreatePipe, &context, &flags, &pipePacketSize, &pipeMaxPackets, &properties, &errcodeRet);
5029 cl_mem pipe = nullptr;
5030 cl_int retVal = CL_SUCCESS;
5031 API_ENTER(&retVal);
5032
5033 DBG_LOG_INPUTS("cl_context", context,
5034 "cl_mem_flags", flags,
5035 "cl_uint", pipePacketSize,
5036 "cl_uint", pipeMaxPackets,
5037 "const cl_pipe_properties", properties,
5038 "cl_int", errcodeRet);
5039
5040 Context *pContext = nullptr;
5041
5042 const cl_mem_flags allValidFlags =
5043 CL_MEM_READ_WRITE | CL_MEM_HOST_NO_ACCESS;
5044
5045 do {
5046 if ((pipePacketSize == 0) || (pipeMaxPackets == 0)) {
5047 retVal = CL_INVALID_PIPE_SIZE;
5048 break;
5049 }
5050
5051 /* Are there some invalid flag bits? */
5052 if ((flags & (~allValidFlags)) != 0) {
5053 retVal = CL_INVALID_VALUE;
5054 break;
5055 }
5056
5057 if (properties != nullptr) {
5058 retVal = CL_INVALID_VALUE;
5059 break;
5060 }
5061
5062 retVal = validateObjects(WithCastToInternal(context, &pContext));
5063 if (retVal != CL_SUCCESS) {
5064 break;
5065 }
5066 auto pDevice = pContext->getDevice(0);
5067
5068 if (pDevice->arePipesSupported() == false) {
5069 retVal = CL_INVALID_OPERATION;
5070 break;
5071 }
5072
5073 if (pipePacketSize > pDevice->getDeviceInfo().pipeMaxPacketSize) {
5074 retVal = CL_INVALID_PIPE_SIZE;
5075 break;
5076 }
5077
5078 // create the pipe
5079 pipe = Pipe::create(pContext, flags, pipePacketSize, pipeMaxPackets, properties, retVal);
5080 } while (false);
5081
5082 if (errcodeRet) {
5083 *errcodeRet = retVal;
5084 }
5085 DBG_LOG_INPUTS("pipe", pipe);
5086 TRACING_EXIT(clCreatePipe, &pipe);
5087 return pipe;
5088 }
5089
clGetPipeInfo(cl_mem pipe,cl_pipe_info paramName,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)5090 cl_int CL_API_CALL clGetPipeInfo(cl_mem pipe,
5091 cl_pipe_info paramName,
5092 size_t paramValueSize,
5093 void *paramValue,
5094 size_t *paramValueSizeRet) {
5095 TRACING_ENTER(clGetPipeInfo, &pipe, ¶mName, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
5096
5097 cl_int retVal = CL_SUCCESS;
5098 API_ENTER(&retVal);
5099
5100 DBG_LOG_INPUTS("cl_mem", pipe,
5101 "cl_pipe_info", paramName,
5102 "size_t", paramValueSize,
5103 "void *", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
5104 "size_t*", paramValueSizeRet);
5105
5106 retVal = validateObjects(pipe);
5107 if (CL_SUCCESS != retVal) {
5108 TRACING_EXIT(clGetPipeInfo, &retVal);
5109 return retVal;
5110 }
5111
5112 auto pPipeObj = castToObject<Pipe>(pipe);
5113
5114 if (pPipeObj == nullptr) {
5115 retVal = CL_INVALID_MEM_OBJECT;
5116 TRACING_EXIT(clGetPipeInfo, &retVal);
5117 return retVal;
5118 }
5119
5120 retVal = pPipeObj->getPipeInfo(paramName, paramValueSize, paramValue, paramValueSizeRet);
5121 TRACING_EXIT(clGetPipeInfo, &retVal);
5122 return retVal;
5123 }
5124
clCreateCommandQueueWithProperties(cl_context context,cl_device_id device,const cl_queue_properties * properties,cl_int * errcodeRet)5125 cl_command_queue CL_API_CALL clCreateCommandQueueWithProperties(cl_context context,
5126 cl_device_id device,
5127 const cl_queue_properties *properties,
5128 cl_int *errcodeRet) {
5129 TRACING_ENTER(clCreateCommandQueueWithProperties, &context, &device, &properties, &errcodeRet);
5130 cl_int retVal = CL_SUCCESS;
5131 API_ENTER(&retVal);
5132 DBG_LOG_INPUTS("context", context,
5133 "device", device,
5134 "properties", properties);
5135
5136 cl_command_queue commandQueue = nullptr;
5137 ErrorCodeHelper err(errcodeRet, CL_SUCCESS);
5138
5139 Context *pContext = nullptr;
5140 ClDevice *pDevice = nullptr;
5141
5142 retVal = validateObjects(
5143 WithCastToInternal(context, &pContext),
5144 WithCastToInternal(device, &pDevice));
5145
5146 if (CL_SUCCESS != retVal) {
5147 err.set(retVal);
5148 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5149 return commandQueue;
5150 }
5151
5152 if (!pContext->isDeviceAssociated(*pDevice)) {
5153 err.set(CL_INVALID_DEVICE);
5154 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5155 return commandQueue;
5156 }
5157
5158 auto minimumCreateDeviceQueueFlags = static_cast<cl_command_queue_properties>(CL_QUEUE_ON_DEVICE |
5159 CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE);
5160 auto tokenValue = properties ? *properties : 0;
5161 auto propertiesAddress = properties;
5162
5163 while (tokenValue != 0) {
5164 if (tokenValue != CL_QUEUE_PROPERTIES &&
5165 tokenValue != CL_QUEUE_SIZE &&
5166 tokenValue != CL_QUEUE_PRIORITY_KHR &&
5167 tokenValue != CL_QUEUE_THROTTLE_KHR &&
5168 tokenValue != CL_QUEUE_SLICE_COUNT_INTEL &&
5169 tokenValue != CL_QUEUE_FAMILY_INTEL &&
5170 tokenValue != CL_QUEUE_INDEX_INTEL) {
5171 err.set(CL_INVALID_VALUE);
5172 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5173 return commandQueue;
5174 }
5175
5176 propertiesAddress += 2;
5177 tokenValue = *propertiesAddress;
5178 }
5179
5180 auto commandQueueProperties = getCmdQueueProperties<cl_command_queue_properties>(properties);
5181 uint32_t maxOnDeviceQueueSize = pDevice->getDeviceInfo().queueOnDeviceMaxSize;
5182
5183 if (commandQueueProperties & static_cast<cl_command_queue_properties>(CL_QUEUE_ON_DEVICE)) {
5184 if (!(commandQueueProperties & static_cast<cl_command_queue_properties>(CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE))) {
5185 err.set(CL_INVALID_VALUE);
5186 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5187 return commandQueue;
5188 }
5189 if (!pDevice->isDeviceEnqueueSupported()) {
5190 err.set(CL_INVALID_QUEUE_PROPERTIES);
5191 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5192 return commandQueue;
5193 }
5194 }
5195
5196 if (commandQueueProperties & static_cast<cl_command_queue_properties>(CL_QUEUE_ON_DEVICE_DEFAULT)) {
5197 if (!(commandQueueProperties & static_cast<cl_command_queue_properties>(CL_QUEUE_ON_DEVICE))) {
5198 err.set(CL_INVALID_VALUE);
5199 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5200 return commandQueue;
5201 }
5202 } else if (commandQueueProperties & static_cast<cl_command_queue_properties>(CL_QUEUE_ON_DEVICE)) {
5203 if (pContext->getDefaultDeviceQueue()) {
5204 err.set(CL_OUT_OF_RESOURCES);
5205 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5206 return commandQueue;
5207 }
5208 }
5209
5210 if (getCmdQueueProperties<cl_command_queue_properties>(properties, CL_QUEUE_SIZE) > maxOnDeviceQueueSize) {
5211 err.set(CL_INVALID_QUEUE_PROPERTIES);
5212 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5213 return commandQueue;
5214 }
5215
5216 if (commandQueueProperties & static_cast<cl_command_queue_properties>(CL_QUEUE_ON_DEVICE)) {
5217 if (getCmdQueueProperties<cl_queue_priority_khr>(properties, CL_QUEUE_PRIORITY_KHR)) {
5218 err.set(CL_INVALID_QUEUE_PROPERTIES);
5219 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5220 return commandQueue;
5221 }
5222 }
5223
5224 if (commandQueueProperties & static_cast<cl_command_queue_properties>(CL_QUEUE_ON_DEVICE)) {
5225 if (getCmdQueueProperties<cl_queue_throttle_khr>(properties, CL_QUEUE_THROTTLE_KHR)) {
5226 err.set(CL_INVALID_QUEUE_PROPERTIES);
5227 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5228 return commandQueue;
5229 }
5230 }
5231
5232 if (getCmdQueueProperties<cl_command_queue_properties>(properties, CL_QUEUE_SLICE_COUNT_INTEL) > pDevice->getDeviceInfo().maxSliceCount) {
5233 err.set(CL_INVALID_QUEUE_PROPERTIES);
5234 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5235 return commandQueue;
5236 }
5237
5238 bool queueFamilySelected = false;
5239 bool queueSelected = false;
5240 const auto queueFamilyIndex = getCmdQueueProperties<cl_uint>(properties, CL_QUEUE_FAMILY_INTEL, &queueFamilySelected);
5241 const auto queueIndex = getCmdQueueProperties<cl_uint>(properties, CL_QUEUE_INDEX_INTEL, &queueSelected);
5242 if (queueFamilySelected != queueSelected) {
5243 err.set(CL_INVALID_QUEUE_PROPERTIES);
5244 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5245 return commandQueue;
5246 }
5247 if (queueFamilySelected &&
5248 (queueFamilyIndex >= pDevice->getDeviceInfo().queueFamilyProperties.size() ||
5249 queueIndex >= pDevice->getDeviceInfo().queueFamilyProperties[queueFamilyIndex].count)) {
5250 err.set(CL_INVALID_QUEUE_PROPERTIES);
5251 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5252 return commandQueue;
5253 }
5254
5255 auto maskedFlags = commandQueueProperties & minimumCreateDeviceQueueFlags;
5256
5257 if (maskedFlags == minimumCreateDeviceQueueFlags) {
5258 commandQueue = DeviceQueue::create(
5259 pContext,
5260 pDevice,
5261 *properties,
5262 retVal);
5263
5264 } else {
5265 commandQueue = CommandQueue::create(
5266 pContext,
5267 pDevice,
5268 properties,
5269 false,
5270 retVal);
5271 if (pContext->isProvidingPerformanceHints()) {
5272 pContext->providePerformanceHint(CL_CONTEXT_DIAGNOSTICS_LEVEL_NEUTRAL_INTEL, DRIVER_CALLS_INTERNAL_CL_FLUSH);
5273 if (castToObjectOrAbort<CommandQueue>(commandQueue)->isProfilingEnabled()) {
5274 pContext->providePerformanceHint(CL_CONTEXT_DIAGNOSTICS_LEVEL_NEUTRAL_INTEL, PROFILING_ENABLED);
5275 if (pDevice->getDeviceInfo().preemptionSupported && pDevice->getHardwareInfo().platform.eProductFamily < IGFX_SKYLAKE) {
5276 pContext->providePerformanceHint(CL_CONTEXT_DIAGNOSTICS_LEVEL_NEUTRAL_INTEL, PROFILING_ENABLED_WITH_DISABLED_PREEMPTION);
5277 }
5278 }
5279 }
5280 }
5281
5282 if (!commandQueue)
5283 retVal = CL_OUT_OF_HOST_MEMORY;
5284
5285 DBG_LOG_INPUTS("commandQueue", commandQueue, "properties", static_cast<int>(getCmdQueueProperties<cl_command_queue_properties>(properties)));
5286
5287 err.set(retVal);
5288
5289 TRACING_EXIT(clCreateCommandQueueWithProperties, &commandQueue);
5290 return commandQueue;
5291 }
5292
clCreateSamplerWithProperties(cl_context context,const cl_sampler_properties * samplerProperties,cl_int * errcodeRet)5293 cl_sampler CL_API_CALL clCreateSamplerWithProperties(cl_context context,
5294 const cl_sampler_properties *samplerProperties,
5295 cl_int *errcodeRet) {
5296 TRACING_ENTER(clCreateSamplerWithProperties, &context, &samplerProperties, &errcodeRet);
5297 cl_int retVal = CL_SUCCESS;
5298 API_ENTER(&retVal);
5299 DBG_LOG_INPUTS("context", context,
5300 "samplerProperties", samplerProperties);
5301 cl_sampler sampler = nullptr;
5302 retVal = validateObjects(context);
5303
5304 if (CL_SUCCESS == retVal) {
5305 sampler = Sampler::create(
5306 castToObject<Context>(context),
5307 samplerProperties,
5308 retVal);
5309 }
5310
5311 if (errcodeRet) {
5312 *errcodeRet = retVal;
5313 }
5314
5315 TRACING_EXIT(clCreateSamplerWithProperties, &sampler);
5316 return sampler;
5317 }
5318
clUnloadCompiler()5319 cl_int CL_API_CALL clUnloadCompiler() {
5320 TRACING_ENTER(clUnloadCompiler);
5321 cl_int retVal = CL_SUCCESS;
5322 API_ENTER(&retVal);
5323 TRACING_EXIT(clUnloadCompiler, &retVal);
5324 return retVal;
5325 }
5326
clGetKernelSubGroupInfoKHR(cl_kernel kernel,cl_device_id device,cl_kernel_sub_group_info paramName,size_t inputValueSize,const void * inputValue,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)5327 cl_int CL_API_CALL clGetKernelSubGroupInfoKHR(cl_kernel kernel,
5328 cl_device_id device,
5329 cl_kernel_sub_group_info paramName,
5330 size_t inputValueSize,
5331 const void *inputValue,
5332 size_t paramValueSize,
5333 void *paramValue,
5334 size_t *paramValueSizeRet) {
5335 cl_int retVal = CL_SUCCESS;
5336 API_ENTER(&retVal);
5337 DBG_LOG_INPUTS("kernel", kernel,
5338 "device", device,
5339 "paramName", paramName,
5340 "inputValueSize", inputValueSize,
5341 "inputValue", NEO::FileLoggerInstance().infoPointerToString(inputValue, inputValueSize),
5342 "paramValueSize", paramValueSize,
5343 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
5344 "paramValueSizeRet", paramValueSizeRet);
5345
5346 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
5347 retVal = validateObjects(WithCastToInternal(kernel, &pMultiDeviceKernel));
5348
5349 ClDevice *pClDevice = nullptr;
5350 if (CL_SUCCESS == retVal) {
5351 if (pMultiDeviceKernel->getDevices().size() == 1u && !device) {
5352 pClDevice = pMultiDeviceKernel->getDevices()[0];
5353 } else {
5354 retVal = validateObjects(WithCastToInternal(device, &pClDevice));
5355 }
5356 }
5357
5358 if (CL_SUCCESS != retVal) {
5359 return retVal;
5360 }
5361 auto pKernel = pMultiDeviceKernel->getKernel(pClDevice->getRootDeviceIndex());
5362
5363 switch (paramName) {
5364 case CL_KERNEL_MAX_SUB_GROUP_SIZE_FOR_NDRANGE:
5365 case CL_KERNEL_SUB_GROUP_COUNT_FOR_NDRANGE:
5366 case CL_KERNEL_COMPILE_SUB_GROUP_SIZE_INTEL:
5367 return pKernel->getSubGroupInfo(paramName,
5368 inputValueSize, inputValue,
5369 paramValueSize, paramValue,
5370 paramValueSizeRet);
5371 default: {
5372 retVal = CL_INVALID_VALUE;
5373 return retVal;
5374 }
5375 }
5376 }
5377
clGetDeviceAndHostTimer(cl_device_id device,cl_ulong * deviceTimestamp,cl_ulong * hostTimestamp)5378 cl_int CL_API_CALL clGetDeviceAndHostTimer(cl_device_id device,
5379 cl_ulong *deviceTimestamp,
5380 cl_ulong *hostTimestamp) {
5381 TRACING_ENTER(clGetDeviceAndHostTimer, &device, &deviceTimestamp, &hostTimestamp);
5382 cl_int retVal = CL_SUCCESS;
5383 API_ENTER(&retVal);
5384 DBG_LOG_INPUTS("device", device,
5385 "deviceTimestamp", deviceTimestamp,
5386 "hostTimestamp", hostTimestamp);
5387 do {
5388 ClDevice *pDevice = castToObject<ClDevice>(device);
5389 if (pDevice == nullptr) {
5390 retVal = CL_INVALID_DEVICE;
5391 break;
5392 }
5393 if (deviceTimestamp == nullptr || hostTimestamp == nullptr) {
5394 retVal = CL_INVALID_VALUE;
5395 break;
5396 }
5397 if (!pDevice->getDeviceAndHostTimer(static_cast<uint64_t *>(deviceTimestamp), static_cast<uint64_t *>(hostTimestamp))) {
5398 retVal = CL_OUT_OF_RESOURCES;
5399 break;
5400 }
5401 } while (false);
5402
5403 TRACING_EXIT(clGetDeviceAndHostTimer, &retVal);
5404 return retVal;
5405 }
5406
clGetHostTimer(cl_device_id device,cl_ulong * hostTimestamp)5407 cl_int CL_API_CALL clGetHostTimer(cl_device_id device,
5408 cl_ulong *hostTimestamp) {
5409 TRACING_ENTER(clGetHostTimer, &device, &hostTimestamp);
5410 cl_int retVal = CL_SUCCESS;
5411 API_ENTER(&retVal);
5412 DBG_LOG_INPUTS("device", device,
5413 "hostTimestamp", hostTimestamp);
5414
5415 do {
5416 ClDevice *pDevice = castToObject<ClDevice>(device);
5417 if (pDevice == nullptr) {
5418 retVal = CL_INVALID_DEVICE;
5419 break;
5420 }
5421 if (hostTimestamp == nullptr) {
5422 retVal = CL_INVALID_VALUE;
5423 break;
5424 }
5425 if (!pDevice->getHostTimer(static_cast<uint64_t *>(hostTimestamp))) {
5426 retVal = CL_OUT_OF_RESOURCES;
5427 break;
5428 }
5429 } while (false);
5430
5431 TRACING_EXIT(clGetHostTimer, &retVal);
5432 return retVal;
5433 }
5434
clGetKernelSubGroupInfo(cl_kernel kernel,cl_device_id device,cl_kernel_sub_group_info paramName,size_t inputValueSize,const void * inputValue,size_t paramValueSize,void * paramValue,size_t * paramValueSizeRet)5435 cl_int CL_API_CALL clGetKernelSubGroupInfo(cl_kernel kernel,
5436 cl_device_id device,
5437 cl_kernel_sub_group_info paramName,
5438 size_t inputValueSize,
5439 const void *inputValue,
5440 size_t paramValueSize,
5441 void *paramValue,
5442 size_t *paramValueSizeRet) {
5443 TRACING_ENTER(clGetKernelSubGroupInfo, &kernel, &device, ¶mName, &inputValueSize, &inputValue, ¶mValueSize, ¶mValue, ¶mValueSizeRet);
5444 cl_int retVal = CL_SUCCESS;
5445 API_ENTER(&retVal);
5446 DBG_LOG_INPUTS("kernel", kernel,
5447 "device", device,
5448 "paramName", paramName,
5449 "inputValueSize", inputValueSize,
5450 "inputValue", NEO::FileLoggerInstance().infoPointerToString(inputValue, inputValueSize),
5451 "paramValueSize", paramValueSize,
5452 "paramValue", NEO::FileLoggerInstance().infoPointerToString(paramValue, paramValueSize),
5453 "paramValueSizeRet", paramValueSizeRet);
5454
5455 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
5456 retVal = validateObjects(WithCastToInternal(kernel, &pMultiDeviceKernel));
5457
5458 ClDevice *pClDevice = nullptr;
5459 if (CL_SUCCESS == retVal) {
5460 if (pMultiDeviceKernel->getDevices().size() == 1u && !device) {
5461 pClDevice = pMultiDeviceKernel->getDevices()[0];
5462 } else {
5463 retVal = validateObjects(WithCastToInternal(device, &pClDevice));
5464 }
5465 }
5466
5467 if (CL_SUCCESS != retVal) {
5468 TRACING_EXIT(clGetKernelSubGroupInfo, &retVal);
5469 return retVal;
5470 }
5471
5472 auto pKernel = pMultiDeviceKernel->getKernel(pClDevice->getRootDeviceIndex());
5473 retVal = pKernel->getSubGroupInfo(paramName,
5474 inputValueSize, inputValue,
5475 paramValueSize, paramValue,
5476 paramValueSizeRet);
5477 TRACING_EXIT(clGetKernelSubGroupInfo, &retVal);
5478 return retVal;
5479 }
5480
clSetDefaultDeviceCommandQueue(cl_context context,cl_device_id device,cl_command_queue commandQueue)5481 cl_int CL_API_CALL clSetDefaultDeviceCommandQueue(cl_context context,
5482 cl_device_id device,
5483 cl_command_queue commandQueue) {
5484 TRACING_ENTER(clSetDefaultDeviceCommandQueue, &context, &device, &commandQueue);
5485
5486 cl_int retVal = CL_SUCCESS;
5487 API_ENTER(&retVal);
5488 DBG_LOG_INPUTS("context", context,
5489 "device", device,
5490 "commandQueue", commandQueue);
5491
5492 Context *pContext = nullptr;
5493 ClDevice *pClDevice = nullptr;
5494
5495 retVal = validateObjects(WithCastToInternal(context, &pContext),
5496 WithCastToInternal(device, &pClDevice));
5497
5498 if (CL_SUCCESS != retVal) {
5499 TRACING_EXIT(clSetDefaultDeviceCommandQueue, &retVal);
5500 return retVal;
5501 }
5502
5503 if (pClDevice->isDeviceEnqueueSupported() == false) {
5504 retVal = CL_INVALID_OPERATION;
5505 TRACING_EXIT(clSetDefaultDeviceCommandQueue, &retVal);
5506 return retVal;
5507 }
5508
5509 auto pDeviceQueue = castToObject<DeviceQueue>(static_cast<_device_queue *>(commandQueue));
5510
5511 if (!pDeviceQueue) {
5512 retVal = CL_INVALID_COMMAND_QUEUE;
5513 TRACING_EXIT(clSetDefaultDeviceCommandQueue, &retVal);
5514 return retVal;
5515 }
5516
5517 if (&pDeviceQueue->getContext() != pContext) {
5518 retVal = CL_INVALID_COMMAND_QUEUE;
5519 TRACING_EXIT(clSetDefaultDeviceCommandQueue, &retVal);
5520 return retVal;
5521 }
5522
5523 pContext->setDefaultDeviceQueue(pDeviceQueue);
5524
5525 retVal = CL_SUCCESS;
5526 TRACING_EXIT(clSetDefaultDeviceCommandQueue, &retVal);
5527 return retVal;
5528 }
5529
clEnqueueSVMMigrateMem(cl_command_queue commandQueue,cl_uint numSvmPointers,const void ** svmPointers,const size_t * sizes,const cl_mem_migration_flags flags,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)5530 cl_int CL_API_CALL clEnqueueSVMMigrateMem(cl_command_queue commandQueue,
5531 cl_uint numSvmPointers,
5532 const void **svmPointers,
5533 const size_t *sizes,
5534 const cl_mem_migration_flags flags,
5535 cl_uint numEventsInWaitList,
5536 const cl_event *eventWaitList,
5537 cl_event *event) {
5538 TRACING_ENTER(clEnqueueSVMMigrateMem, &commandQueue, &numSvmPointers, &svmPointers, &sizes, &flags, &numEventsInWaitList, &eventWaitList, &event);
5539 cl_int retVal = CL_SUCCESS;
5540 API_ENTER(&retVal);
5541 DBG_LOG_INPUTS("commandQueue", commandQueue,
5542 "numSvmPointers", numSvmPointers,
5543 "svmPointers", NEO::FileLoggerInstance().infoPointerToString(svmPointers ? svmPointers[0] : 0, NEO::FileLoggerInstance().getInput(sizes, 0)),
5544 "sizes", NEO::FileLoggerInstance().getInput(sizes, 0),
5545 "flags", flags,
5546 "numEventsInWaitList", numEventsInWaitList,
5547 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
5548 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
5549
5550 CommandQueue *pCommandQueue = nullptr;
5551 retVal = validateObjects(
5552 WithCastToInternal(commandQueue, &pCommandQueue),
5553 EventWaitList(numEventsInWaitList, eventWaitList));
5554
5555 if (CL_SUCCESS != retVal) {
5556 TRACING_EXIT(clEnqueueSVMMigrateMem, &retVal);
5557 return retVal;
5558 }
5559
5560 auto &device = pCommandQueue->getDevice();
5561 if (!device.getHardwareInfo().capabilityTable.ftrSvm) {
5562 retVal = CL_INVALID_OPERATION;
5563 TRACING_EXIT(clEnqueueSVMMigrateMem, &retVal);
5564 return retVal;
5565 }
5566
5567 if (numSvmPointers == 0 || svmPointers == nullptr) {
5568 retVal = CL_INVALID_VALUE;
5569 TRACING_EXIT(clEnqueueSVMMigrateMem, &retVal);
5570 return retVal;
5571 }
5572
5573 const cl_mem_migration_flags allValidFlags =
5574 CL_MIGRATE_MEM_OBJECT_HOST | CL_MIGRATE_MEM_OBJECT_CONTENT_UNDEFINED;
5575
5576 if ((flags & (~allValidFlags)) != 0) {
5577 retVal = CL_INVALID_VALUE;
5578 TRACING_EXIT(clEnqueueSVMMigrateMem, &retVal);
5579 return retVal;
5580 }
5581
5582 auto pSvmAllocMgr = pCommandQueue->getContext().getSVMAllocsManager();
5583 UNRECOVERABLE_IF(pSvmAllocMgr == nullptr);
5584
5585 for (uint32_t i = 0; i < numSvmPointers; i++) {
5586 auto svmData = pSvmAllocMgr->getSVMAlloc(svmPointers[i]);
5587 if (svmData == nullptr) {
5588 retVal = CL_INVALID_VALUE;
5589 TRACING_EXIT(clEnqueueSVMMigrateMem, &retVal);
5590 return retVal;
5591 }
5592 if (sizes != nullptr && sizes[i] != 0) {
5593 svmData = pSvmAllocMgr->getSVMAlloc(reinterpret_cast<void *>((size_t)svmPointers[i] + sizes[i] - 1));
5594 if (svmData == nullptr) {
5595 retVal = CL_INVALID_VALUE;
5596 TRACING_EXIT(clEnqueueSVMMigrateMem, &retVal);
5597 return retVal;
5598 }
5599 }
5600 }
5601
5602 for (uint32_t i = 0; i < numEventsInWaitList; i++) {
5603 auto pEvent = castToObject<Event>(eventWaitList[i]);
5604 if (pEvent->getContext() != &pCommandQueue->getContext()) {
5605 retVal = CL_INVALID_CONTEXT;
5606 TRACING_EXIT(clEnqueueSVMMigrateMem, &retVal);
5607 return retVal;
5608 }
5609 }
5610 retVal = pCommandQueue->enqueueSVMMigrateMem(numSvmPointers,
5611 svmPointers,
5612 sizes,
5613 flags,
5614 numEventsInWaitList,
5615 eventWaitList,
5616 event);
5617 TRACING_EXIT(clEnqueueSVMMigrateMem, &retVal);
5618 return retVal;
5619 }
5620
clCloneKernel(cl_kernel sourceKernel,cl_int * errcodeRet)5621 cl_kernel CL_API_CALL clCloneKernel(cl_kernel sourceKernel,
5622 cl_int *errcodeRet) {
5623 TRACING_ENTER(clCloneKernel, &sourceKernel, &errcodeRet);
5624 MultiDeviceKernel *pSourceMultiDeviceKernel = nullptr;
5625 MultiDeviceKernel *pClonedMultiDeviceKernel = nullptr;
5626
5627 auto retVal = validateObjects(WithCastToInternal(sourceKernel, &pSourceMultiDeviceKernel));
5628 API_ENTER(&retVal);
5629 DBG_LOG_INPUTS("sourceKernel", sourceKernel);
5630
5631 if (CL_SUCCESS == retVal) {
5632 pClonedMultiDeviceKernel = MultiDeviceKernel::create(pSourceMultiDeviceKernel->getProgram(),
5633 pSourceMultiDeviceKernel->getKernelInfos(),
5634 &retVal);
5635 UNRECOVERABLE_IF((pClonedMultiDeviceKernel == nullptr) || (retVal != CL_SUCCESS));
5636
5637 retVal = pClonedMultiDeviceKernel->cloneKernel(pSourceMultiDeviceKernel);
5638 }
5639
5640 if (errcodeRet) {
5641 *errcodeRet = retVal;
5642 }
5643 if (pClonedMultiDeviceKernel != nullptr) {
5644 gtpinNotifyKernelCreate(pClonedMultiDeviceKernel);
5645 }
5646
5647 TRACING_EXIT(clCloneKernel, (cl_kernel *)&pClonedMultiDeviceKernel);
5648 return pClonedMultiDeviceKernel;
5649 }
5650
clEnqueueVerifyMemoryINTEL(cl_command_queue commandQueue,const void * allocationPtr,const void * expectedData,size_t sizeOfComparison,cl_uint comparisonMode)5651 CL_API_ENTRY cl_int CL_API_CALL clEnqueueVerifyMemoryINTEL(cl_command_queue commandQueue,
5652 const void *allocationPtr,
5653 const void *expectedData,
5654 size_t sizeOfComparison,
5655 cl_uint comparisonMode) {
5656 cl_int retVal = CL_SUCCESS;
5657 API_ENTER(&retVal);
5658 DBG_LOG_INPUTS("commandQueue", commandQueue,
5659 "allocationPtr", allocationPtr,
5660 "expectedData", expectedData,
5661 "sizeOfComparison", sizeOfComparison,
5662 "comparisonMode", comparisonMode);
5663
5664 if (sizeOfComparison == 0 || expectedData == nullptr || allocationPtr == nullptr) {
5665 retVal = CL_INVALID_VALUE;
5666 return retVal;
5667 }
5668
5669 CommandQueue *pCommandQueue = nullptr;
5670 retVal = validateObjects(WithCastToInternal(commandQueue, &pCommandQueue));
5671 if (retVal != CL_SUCCESS) {
5672 return retVal;
5673 }
5674
5675 auto &csr = pCommandQueue->getGpgpuCommandStreamReceiver();
5676 auto status = csr.expectMemory(allocationPtr, expectedData, sizeOfComparison, comparisonMode);
5677 return status ? CL_SUCCESS : CL_INVALID_VALUE;
5678 }
5679
clAddCommentINTEL(cl_device_id device,const char * comment)5680 cl_int CL_API_CALL clAddCommentINTEL(cl_device_id device, const char *comment) {
5681 cl_int retVal = CL_SUCCESS;
5682 API_ENTER(&retVal);
5683 DBG_LOG_INPUTS("device", device, "comment", comment);
5684
5685 ClDevice *pDevice = nullptr;
5686 retVal = validateObjects(WithCastToInternal(device, &pDevice));
5687 if (retVal != CL_SUCCESS) {
5688 return retVal;
5689 }
5690 auto aubCenter = pDevice->getRootDeviceEnvironment().aubCenter.get();
5691
5692 if (!comment || (aubCenter && !aubCenter->getAubManager())) {
5693 retVal = CL_INVALID_VALUE;
5694 }
5695
5696 if (retVal == CL_SUCCESS && aubCenter) {
5697 aubCenter->getAubManager()->addComment(comment);
5698 }
5699
5700 return retVal;
5701 }
5702
clGetDeviceGlobalVariablePointerINTEL(cl_device_id device,cl_program program,const char * globalVariableName,size_t * globalVariableSizeRet,void ** globalVariablePointerRet)5703 cl_int CL_API_CALL clGetDeviceGlobalVariablePointerINTEL(
5704 cl_device_id device,
5705 cl_program program,
5706 const char *globalVariableName,
5707 size_t *globalVariableSizeRet,
5708 void **globalVariablePointerRet) {
5709 cl_int retVal = CL_SUCCESS;
5710 API_ENTER(&retVal);
5711 DBG_LOG_INPUTS("device", device, "program", program,
5712 "globalVariableName", globalVariableName,
5713 "globalVariablePointerRet", globalVariablePointerRet);
5714 Program *pProgram = nullptr;
5715 ClDevice *pDevice = nullptr;
5716 retVal = validateObjects(WithCastToInternal(program, &pProgram), WithCastToInternal(device, &pDevice));
5717 if (globalVariablePointerRet == nullptr) {
5718 retVal = CL_INVALID_ARG_VALUE;
5719 }
5720
5721 if (CL_SUCCESS == retVal) {
5722 const auto &symbols = pProgram->getSymbols(pDevice->getRootDeviceIndex());
5723 auto symbolIt = symbols.find(globalVariableName);
5724 if ((symbolIt == symbols.end()) || (symbolIt->second.symbol.segment == NEO::SegmentType::Instructions)) {
5725 retVal = CL_INVALID_ARG_VALUE;
5726 } else {
5727 if (globalVariableSizeRet != nullptr) {
5728 *globalVariableSizeRet = symbolIt->second.symbol.size;
5729 }
5730 *globalVariablePointerRet = reinterpret_cast<void *>(symbolIt->second.gpuAddress);
5731 }
5732 }
5733
5734 return retVal;
5735 }
5736
clGetDeviceFunctionPointerINTEL(cl_device_id device,cl_program program,const char * functionName,cl_ulong * functionPointerRet)5737 cl_int CL_API_CALL clGetDeviceFunctionPointerINTEL(
5738 cl_device_id device,
5739 cl_program program,
5740 const char *functionName,
5741 cl_ulong *functionPointerRet) {
5742 cl_int retVal = CL_SUCCESS;
5743 API_ENTER(&retVal);
5744 DBG_LOG_INPUTS("device", device, "program", program,
5745 "functionName", functionName,
5746 "functionPointerRet", functionPointerRet);
5747
5748 Program *pProgram = nullptr;
5749 ClDevice *pDevice = nullptr;
5750 retVal = validateObjects(WithCastToInternal(program, &pProgram), WithCastToInternal(device, &pDevice));
5751 if ((CL_SUCCESS == retVal) && (functionPointerRet == nullptr)) {
5752 retVal = CL_INVALID_ARG_VALUE;
5753 }
5754
5755 if (CL_SUCCESS == retVal) {
5756 const auto &symbols = pProgram->getSymbols(pDevice->getRootDeviceIndex());
5757 auto symbolIt = symbols.find(functionName);
5758 if ((symbolIt == symbols.end()) || (symbolIt->second.symbol.segment != NEO::SegmentType::Instructions)) {
5759 retVal = CL_INVALID_ARG_VALUE;
5760 } else {
5761 *functionPointerRet = static_cast<cl_ulong>(symbolIt->second.gpuAddress);
5762 }
5763 }
5764
5765 return retVal;
5766 }
5767
clSetProgramReleaseCallback(cl_program program,void (CL_CALLBACK * pfnNotify)(cl_program,void *),void * userData)5768 cl_int CL_API_CALL clSetProgramReleaseCallback(cl_program program,
5769 void(CL_CALLBACK *pfnNotify)(cl_program /* program */, void * /* user_data */),
5770 void *userData) {
5771 DBG_LOG_INPUTS("program", program,
5772 "pfnNotify", pfnNotify,
5773 "userData", userData);
5774
5775 cl_int retVal = CL_SUCCESS;
5776 API_ENTER(&retVal);
5777
5778 Program *pProgram = nullptr;
5779 retVal = validateObjects(WithCastToInternal(program, &pProgram),
5780 reinterpret_cast<void *>(pfnNotify));
5781
5782 if (retVal == CL_SUCCESS) {
5783 retVal = CL_INVALID_OPERATION;
5784 }
5785
5786 return retVal;
5787 }
5788
clSetProgramSpecializationConstant(cl_program program,cl_uint specId,size_t specSize,const void * specValue)5789 cl_int CL_API_CALL clSetProgramSpecializationConstant(cl_program program, cl_uint specId, size_t specSize, const void *specValue) {
5790 cl_int retVal = CL_SUCCESS;
5791 API_ENTER(&retVal);
5792 DBG_LOG_INPUTS("program", program,
5793 "specId", specId,
5794 "specSize", specSize,
5795 "specValue", specValue);
5796
5797 Program *pProgram = nullptr;
5798 retVal = validateObjects(WithCastToInternal(program, &pProgram), specValue);
5799
5800 if (retVal == CL_SUCCESS) {
5801 retVal = pProgram->setProgramSpecializationConstant(specId, specSize, specValue);
5802 }
5803
5804 return retVal;
5805 }
5806
clGetKernelSuggestedLocalWorkSizeINTEL(cl_command_queue commandQueue,cl_kernel kernel,cl_uint workDim,const size_t * globalWorkOffset,const size_t * globalWorkSize,size_t * suggestedLocalWorkSize)5807 cl_int CL_API_CALL clGetKernelSuggestedLocalWorkSizeINTEL(cl_command_queue commandQueue,
5808 cl_kernel kernel,
5809 cl_uint workDim,
5810 const size_t *globalWorkOffset,
5811 const size_t *globalWorkSize,
5812 size_t *suggestedLocalWorkSize) {
5813 cl_int retVal = CL_SUCCESS;
5814 API_ENTER(&retVal);
5815 DBG_LOG_INPUTS("commandQueue", commandQueue, "cl_kernel", kernel,
5816 "globalWorkOffset[0]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 0),
5817 "globalWorkOffset[1]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 1),
5818 "globalWorkOffset[2]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 2),
5819 "globalWorkSize", NEO::FileLoggerInstance().getSizes(globalWorkSize, workDim, true),
5820 "suggestedLocalWorkSize", suggestedLocalWorkSize);
5821
5822 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
5823 CommandQueue *pCommandQueue = nullptr;
5824 retVal = validateObjects(WithCastToInternal(commandQueue, &pCommandQueue), WithCastToInternal(kernel, &pMultiDeviceKernel));
5825
5826 if (CL_SUCCESS != retVal) {
5827 return retVal;
5828 }
5829
5830 if ((workDim == 0) || (workDim > 3)) {
5831 retVal = CL_INVALID_WORK_DIMENSION;
5832 return retVal;
5833 }
5834
5835 if (globalWorkSize == nullptr ||
5836 globalWorkSize[0] == 0 ||
5837 (workDim > 1 && globalWorkSize[1] == 0) ||
5838 (workDim > 2 && globalWorkSize[2] == 0)) {
5839 retVal = CL_INVALID_GLOBAL_WORK_SIZE;
5840 return retVal;
5841 }
5842
5843 auto pKernel = pMultiDeviceKernel->getKernel(pCommandQueue->getDevice().getRootDeviceIndex());
5844 if (!pKernel->isPatched()) {
5845 retVal = CL_INVALID_KERNEL;
5846 return retVal;
5847 }
5848
5849 if (suggestedLocalWorkSize == nullptr) {
5850 retVal = CL_INVALID_VALUE;
5851 return retVal;
5852 }
5853
5854 pKernel->getSuggestedLocalWorkSize(workDim, globalWorkSize, globalWorkOffset, suggestedLocalWorkSize);
5855
5856 return retVal;
5857 }
5858
clGetKernelMaxConcurrentWorkGroupCountINTEL(cl_command_queue commandQueue,cl_kernel kernel,cl_uint workDim,const size_t * globalWorkOffset,const size_t * localWorkSize,size_t * suggestedWorkGroupCount)5859 cl_int CL_API_CALL clGetKernelMaxConcurrentWorkGroupCountINTEL(cl_command_queue commandQueue,
5860 cl_kernel kernel,
5861 cl_uint workDim,
5862 const size_t *globalWorkOffset,
5863 const size_t *localWorkSize,
5864 size_t *suggestedWorkGroupCount) {
5865
5866 cl_int retVal = CL_SUCCESS;
5867 API_ENTER(&retVal);
5868 DBG_LOG_INPUTS("commandQueue", commandQueue, "cl_kernel", kernel,
5869 "globalWorkOffset[0]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 0),
5870 "globalWorkOffset[1]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 1),
5871 "globalWorkOffset[2]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 2),
5872 "localWorkSize", NEO::FileLoggerInstance().getSizes(localWorkSize, workDim, true),
5873 "suggestedWorkGroupCount", suggestedWorkGroupCount);
5874
5875 CommandQueue *pCommandQueue = nullptr;
5876 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
5877
5878 retVal = validateObjects(WithCastToInternal(commandQueue, &pCommandQueue), WithCastToInternal(kernel, &pMultiDeviceKernel));
5879
5880 if (CL_SUCCESS != retVal) {
5881 return retVal;
5882 }
5883
5884 if ((workDim == 0) || (workDim > 3)) {
5885 retVal = CL_INVALID_WORK_DIMENSION;
5886 return retVal;
5887 }
5888
5889 if (globalWorkOffset == nullptr) {
5890 retVal = CL_INVALID_GLOBAL_OFFSET;
5891 return retVal;
5892 }
5893
5894 if (localWorkSize == nullptr) {
5895 retVal = CL_INVALID_WORK_GROUP_SIZE;
5896 return retVal;
5897 }
5898
5899 auto pKernel = pMultiDeviceKernel->getKernel(pCommandQueue->getDevice().getRootDeviceIndex());
5900 if (!pKernel->isPatched()) {
5901 retVal = CL_INVALID_KERNEL;
5902 return retVal;
5903 }
5904
5905 if (suggestedWorkGroupCount == nullptr) {
5906 retVal = CL_INVALID_VALUE;
5907 return retVal;
5908 }
5909
5910 WithCastToInternal(commandQueue, &pCommandQueue);
5911 *suggestedWorkGroupCount = pKernel->getMaxWorkGroupCount(workDim, localWorkSize, pCommandQueue);
5912
5913 return retVal;
5914 }
5915
clEnqueueNDCountKernelINTEL(cl_command_queue commandQueue,cl_kernel kernel,cl_uint workDim,const size_t * globalWorkOffset,const size_t * workgroupCount,const size_t * localWorkSize,cl_uint numEventsInWaitList,const cl_event * eventWaitList,cl_event * event)5916 cl_int CL_API_CALL clEnqueueNDCountKernelINTEL(cl_command_queue commandQueue,
5917 cl_kernel kernel,
5918 cl_uint workDim,
5919 const size_t *globalWorkOffset,
5920 const size_t *workgroupCount,
5921 const size_t *localWorkSize,
5922 cl_uint numEventsInWaitList,
5923 const cl_event *eventWaitList,
5924 cl_event *event) {
5925 cl_int retVal = CL_SUCCESS;
5926 API_ENTER(&retVal);
5927 DBG_LOG_INPUTS("commandQueue", commandQueue, "cl_kernel", kernel,
5928 "globalWorkOffset[0]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 0),
5929 "globalWorkOffset[1]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 1),
5930 "globalWorkOffset[2]", NEO::FileLoggerInstance().getInput(globalWorkOffset, 2),
5931 "workgroupCount", NEO::FileLoggerInstance().getSizes(workgroupCount, workDim, false),
5932 "localWorkSize", NEO::FileLoggerInstance().getSizes(localWorkSize, workDim, true),
5933 "numEventsInWaitList", numEventsInWaitList,
5934 "eventWaitList", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(eventWaitList), numEventsInWaitList),
5935 "event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1));
5936
5937 CommandQueue *pCommandQueue = nullptr;
5938 Kernel *pKernel = nullptr;
5939 MultiDeviceKernel *pMultiDeviceKernel = nullptr;
5940
5941 retVal = validateObjects(
5942 WithCastToInternal(commandQueue, &pCommandQueue),
5943 WithCastToInternal(kernel, &pMultiDeviceKernel),
5944 EventWaitList(numEventsInWaitList, eventWaitList));
5945
5946 if (CL_SUCCESS != retVal) {
5947 return retVal;
5948 }
5949
5950 auto &device = pCommandQueue->getClDevice();
5951 auto rootDeviceIndex = device.getRootDeviceIndex();
5952
5953 pKernel = pMultiDeviceKernel->getKernel(rootDeviceIndex);
5954 size_t globalWorkSize[3];
5955 for (size_t i = 0; i < workDim; i++) {
5956 globalWorkSize[i] = workgroupCount[i] * localWorkSize[i];
5957 }
5958
5959 if (pKernel->usesSyncBuffer()) {
5960 if (pKernel->getExecutionType() != KernelExecutionType::Concurrent) {
5961 retVal = CL_INVALID_KERNEL;
5962 return retVal;
5963 }
5964
5965 auto &hardwareInfo = device.getHardwareInfo();
5966 auto &hwHelper = HwHelper::get(hardwareInfo.platform.eRenderCoreFamily);
5967 auto engineGroupType = hwHelper.getEngineGroupType(pCommandQueue->getGpgpuEngine().getEngineType(),
5968 pCommandQueue->getGpgpuEngine().getEngineUsage(), hardwareInfo);
5969 if (!hwHelper.isCooperativeDispatchSupported(engineGroupType, hardwareInfo)) {
5970 retVal = CL_INVALID_COMMAND_QUEUE;
5971 return retVal;
5972 }
5973 }
5974
5975 if (pKernel->getExecutionType() == KernelExecutionType::Concurrent) {
5976 size_t requestedNumberOfWorkgroups = 1;
5977 for (size_t i = 0; i < workDim; i++) {
5978 requestedNumberOfWorkgroups *= workgroupCount[i];
5979 }
5980 size_t maximalNumberOfWorkgroupsAllowed = pKernel->getMaxWorkGroupCount(workDim, localWorkSize, pCommandQueue);
5981 if (requestedNumberOfWorkgroups > maximalNumberOfWorkgroupsAllowed) {
5982 retVal = CL_INVALID_VALUE;
5983 return retVal;
5984 }
5985 }
5986
5987 if (!pCommandQueue->validateCapabilityForOperation(CL_QUEUE_CAPABILITY_KERNEL_INTEL, numEventsInWaitList, eventWaitList, event)) {
5988 retVal = CL_INVALID_OPERATION;
5989 return retVal;
5990 }
5991
5992 if (pKernel->usesSyncBuffer()) {
5993 device.getDevice().allocateSyncBufferHandler();
5994 }
5995
5996 TakeOwnershipWrapper<MultiDeviceKernel> kernelOwnership(*pMultiDeviceKernel, gtpinIsGTPinInitialized());
5997 if (gtpinIsGTPinInitialized()) {
5998 gtpinNotifyKernelSubmit(kernel, pCommandQueue);
5999 }
6000
6001 retVal = pCommandQueue->enqueueKernel(
6002 pKernel,
6003 workDim,
6004 globalWorkOffset,
6005 globalWorkSize,
6006 localWorkSize,
6007 numEventsInWaitList,
6008 eventWaitList,
6009 event);
6010
6011 DBG_LOG_INPUTS("event", getClFileLogger().getEvents(reinterpret_cast<const uintptr_t *>(event), 1u));
6012 return retVal;
6013 }
6014
clSetContextDestructorCallback(cl_context context,void (CL_CALLBACK * pfnNotify)(cl_context,void *),void * userData)6015 cl_int CL_API_CALL clSetContextDestructorCallback(cl_context context,
6016 void(CL_CALLBACK *pfnNotify)(cl_context /* context */, void * /* user_data */),
6017 void *userData) {
6018 DBG_LOG_INPUTS("program", context,
6019 "pfnNotify", pfnNotify,
6020 "userData", userData);
6021
6022 cl_int retVal = CL_SUCCESS;
6023 API_ENTER(&retVal);
6024
6025 Context *pContext = nullptr;
6026 retVal = validateObjects(WithCastToInternal(context, &pContext),
6027 reinterpret_cast<void *>(pfnNotify));
6028
6029 if (retVal == CL_SUCCESS) {
6030 retVal = pContext->setDestructorCallback(pfnNotify, userData);
6031 }
6032
6033 return retVal;
6034 }
6035