diff --git a/sycl/plugins/level_zero/pi_level_zero.cpp b/sycl/plugins/level_zero/pi_level_zero.cpp index 2793c76798981..0db7be0174041 100644 --- a/sycl/plugins/level_zero/pi_level_zero.cpp +++ b/sycl/plugins/level_zero/pi_level_zero.cpp @@ -161,7 +161,7 @@ static const pi_uint32 MaxNumEventsPerPool = [] { // template ze_result_t zeHostSynchronizeImpl(Func Api, T Handle) { - if (!ZeDebug) { + if (!UrL0Debug) { return Api(Handle, UINT64_MAX); } @@ -259,7 +259,7 @@ _pi_context::getFreeSlotInExistingOrNewPool(ze_event_pool_handle_t &Pool, ZeEventPoolDesc.flags |= ZE_EVENT_POOL_FLAG_HOST_VISIBLE; if (ProfilingEnabled) ZeEventPoolDesc.flags |= ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP; - zePrint("ze_event_pool_desc_t flags set to: %d\n", ZeEventPoolDesc.flags); + urPrint("ze_event_pool_desc_t flags set to: %d\n", ZeEventPoolDesc.flags); std::vector ZeDevices; std::for_each(Devices.begin(), Devices.end(), [&](const pi_device &D) { @@ -801,10 +801,10 @@ static const zeCommandListBatchConfig ZeCommandListBatchConfig(bool IsCopy) { Val = std::stoi(BatchConfig.substr(Pos)); } catch (...) { if (IsCopy) - zePrint( + urPrint( "SYCL_PI_LEVEL_ZERO_COPY_BATCH_SIZE: failed to parse value\n"); else - zePrint("SYCL_PI_LEVEL_ZERO_BATCH_SIZE: failed to parse value\n"); + urPrint("SYCL_PI_LEVEL_ZERO_BATCH_SIZE: failed to parse value\n"); break; } switch (Ord) { @@ -827,11 +827,11 @@ static const zeCommandListBatchConfig ZeCommandListBatchConfig(bool IsCopy) { die("Unexpected batch config"); } if (IsCopy) - zePrint("SYCL_PI_LEVEL_ZERO_COPY_BATCH_SIZE: dynamic batch param " + urPrint("SYCL_PI_LEVEL_ZERO_COPY_BATCH_SIZE: dynamic batch param " "#%d: %d\n", (int)Ord, (int)Val); else - zePrint( + urPrint( "SYCL_PI_LEVEL_ZERO_BATCH_SIZE: dynamic batch param #%d: %d\n", (int)Ord, (int)Val); }; @@ -839,9 +839,9 @@ static const zeCommandListBatchConfig ZeCommandListBatchConfig(bool IsCopy) { } else { // Negative batch sizes are silently ignored. if (IsCopy) - zePrint("SYCL_PI_LEVEL_ZERO_COPY_BATCH_SIZE: ignored negative value\n"); + urPrint("SYCL_PI_LEVEL_ZERO_COPY_BATCH_SIZE: ignored negative value\n"); else - zePrint("SYCL_PI_LEVEL_ZERO_BATCH_SIZE: ignored negative value\n"); + urPrint("SYCL_PI_LEVEL_ZERO_BATCH_SIZE: ignored negative value\n"); } } return Config; @@ -1312,7 +1312,7 @@ void _pi_queue::adjustBatchSizeForFullBatch(bool IsCopy) { ZeCommandListBatchConfig.NumTimesClosedFullThreshold) { if (QueueBatchSize < ZeCommandListBatchConfig.DynamicSizeMax) { QueueBatchSize += ZeCommandListBatchConfig.DynamicSizeStep; - zePrint("Raising QueueBatchSize to %d\n", QueueBatchSize); + urPrint("Raising QueueBatchSize to %d\n", QueueBatchSize); } CommandBatch.NumTimesClosedEarly = 0; CommandBatch.NumTimesClosedFull = 0; @@ -1339,7 +1339,7 @@ void _pi_queue::adjustBatchSizeForPartialBatch(bool IsCopy) { QueueBatchSize = CommandBatch.OpenCommandList->second.size() - 1; if (QueueBatchSize < 1) QueueBatchSize = 1; - zePrint("Lowering QueueBatchSize to %d\n", QueueBatchSize); + urPrint("Lowering QueueBatchSize to %d\n", QueueBatchSize); CommandBatch.NumTimesClosedEarly = 0; CommandBatch.NumTimesClosedFull = 0; } @@ -1567,7 +1567,7 @@ pi_result _pi_queue::executeCommandList(pi_command_list_ptr_t CommandList, } // Check global control to make every command blocking for debugging. - if (IsBlocking || (ZeSerialize & ZeSerializeBlock) != 0) { + if (IsBlocking || (UrL0Serialize & UrL0SerializeBlock) != 0) { if (Device->ImmCommandListUsed) { synchronize(); } else { @@ -1581,7 +1581,7 @@ pi_result _pi_queue::executeCommandList(pi_command_list_ptr_t CommandList, bool _pi_queue::isBatchingAllowed(bool IsCopy) const { auto &CommandBatch = IsCopy ? CopyCommandBatch : ComputeCommandBatch; return (CommandBatch.QueueBatchSize > 0 && - ((ZeSerialize & ZeSerializeBlock) == 0)); + ((UrL0Serialize & UrL0SerializeBlock) == 0)); } // Return the index of the next queue to use based on a @@ -1663,7 +1663,7 @@ _pi_queue::pi_queue_group_t::getZeQueue(uint32_t *QueueGroupOrdinal) { ZeCommandQueueDesc.flags = ZE_COMMAND_QUEUE_FLAG_EXPLICIT_ONLY; } - zePrint("[getZeQueue]: create queue ordinal = %d, index = %d " + urPrint("[getZeQueue]: create queue ordinal = %d, index = %d " "(round robin in [%d, %d]) priority = %s\n", ZeCommandQueueDesc.ordinal, ZeCommandQueueDesc.index, LowerIndex, UpperIndex, Priority); @@ -1706,7 +1706,7 @@ pi_command_list_ptr_t &_pi_queue::pi_queue_group_t::getImmCmdList() { ZeCommandQueueDesc.flags = ZE_COMMAND_QUEUE_FLAG_EXPLICIT_ONLY; } - zePrint("[getZeQueue]: create queue ordinal = %d, index = %d " + urPrint("[getZeQueue]: create queue ordinal = %d, index = %d " "(round robin in [%d, %d]) priority = %s\n", ZeCommandQueueDesc.ordinal, ZeCommandQueueDesc.index, LowerIndex, UpperIndex, Priority); @@ -1997,13 +1997,13 @@ pi_result _pi_ze_event_list_t::createAndRetainPiZeEventList( } static void printZeEventList(const _pi_ze_event_list_t &PiZeEventList) { - zePrint(" NumEventsInWaitList %d:", PiZeEventList.Length); + urPrint(" NumEventsInWaitList %d:", PiZeEventList.Length); for (pi_uint32 I = 0; I < PiZeEventList.Length; I++) { - zePrint(" %#llx", pi_cast(PiZeEventList.ZeEventList[I])); + urPrint(" %#llx", pi_cast(PiZeEventList.ZeEventList[I])); } - zePrint("\n"); + urPrint("\n"); } pi_result _pi_ze_event_list_t::collectEventsForReleaseAndDestroyPiZeEventList( @@ -2065,9 +2065,9 @@ pi_result piPlatformsGet(pi_uint32 NumEntries, pi_platform *Platforms, pi_result piPlatformGetInfo(pi_platform Platform, pi_platform_info ParamName, size_t ParamValueSize, void *ParamValue, size_t *ParamValueSizeRet) { - zePrint("==========================\n"); - zePrint("SYCL over Level-Zero %s\n", Platform->ZeDriverVersion.c_str()); - zePrint("==========================\n"); + urPrint("==========================\n"); + urPrint("SYCL over Level-Zero %s\n", Platform->ZeDriverVersion.c_str()); + urPrint("==========================\n"); // To distinguish this L0 platform from Unified Runtime one. if (ParamName == PI_PLATFORM_INFO_NAME) { @@ -2639,7 +2639,7 @@ pi_result piQueueGetInfo(pi_queue Queue, pi_queue_info ParamName, return ReturnValue(pi_bool{true}); } default: - zePrint("Unsupported ParamName in piQueueGetInfo: ParamName=%d(0x%x)\n", + urPrint("Unsupported ParamName in piQueueGetInfo: ParamName=%d(0x%x)\n", ParamName, ParamName); return PI_ERROR_INVALID_VALUE; } @@ -2737,11 +2737,11 @@ static pi_result piQueueReleaseInternal(pi_queue Queue) { ZE_CALL(zeCommandQueueDestroy, (ZeQueue)); } - zePrint("piQueueRelease(compute) NumTimesClosedFull %d, " + urPrint("piQueueRelease(compute) NumTimesClosedFull %d, " "NumTimesClosedEarly %d\n", Queue->ComputeCommandBatch.NumTimesClosedFull, Queue->ComputeCommandBatch.NumTimesClosedEarly); - zePrint("piQueueRelease(copy) NumTimesClosedFull %d, NumTimesClosedEarly " + urPrint("piQueueRelease(copy) NumTimesClosedFull %d, NumTimesClosedEarly " "%d\n", Queue->CopyCommandBatch.NumTimesClosedFull, Queue->CopyCommandBatch.NumTimesClosedEarly); @@ -3183,7 +3183,7 @@ pi_result piMemImageCreate(pi_context Context, pi_mem_flags Flags, ZeImageFormatTypeSize = 8; break; default: - zePrint("piMemImageCreate: unsupported image data type: data type = %d\n", + urPrint("piMemImageCreate: unsupported image data type: data type = %d\n", ImageFormat->image_channel_data_type); return PI_ERROR_INVALID_VALUE; } @@ -3203,12 +3203,12 @@ pi_result piMemImageCreate(pi_context Context, pi_mem_flags Flags, ZeImageFormatLayout = ZE_IMAGE_FORMAT_LAYOUT_32_32_32_32; break; default: - zePrint("piMemImageCreate: unexpected data type Size\n"); + urPrint("piMemImageCreate: unexpected data type Size\n"); return PI_ERROR_INVALID_VALUE; } break; default: - zePrint("format layout = %d\n", ImageFormat->image_channel_order); + urPrint("format layout = %d\n", ImageFormat->image_channel_order); die("piMemImageCreate: unsupported image format layout\n"); break; } @@ -3237,7 +3237,7 @@ pi_result piMemImageCreate(pi_context Context, pi_mem_flags Flags, ZeImageType = ZE_IMAGE_TYPE_2DARRAY; break; default: - zePrint("piMemImageCreate: unsupported image type\n"); + urPrint("piMemImageCreate: unsupported image type\n"); return PI_ERROR_INVALID_VALUE; } @@ -3428,7 +3428,7 @@ pi_result piProgramCreateWithBinary( // For now we support only one device. if (NumDevices != 1) { - zePrint("piProgramCreateWithBinary: level_zero supports only one device."); + urPrint("piProgramCreateWithBinary: level_zero supports only one device."); return PI_ERROR_INVALID_VALUE; } if (!Binaries[0] || !Lengths[0]) { @@ -3474,7 +3474,7 @@ pi_result piclProgramCreateWithSource(pi_context Context, pi_uint32 Count, (void)Strings; (void)Lengths; (void)RetProgram; - zePrint("piclProgramCreateWithSource: not supported in Level Zero\n"); + urPrint("piclProgramCreateWithSource: not supported in Level Zero\n"); return PI_ERROR_INVALID_OPERATION; } @@ -3590,7 +3590,7 @@ pi_result piProgramLink(pi_context Context, pi_uint32 NumDevices, void *UserData, pi_program *RetProgram) { // We only support one device with Level Zero currently. if (NumDevices != 1) { - zePrint("piProgramLink: level_zero supports only one device."); + urPrint("piProgramLink: level_zero supports only one device."); return PI_ERROR_INVALID_VALUE; } @@ -3702,7 +3702,7 @@ pi_result piProgramLink(pi_context Context, pi_uint32 NumDevices, ZeModuleDesc.pBuildFlags = ZeExtModuleDesc.pBuildFlags[0]; ZeModuleDesc.pConstants = ZeExtModuleDesc.pConstants[0]; } else { - zePrint("piProgramLink: level_zero driver does not have static linking " + urPrint("piProgramLink: level_zero driver does not have static linking " "support."); return PI_ERROR_INVALID_VALUE; } @@ -3805,7 +3805,7 @@ pi_result piProgramBuild(pi_program Program, pi_uint32 NumDevices, // TODO: we should eventually build to the possibly multiple root // devices in the context. if (NumDevices != 1) { - zePrint("piProgramBuild: level_zero supports only one device."); + urPrint("piProgramBuild: level_zero supports only one device."); return PI_ERROR_INVALID_VALUE; } @@ -3931,7 +3931,7 @@ pi_result piProgramGetBuildInfo(pi_program Program, pi_device Device, // program. return ReturnValue(""); } else { - zePrint("piProgramGetBuildInfo: unsupported ParamName\n"); + urPrint("piProgramGetBuildInfo: unsupported ParamName\n"); return PI_ERROR_INVALID_VALUE; } return PI_SUCCESS; @@ -4214,7 +4214,7 @@ pi_result piKernelGetInfo(pi_kernel Kernel, pi_kernel_info ParamName, return PI_ERROR_UNKNOWN; } default: - zePrint("Unsupported ParamName in piKernelGetInfo: ParamName=%d(0x%x)\n", + urPrint("Unsupported ParamName in piKernelGetInfo: ParamName=%d(0x%x)\n", ParamName, ParamName); return PI_ERROR_INVALID_VALUE; } @@ -4269,7 +4269,7 @@ pi_result piKernelGetGroupInfo(pi_kernel Kernel, pi_device Device, break; } default: - zePrint("Unknown ParamName in piKernelGetGroupInfo: ParamName=%d(0x%x)\n", + urPrint("Unknown ParamName in piKernelGetGroupInfo: ParamName=%d(0x%x)\n", ParamName, ParamName); return PI_ERROR_INVALID_VALUE; } @@ -4345,7 +4345,7 @@ piEnqueueKernelLaunch(pi_queue Queue, pi_kernel Kernel, pi_uint32 WorkDim, Queue->Mutex, Kernel->Mutex, Kernel->Program->Mutex); if (GlobalWorkOffset != NULL) { if (!Queue->Device->Platform->ZeDriverGlobalOffsetExtensionFound) { - zePrint("No global offset extension found on this driver\n"); + urPrint("No global offset extension found on this driver\n"); return PI_ERROR_INVALID_VALUE; } @@ -4406,13 +4406,13 @@ piEnqueueKernelLaunch(pi_queue Queue, pi_kernel Kernel, pi_uint32 WorkDim, --GroupSize[I]; } if (GlobalWorkSize[I] / GroupSize[I] > UINT32_MAX) { - zePrint("piEnqueueKernelLaunch: can't find a WG size " + urPrint("piEnqueueKernelLaunch: can't find a WG size " "suitable for global work size > UINT32_MAX\n"); return PI_ERROR_INVALID_WORK_GROUP_SIZE; } WG[I] = GroupSize[I]; } - zePrint("piEnqueueKernelLaunch: using computed WG size = {%d, %d, %d}\n", + urPrint("piEnqueueKernelLaunch: using computed WG size = {%d, %d, %d}\n", WG[0], WG[1], WG[2]); } } @@ -4441,26 +4441,26 @@ piEnqueueKernelLaunch(pi_queue Queue, pi_kernel Kernel, pi_uint32 WorkDim, break; default: - zePrint("piEnqueueKernelLaunch: unsupported work_dim\n"); + urPrint("piEnqueueKernelLaunch: unsupported work_dim\n"); return PI_ERROR_INVALID_VALUE; } // Error handling for non-uniform group size case if (GlobalWorkSize[0] != size_t(ZeThreadGroupDimensions.groupCountX) * WG[0]) { - zePrint("piEnqueueKernelLaunch: invalid work_dim. The range is not a " + urPrint("piEnqueueKernelLaunch: invalid work_dim. The range is not a " "multiple of the group size in the 1st dimension\n"); return PI_ERROR_INVALID_WORK_GROUP_SIZE; } if (GlobalWorkSize[1] != size_t(ZeThreadGroupDimensions.groupCountY) * WG[1]) { - zePrint("piEnqueueKernelLaunch: invalid work_dim. The range is not a " + urPrint("piEnqueueKernelLaunch: invalid work_dim. The range is not a " "multiple of the group size in the 2nd dimension\n"); return PI_ERROR_INVALID_WORK_GROUP_SIZE; } if (GlobalWorkSize[2] != size_t(ZeThreadGroupDimensions.groupCountZ) * WG[2]) { - zePrint("piEnqueueKernelLaunch: invalid work_dim. The range is not a " + urPrint("piEnqueueKernelLaunch: invalid work_dim. The range is not a " "multiple of the group size in the 3rd dimension\n"); return PI_ERROR_INVALID_WORK_GROUP_SIZE; } @@ -4534,7 +4534,7 @@ piEnqueueKernelLaunch(pi_queue Queue, pi_kernel Kernel, pi_uint32 WorkDim, (*Event)->WaitList.ZeEventList)); } - zePrint("calling zeCommandListAppendLaunchKernel() with" + urPrint("calling zeCommandListAppendLaunchKernel() with" " ZeEvent %#llx\n", pi_cast(ZeEvent)); printZeEventList((*Event)->WaitList); @@ -4796,7 +4796,7 @@ pi_result piEventGetInfo(pi_event Event, pi_event_info ParamName, case PI_EVENT_INFO_REFERENCE_COUNT: return ReturnValue(pi_uint32{Event->RefCount.load()}); default: - zePrint("Unsupported ParamName in piEventGetInfo: ParamName=%d(%x)\n", + urPrint("Unsupported ParamName in piEventGetInfo: ParamName=%d(%x)\n", ParamName, ParamName); return PI_ERROR_INVALID_VALUE; } @@ -4861,7 +4861,7 @@ pi_result piEventGetProfilingInfo(pi_event Event, pi_profiling_info ParamName, // before submitting command to device return ReturnValue(uint64_t{0}); default: - zePrint("piEventGetProfilingInfo: not supported ParamName\n"); + urPrint("piEventGetProfilingInfo: not supported ParamName\n"); return PI_ERROR_INVALID_VALUE; } @@ -5045,7 +5045,7 @@ pi_result piEventsWait(pi_uint32 NumEvents, const pi_event *EventList) { die("The host-visible proxy event missing"); ze_event_handle_t ZeEvent = HostVisibleEvent->ZeEvent; - zePrint("ZeEvent = %#llx\n", pi_cast(ZeEvent)); + urPrint("ZeEvent = %#llx\n", pi_cast(ZeEvent)); ZE_CALL(zeHostSynchronize, (ZeEvent)); EventList[I]->Completed = true; } @@ -5274,7 +5274,7 @@ pi_result piSamplerCreate(pi_context Context, else if (CurValueBool == PI_FALSE) ZeSamplerDesc.isNormalized = PI_FALSE; else { - zePrint("piSamplerCreate: unsupported " + urPrint("piSamplerCreate: unsupported " "PI_SAMPLER_NORMALIZED_COORDS value\n"); return PI_ERROR_INVALID_VALUE; } @@ -5316,9 +5316,9 @@ pi_result piSamplerCreate(pi_context Context, ZeSamplerDesc.addressMode = ZE_SAMPLER_ADDRESS_MODE_MIRROR; break; default: - zePrint("piSamplerCreate: unsupported PI_SAMPLER_ADDRESSING_MODE " + urPrint("piSamplerCreate: unsupported PI_SAMPLER_ADDRESSING_MODE " "value\n"); - zePrint("PI_SAMPLER_ADDRESSING_MODE=%d\n", CurValueAddressingMode); + urPrint("PI_SAMPLER_ADDRESSING_MODE=%d\n", CurValueAddressingMode); return PI_ERROR_INVALID_VALUE; } } break; @@ -5333,8 +5333,8 @@ pi_result piSamplerCreate(pi_context Context, else if (CurValueFilterMode == PI_SAMPLER_FILTER_MODE_LINEAR) ZeSamplerDesc.filterMode = ZE_SAMPLER_FILTER_MODE_LINEAR; else { - zePrint("PI_SAMPLER_FILTER_MODE=%d\n", CurValueFilterMode); - zePrint( + urPrint("PI_SAMPLER_FILTER_MODE=%d\n", CurValueFilterMode); + urPrint( "piSamplerCreate: unsupported PI_SAMPLER_FILTER_MODE value\n"); return PI_ERROR_INVALID_VALUE; } @@ -5822,7 +5822,7 @@ enqueueMemCopyHelper(pi_command_type CommandType, pi_queue Queue, void *Dst, const auto &ZeCommandList = CommandList->first; const auto &WaitList = (*Event)->WaitList; - zePrint("calling zeCommandListAppendMemoryCopy() with\n" + urPrint("calling zeCommandListAppendMemoryCopy() with\n" " ZeEvent %#llx\n", pi_cast(ZeEvent)); printZeEventList(WaitList); @@ -5881,7 +5881,7 @@ static pi_result enqueueMemCopyRectHelper( const auto &ZeCommandList = CommandList->first; const auto &WaitList = (*Event)->WaitList; - zePrint("calling zeCommandListAppendMemoryCopy() with\n" + urPrint("calling zeCommandListAppendMemoryCopy() with\n" " ZeEvent %#llx\n", pi_cast(ZeEvent)); printZeEventList(WaitList); @@ -5922,11 +5922,11 @@ static pi_result enqueueMemCopyRectHelper( SrcBuffer, &ZeSrcRegion, SrcPitch, SrcSlicePitch, nullptr, WaitList.Length, WaitList.ZeEventList)); - zePrint("calling zeCommandListAppendMemoryCopyRegion()\n"); + urPrint("calling zeCommandListAppendMemoryCopyRegion()\n"); ZE_CALL(zeCommandListAppendBarrier, (ZeCommandList, ZeEvent, 0, nullptr)); - zePrint("calling zeCommandListAppendBarrier() with Event %#llx\n", + urPrint("calling zeCommandListAppendBarrier() with Event %#llx\n", pi_cast(ZeEvent)); if (auto Res = Queue->executeCommandList(CommandList, Blocking, OkToBatch)) @@ -6139,7 +6139,7 @@ enqueueMemFillHelper(pi_command_type CommandType, pi_queue Queue, void *Ptr, (ZeCommandList, Ptr, Pattern, PatternSize, Size, ZeEvent, WaitList.Length, WaitList.ZeEventList)); - zePrint("calling zeCommandListAppendMemoryFill() with\n" + urPrint("calling zeCommandListAppendMemoryFill() with\n" " ZeEvent %#llx\n", pi_cast(ZeEvent)); printZeEventList(WaitList); @@ -6273,7 +6273,7 @@ pi_result piEnqueueMemBufferMap(pi_queue Queue, pi_mem Mem, pi_bool BlockingMap, // False as the second value in pair means that mapping was not inserted // because mapping already exists. if (!Res.second) { - zePrint("piEnqueueMemBufferMap: duplicate mapping detected\n"); + urPrint("piEnqueueMemBufferMap: duplicate mapping detected\n"); return PI_ERROR_INVALID_VALUE; } @@ -6329,7 +6329,7 @@ pi_result piEnqueueMemBufferMap(pi_queue Queue, pi_mem Mem, pi_bool BlockingMap, // False as the second value in pair means that mapping was not inserted // because mapping already exists. if (!Res.second) { - zePrint("piEnqueueMemBufferMap: duplicate mapping detected\n"); + urPrint("piEnqueueMemBufferMap: duplicate mapping detected\n"); return PI_ERROR_INVALID_VALUE; } return PI_SUCCESS; @@ -6374,7 +6374,7 @@ pi_result piEnqueueMemUnmap(pi_queue Queue, pi_mem Mem, void *MappedPtr, std::scoped_lock Guard(Buffer->Mutex); auto It = Buffer->Mappings.find(MappedPtr); if (It == Buffer->Mappings.end()) { - zePrint("piEnqueueMemUnmap: unknown memory mapping\n"); + urPrint("piEnqueueMemUnmap: unknown memory mapping\n"); return PI_ERROR_INVALID_VALUE; } MapInfo = It->second; @@ -6645,7 +6645,7 @@ static pi_result enqueueMemImageCommandHelper( pi_cast(ZeHandleSrc), &ZeDstRegion, &ZeSrcRegion, ZeEvent, 0, nullptr)); } else { - zePrint("enqueueMemImageUpdate: unsupported image command type\n"); + urPrint("enqueueMemImageUpdate: unsupported image command type\n"); return PI_ERROR_INVALID_OPERATION; } @@ -7841,7 +7841,7 @@ pi_result piextUSMGetMemAllocInfo(pi_context Context, const void *Ptr, MemAllocaType = PI_MEM_TYPE_SHARED; break; default: - zePrint("piextUSMGetMemAllocInfo: unexpected usm memory type\n"); + urPrint("piextUSMGetMemAllocInfo: unexpected usm memory type\n"); return PI_ERROR_INVALID_VALUE; } return ReturnValue(MemAllocaType); @@ -7865,7 +7865,7 @@ pi_result piextUSMGetMemAllocInfo(pi_context Context, const void *Ptr, return ReturnValue(Size); } default: - zePrint("piextUSMGetMemAllocInfo: unsupported ParamName\n"); + urPrint("piextUSMGetMemAllocInfo: unsupported ParamName\n"); return PI_ERROR_INVALID_VALUE; } return PI_SUCCESS; @@ -8061,7 +8061,7 @@ pi_result piKernelSetExecInfo(pi_kernel Kernel, pi_kernel_exec_info ParamName, } ZE_CALL(zeKernelSetCacheConfig, (Kernel->ZeKernel, ZeCacheConfig);); } else { - zePrint("piKernelSetExecInfo: unsupported ParamName\n"); + urPrint("piKernelSetExecInfo: unsupported ParamName\n"); return PI_ERROR_INVALID_VALUE; } @@ -8135,7 +8135,7 @@ pi_result piTearDown(void *PluginParameter) { // Print the balance of various create/destroy native calls. // The idea is to verify if the number of create(+) and destroy(-) calls are // matched. - if (ZeDebug & ZE_DEBUG_CALL_COUNT) { + if (UrL0Debug & UR_L0_DEBUG_CALL_COUNT) { // clang-format off // // The format of this table is such that each row accounts for a @@ -8177,7 +8177,7 @@ pi_result piTearDown(void *PluginParameter) { // clang-format on fprintf(stderr, "ZE_DEBUG=%d: check balance of create/destroy calls\n", - ZE_DEBUG_CALL_COUNT); + UR_L0_DEBUG_CALL_COUNT); fprintf(stderr, "----------------------------------------------------------\n"); for (const auto &Row : CreateDestroySet) { @@ -8447,7 +8447,7 @@ pi_result _pi_buffer::getZeHandle(char *&ZeHandle, access_mode_t AccessMode, } } - zePrint("getZeHandle(pi_device{%p}) = %p\n", (void *)Device, + urPrint("getZeHandle(pi_device{%p}) = %p\n", (void *)Device, (void *)Allocation.ZeHandle); return PI_SUCCESS; } diff --git a/sycl/plugins/unified_runtime/CMakeLists.txt b/sycl/plugins/unified_runtime/CMakeLists.txt index 76b3916549a38..ab7eaab4b1a06 100755 --- a/sycl/plugins/unified_runtime/CMakeLists.txt +++ b/sycl/plugins/unified_runtime/CMakeLists.txt @@ -4,7 +4,7 @@ if (NOT DEFINED UNIFIED_RUNTIME_LIBRARY OR NOT DEFINED UNIFIED_RUNTIME_INCLUDE_D include(FetchContent) set(UNIFIED_RUNTIME_REPO "https://github.com/oneapi-src/unified-runtime.git") - set(UNIFIED_RUNTIME_TAG 6fb8e2620f1474428a539ef46d6dd47043c7d59b) + set(UNIFIED_RUNTIME_TAG d6af758779db6eebdc419fd5e249302f566eb5de) message(STATUS "Will fetch Unified Runtime from ${UNIFIED_RUNTIME_REPO}") FetchContent_Declare(unified-runtime diff --git a/sycl/plugins/unified_runtime/pi2ur.hpp b/sycl/plugins/unified_runtime/pi2ur.hpp index ac6b9ff0466a4..1b34d62ed7a70 100644 --- a/sycl/plugins/unified_runtime/pi2ur.hpp +++ b/sycl/plugins/unified_runtime/pi2ur.hpp @@ -29,7 +29,7 @@ static pi_result ur2piResult(ur_result_t urResult) { {UR_RESULT_ERROR_INVALID_FUNCTION_NAME, PI_ERROR_BUILD_PROGRAM_FAILURE}, {UR_RESULT_ERROR_INVALID_WORK_GROUP_SIZE, PI_ERROR_INVALID_WORK_GROUP_SIZE}, - {UR_RESULT_ERROR_MODULE_BUILD_FAILURE, PI_ERROR_BUILD_PROGRAM_FAILURE}, + {UR_RESULT_ERROR_PROGRAM_BUILD_FAILURE, PI_ERROR_BUILD_PROGRAM_FAILURE}, {UR_RESULT_ERROR_OUT_OF_DEVICE_MEMORY, PI_ERROR_OUT_OF_RESOURCES}, {UR_RESULT_ERROR_OUT_OF_HOST_MEMORY, PI_ERROR_OUT_OF_HOST_MEMORY}}; diff --git a/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero.cpp b/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero.cpp index 340d1022387f1..182be4cf149e8 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero.cpp @@ -18,32 +18,6 @@ std::mutex ZeCall::GlobalLock; ZeUSMImportExtension ZeUSMImport; -void zePrint(const char *Format, ...) { - if (ZeDebug & ZE_DEBUG_BASIC) { - va_list Args; - va_start(Args, Format); - vfprintf(stderr, Format, Args); - va_end(Args); - } -} - -// This function will ensure compatibility with both Linux and Windows for -// setting environment variables. -bool setEnvVar(const char *name, const char *value) { -#ifdef _WIN32 - int Res = _putenv_s(name, value); -#else - int Res = setenv(name, value, 1); -#endif - if (Res != 0) { - zePrint( - "Level Zero plugin was unable to set the environment variable: %s\n", - name); - return false; - } - return true; -} - // Trace a call to Level-Zero RT #define ZE_CALL(ZeName, ZeArgs) \ { \ @@ -52,182 +26,6 @@ bool setEnvVar(const char *name, const char *value) { return ze2urResult(Result); \ } -// This will count the calls to Level-Zero -std::map *ZeCallCount = nullptr; - -inline void zeParseError(ze_result_t ZeError, const char *&ErrorString) { - switch (ZeError) { -#define ZE_ERRCASE(ERR) \ - case ERR: \ - ErrorString = "" #ERR; \ - break; - - ZE_ERRCASE(ZE_RESULT_SUCCESS) - ZE_ERRCASE(ZE_RESULT_NOT_READY) - ZE_ERRCASE(ZE_RESULT_ERROR_DEVICE_LOST) - ZE_ERRCASE(ZE_RESULT_ERROR_OUT_OF_HOST_MEMORY) - ZE_ERRCASE(ZE_RESULT_ERROR_OUT_OF_DEVICE_MEMORY) - ZE_ERRCASE(ZE_RESULT_ERROR_MODULE_BUILD_FAILURE) - ZE_ERRCASE(ZE_RESULT_ERROR_INSUFFICIENT_PERMISSIONS) - ZE_ERRCASE(ZE_RESULT_ERROR_NOT_AVAILABLE) - ZE_ERRCASE(ZE_RESULT_ERROR_UNINITIALIZED) - ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_VERSION) - ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_FEATURE) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_ARGUMENT) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_NULL_HANDLE) - ZE_ERRCASE(ZE_RESULT_ERROR_HANDLE_OBJECT_IN_USE) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_NULL_POINTER) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_SIZE) - ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_SIZE) - ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_ALIGNMENT) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_SYNCHRONIZATION_OBJECT) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_ENUMERATION) - ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_ENUMERATION) - ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_IMAGE_FORMAT) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_NATIVE_BINARY) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_GLOBAL_NAME) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_KERNEL_NAME) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_FUNCTION_NAME) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_GROUP_SIZE_DIMENSION) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_GLOBAL_WIDTH_DIMENSION) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_INDEX) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_SIZE) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_KERNEL_ATTRIBUTE_VALUE) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_COMMAND_LIST_TYPE) - ZE_ERRCASE(ZE_RESULT_ERROR_OVERLAPPING_REGIONS) - ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_MODULE_UNLINKED) - ZE_ERRCASE(ZE_RESULT_ERROR_UNKNOWN) - -#undef ZE_ERRCASE - default: - assert(false && "Unexpected Error code"); - } // switch -} - -ze_result_t ZeCall::doCall(ze_result_t ZeResult, const char *ZeName, - const char *ZeArgs, bool TraceError) { - zePrint("ZE ---> %s%s\n", ZeName, ZeArgs); - - if (ZeDebug & ZE_DEBUG_CALL_COUNT) { - ++(*ZeCallCount)[ZeName]; - } - - if (ZeResult && TraceError) { - const char *ErrorString = "Unknown"; - zeParseError(ZeResult, ErrorString); - zePrint("Error (%s) in %s\n", ErrorString, ZeName); - } - return ZeResult; -} - -// Specializations for various L0 structures -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_EVENT_POOL_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_FENCE_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_COMMAND_LIST_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_CONTEXT_DESC; -} -template <> -ze_structure_type_t -getZeStructureType() { - return ZE_STRUCTURE_TYPE_RELAXED_ALLOCATION_LIMITS_EXP_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_HOST_MEM_ALLOC_DESC; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DEVICE_MEM_ALLOC_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_COMMAND_QUEUE_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_IMAGE_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_MODULE_DESC; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_MODULE_PROGRAM_EXP_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_KERNEL_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_EVENT_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_SAMPLER_DESC; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DRIVER_PROPERTIES; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DEVICE_PROPERTIES; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DEVICE_COMPUTE_PROPERTIES; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_COMMAND_QUEUE_GROUP_PROPERTIES; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DEVICE_IMAGE_PROPERTIES; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DEVICE_MODULE_PROPERTIES; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DEVICE_CACHE_PROPERTIES; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_PROPERTIES; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_EXT_PROPERTIES; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_ACCESS_PROPERTIES; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_MODULE_PROPERTIES; -} -template <> ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_KERNEL_PROPERTIES; -} -template <> -ze_structure_type_t getZeStructureType() { - return ZE_STRUCTURE_TYPE_MEMORY_ALLOCATION_PROPERTIES; -} - -template <> zes_structure_type_t getZesStructureType() { - return ZES_STRUCTURE_TYPE_PCI_PROPERTIES; -} - -template <> zes_structure_type_t getZesStructureType() { - return ZES_STRUCTURE_TYPE_MEM_STATE; -} - -template <> zes_structure_type_t getZesStructureType() { - return ZES_STRUCTURE_TYPE_MEM_PROPERTIES; -} - static const bool ExposeCSliceInAffinityPartitioning = [] { const char *Flag = std::getenv("SYCL_PI_LEVEL_ZERO_EXPOSE_CSLICE_IN_AFFINITY_PARTITIONING"); @@ -303,7 +101,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urPlatformGet( static std::once_flag ZeCallCountInitialized; try { std::call_once(ZeCallCountInitialized, []() { - if (ZeDebug & ZE_DEBUG_CALL_COUNT) { + if (UrL0Debug & UR_L0_DEBUG_CALL_COUNT) { ZeCallCount = new std::map; } }); @@ -315,7 +113,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urPlatformGet( // Setting these environment variables before running zeInit will enable the // validation layer in the Level Zero loader. - if (ZeDebug & ZE_DEBUG_VALIDATION) { + if (UrL0Debug & UR_L0_DEBUG_VALIDATION) { setEnvVar("ZE_ENABLE_VALIDATION_LAYER", "1"); setEnvVar("ZE_ENABLE_PARAMETER_VALIDATION", "1"); } @@ -343,7 +141,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urPlatformGet( } if (ZeResult != ZE_RESULT_SUCCESS) { - zePrint("zeInit: Level Zero initialization failure\n"); + urPrint("zeInit: Level Zero initialization failure\n"); return ze2urResult(ZeResult); } @@ -450,7 +248,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urPlatformGetInfo( // return ReturnValue(Platform->ZeDriverApiVersion.c_str()); default: - zePrint("piPlatformGetInfo: unrecognized ParamName\n"); + urPrint("piPlatformGetInfo: unrecognized ParamName\n"); return UR_RESULT_ERROR_INVALID_VALUE; } @@ -511,7 +309,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGet( break; default: Matched = false; - zePrint("Unknown device type"); + urPrint("Unknown device type"); break; } if (Matched) @@ -561,7 +359,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetInfo( case ZE_DEVICE_TYPE_FPGA: return ReturnValue(UR_DEVICE_TYPE_FPGA); default: - zePrint("This device type is not supported\n"); + urPrint("This device type is not supported\n"); return UR_RESULT_ERROR_INVALID_VALUE; } } @@ -1058,7 +856,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetInfo( return ReturnValue(uint32_t{Device->ZeDeviceProperties->deviceId}); case UR_DEVICE_INFO_PCI_ADDRESS: { if (getenv("ZES_ENABLE_SYSMAN") == nullptr) { - zePrint("Set SYCL_ENABLE_PCI=1 to obtain PCI data.\n"); + urPrint("Set SYCL_ENABLE_PCI=1 to obtain PCI data.\n"); return UR_RESULT_ERROR_INVALID_VALUE; } ZesStruct ZeDevicePciProperties; @@ -1222,8 +1020,8 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetInfo( // TODO: Implement. default: - zePrint("Unsupported ParamName in piGetDeviceInfo\n"); - zePrint("ParamName=%d(0x%x)\n", ParamName, ParamName); + urPrint("Unsupported ParamName in piGetDeviceInfo\n"); + urPrint("ParamName=%d(0x%x)\n", ParamName, ParamName); return UR_RESULT_ERROR_INVALID_VALUE; } @@ -1263,7 +1061,7 @@ getRangeOfAllowedCopyEngines(const ur_device_handle_t &Device) { int UpperCopyEngineIndex = std::stoi(CopyEngineRange.substr(pos + 1)); if ((LowerCopyEngineIndex > UpperCopyEngineIndex) || (LowerCopyEngineIndex < -1) || (UpperCopyEngineIndex < -1)) { - zePrint("SYCL_PI_LEVEL_ZERO_USE_COPY_ENGINE: invalid value provided, " + urPrint("SYCL_PI_LEVEL_ZERO_USE_COPY_ENGINE: invalid value provided, " "default set.\n"); LowerCopyEngineIndex = 0; UpperCopyEngineIndex = INT_MAX; @@ -1340,7 +1138,7 @@ ur_result_t _ur_device_handle_t::initialize(int SubSubDeviceOrdinal, if (numQueueGroups == 0) { return UR_RESULT_ERROR_UNKNOWN; } - zePrint("NOTE: Number of queue groups = %d\n", numQueueGroups); + urPrint("NOTE: Number of queue groups = %d\n", numQueueGroups); std::vector> QueueGroupProperties(numQueueGroups); ZE_CALL(zeDeviceGetCommandQueueGroupProperties, @@ -1393,14 +1191,14 @@ ur_result_t _ur_device_handle_t::initialize(int SubSubDeviceOrdinal, } } if (QueueGroup[queue_group_info_t::MainCopy].ZeOrdinal < 0) - zePrint("NOTE: main blitter/copy engine is not available\n"); + urPrint("NOTE: main blitter/copy engine is not available\n"); else - zePrint("NOTE: main blitter/copy engine is available\n"); + urPrint("NOTE: main blitter/copy engine is available\n"); if (QueueGroup[queue_group_info_t::LinkCopy].ZeOrdinal < 0) - zePrint("NOTE: link blitter/copy engines are not available\n"); + urPrint("NOTE: link blitter/copy engines are not available\n"); else - zePrint("NOTE: link blitter/copy engines are available\n"); + urPrint("NOTE: link blitter/copy engines are available\n"); } } diff --git a/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero.hpp b/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero.hpp index 94e532f2ef6da..10d018a29506b 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero.hpp +++ b/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero.hpp @@ -22,91 +22,6 @@ #include "ur_level_zero_common.hpp" -// Returns the ze_structure_type_t to use in .stype of a structured descriptor. -// Intentionally not defined; will give an error if no proper specialization -template ze_structure_type_t getZeStructureType(); -template zes_structure_type_t getZesStructureType(); - -// The helpers to properly default initialize Level-Zero descriptor and -// properties structures. -template struct ZeStruct : public T { - ZeStruct() : T{} { // zero initializes base struct - this->stype = getZeStructureType(); - this->pNext = nullptr; - } -}; - -template struct ZesStruct : public T { - ZesStruct() : T{} { // zero initializes base struct - this->stype = getZesStructureType(); - this->pNext = nullptr; - } -}; - -// Controls Level Zero calls serialization to w/a Level Zero driver being not MT -// ready. Recognized values (can be used as a bit mask): -enum { - ZeSerializeNone = - 0, // no locking or blocking (except when SYCL RT requested blocking) - ZeSerializeLock = 1, // locking around each ZE_CALL - ZeSerializeBlock = - 2, // blocking ZE calls, where supported (usually in enqueue commands) -}; -static const uint32_t ZeSerialize = [] { - const char *SerializeMode = std::getenv("ZE_SERIALIZE"); - const uint32_t SerializeModeValue = - SerializeMode ? std::atoi(SerializeMode) : 0; - return SerializeModeValue; -}(); - -// This class encapsulates actions taken along with a call to Level Zero API. -class ZeCall { -private: - // The global mutex that is used for total serialization of Level Zero calls. - static std::mutex GlobalLock; - -public: - ZeCall() { - if ((ZeSerialize & ZeSerializeLock) != 0) { - GlobalLock.lock(); - } - } - ~ZeCall() { - if ((ZeSerialize & ZeSerializeLock) != 0) { - GlobalLock.unlock(); - } - } - - // The non-static version just calls static one. - ze_result_t doCall(ze_result_t ZeResult, const char *ZeName, - const char *ZeArgs, bool TraceError = true); -}; - -// Controls Level Zero calls tracing. -enum DebugLevel { - ZE_DEBUG_NONE = 0x0, - ZE_DEBUG_BASIC = 0x1, - ZE_DEBUG_VALIDATION = 0x2, - ZE_DEBUG_CALL_COUNT = 0x4, - ZE_DEBUG_ALL = -1 -}; - -const int ZeDebug = [] { - const char *DebugMode = std::getenv("ZE_DEBUG"); - return DebugMode ? std::atoi(DebugMode) : ZE_DEBUG_NONE; -}(); - -// Prints to stderr if ZE_DEBUG allows it -void zePrint(const char *Format, ...); - -// This function will ensure compatibility with both Linux and Windows for -// setting environment variables. -bool setEnvVar(const char *name, const char *value); - -// Perform traced call to L0 without checking for errors -#define ZE_CALL_NOCHECK(ZeName, ZeArgs) \ - ZeCall().doCall(ZeName ZeArgs, #ZeName, #ZeArgs, false) - struct _ur_platform_handle_t; // using ur_platform_handle_t = _ur_platform_handle_t *; struct _ur_device_handle_t; diff --git a/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero_common.cpp b/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero_common.cpp index afac443636bfd..0a67736409a87 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero_common.cpp +++ b/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero_common.cpp @@ -50,7 +50,7 @@ ur_result_t ze2urResult(ze_result_t ZeResult) { case ZE_RESULT_ERROR_INVALID_GROUP_SIZE_DIMENSION: return UR_RESULT_ERROR_INVALID_WORK_GROUP_SIZE; case ZE_RESULT_ERROR_MODULE_BUILD_FAILURE: - return UR_RESULT_ERROR_MODULE_BUILD_FAILURE; + return UR_RESULT_ERROR_PROGRAM_BUILD_FAILURE; case ZE_RESULT_ERROR_OUT_OF_DEVICE_MEMORY: return UR_RESULT_ERROR_OUT_OF_DEVICE_MEMORY; case ZE_RESULT_ERROR_OUT_OF_HOST_MEMORY: @@ -59,3 +59,205 @@ ur_result_t ze2urResult(ze_result_t ZeResult) { return UR_RESULT_ERROR_UNKNOWN; } } + +void urPrint(const char *Format, ...) { + if (UrL0Debug & UR_L0_DEBUG_BASIC) { + va_list Args; + va_start(Args, Format); + vfprintf(stderr, Format, Args); + va_end(Args); + } +} + +// This function will ensure compatibility with both Linux and Windows for +// setting environment variables. +bool setEnvVar(const char *name, const char *value) { +#ifdef _WIN32 + int Res = _putenv_s(name, value); +#else + int Res = setenv(name, value, 1); +#endif + if (Res != 0) { + urPrint( + "Level Zero plugin was unable to set the environment variable: %s\n", + name); + return false; + } + return true; +} + +// This will count the calls to Level-Zero +std::map *ZeCallCount = nullptr; + +inline void zeParseError(ze_result_t ZeError, const char *&ErrorString) { + switch (ZeError) { +#define ZE_ERRCASE(ERR) \ + case ERR: \ + ErrorString = "" #ERR; \ + break; + + ZE_ERRCASE(ZE_RESULT_SUCCESS) + ZE_ERRCASE(ZE_RESULT_NOT_READY) + ZE_ERRCASE(ZE_RESULT_ERROR_DEVICE_LOST) + ZE_ERRCASE(ZE_RESULT_ERROR_OUT_OF_HOST_MEMORY) + ZE_ERRCASE(ZE_RESULT_ERROR_OUT_OF_DEVICE_MEMORY) + ZE_ERRCASE(ZE_RESULT_ERROR_MODULE_BUILD_FAILURE) + ZE_ERRCASE(ZE_RESULT_ERROR_INSUFFICIENT_PERMISSIONS) + ZE_ERRCASE(ZE_RESULT_ERROR_NOT_AVAILABLE) + ZE_ERRCASE(ZE_RESULT_ERROR_UNINITIALIZED) + ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_VERSION) + ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_FEATURE) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_ARGUMENT) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_NULL_HANDLE) + ZE_ERRCASE(ZE_RESULT_ERROR_HANDLE_OBJECT_IN_USE) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_NULL_POINTER) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_SIZE) + ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_SIZE) + ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_ALIGNMENT) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_SYNCHRONIZATION_OBJECT) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_ENUMERATION) + ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_ENUMERATION) + ZE_ERRCASE(ZE_RESULT_ERROR_UNSUPPORTED_IMAGE_FORMAT) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_NATIVE_BINARY) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_GLOBAL_NAME) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_KERNEL_NAME) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_FUNCTION_NAME) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_GROUP_SIZE_DIMENSION) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_GLOBAL_WIDTH_DIMENSION) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_INDEX) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_SIZE) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_KERNEL_ATTRIBUTE_VALUE) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_COMMAND_LIST_TYPE) + ZE_ERRCASE(ZE_RESULT_ERROR_OVERLAPPING_REGIONS) + ZE_ERRCASE(ZE_RESULT_ERROR_INVALID_MODULE_UNLINKED) + ZE_ERRCASE(ZE_RESULT_ERROR_UNKNOWN) + +#undef ZE_ERRCASE + default: + assert(false && "Unexpected Error code"); + } // switch +} + +ze_result_t ZeCall::doCall(ze_result_t ZeResult, const char *ZeName, + const char *ZeArgs, bool TraceError) { + urPrint("ZE ---> %s%s\n", ZeName, ZeArgs); + + if (UrL0Debug & UR_L0_DEBUG_CALL_COUNT) { + ++(*ZeCallCount)[ZeName]; + } + + if (ZeResult && TraceError) { + const char *ErrorString = "Unknown"; + zeParseError(ZeResult, ErrorString); + urPrint("Error (%s) in %s\n", ErrorString, ZeName); + } + return ZeResult; +} + +// Specializations for various L0 structures +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_EVENT_POOL_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_FENCE_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_COMMAND_LIST_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_CONTEXT_DESC; +} +template <> +ze_structure_type_t +getZeStructureType() { + return ZE_STRUCTURE_TYPE_RELAXED_ALLOCATION_LIMITS_EXP_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_HOST_MEM_ALLOC_DESC; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DEVICE_MEM_ALLOC_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_COMMAND_QUEUE_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_IMAGE_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_MODULE_DESC; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_MODULE_PROGRAM_EXP_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_KERNEL_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_EVENT_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_SAMPLER_DESC; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DRIVER_PROPERTIES; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DEVICE_PROPERTIES; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DEVICE_COMPUTE_PROPERTIES; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_COMMAND_QUEUE_GROUP_PROPERTIES; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DEVICE_IMAGE_PROPERTIES; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DEVICE_MODULE_PROPERTIES; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DEVICE_CACHE_PROPERTIES; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_PROPERTIES; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_EXT_PROPERTIES; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_DEVICE_MEMORY_ACCESS_PROPERTIES; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_MODULE_PROPERTIES; +} +template <> ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_KERNEL_PROPERTIES; +} +template <> +ze_structure_type_t getZeStructureType() { + return ZE_STRUCTURE_TYPE_MEMORY_ALLOCATION_PROPERTIES; +} + +template <> zes_structure_type_t getZesStructureType() { + return ZES_STRUCTURE_TYPE_PCI_PROPERTIES; +} + +template <> zes_structure_type_t getZesStructureType() { + return ZES_STRUCTURE_TYPE_MEM_STATE; +} + +template <> zes_structure_type_t getZesStructureType() { + return ZES_STRUCTURE_TYPE_MEM_PROPERTIES; +} diff --git a/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero_common.hpp b/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero_common.hpp index cbbdb11a24edc..9d13f1a6ec64b 100644 --- a/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero_common.hpp +++ b/sycl/plugins/unified_runtime/ur/adapters/level_zero/ur_level_zero_common.hpp @@ -7,6 +7,11 @@ //===-----------------------------------------------------------------===// #pragma once +#include +#include +#include +#include + #include #include #include @@ -23,5 +28,277 @@ #include "ur_level_zero_queue.hpp" #include "ur_level_zero_sampler.hpp" +template To ur_cast(From Value) { + // TODO: see if more sanity checks are possible. + assert(sizeof(From) == sizeof(To)); + return (To)(Value); +} + +static auto getUrResultString = [](ur_result_t Result) { + switch (Result) { + case UR_RESULT_SUCCESS: + return "UR_RESULT_SUCCESS"; + case UR_RESULT_ERROR_INVALID_OPERATION: + return "UR_RESULT_ERROR_INVALID_OPERATION"; + case UR_RESULT_ERROR_INVALID_QUEUE_PROPERTIES: + return "UR_RESULT_ERROR_INVALID_QUEUE_PROPERTIES"; + case UR_RESULT_ERROR_INVALID_QUEUE: + return "UR_RESULT_ERROR_INVALID_QUEUE"; + case UR_RESULT_ERROR_INVALID_VALUE: + return "UR_RESULT_ERROR_INVALID_VALUE"; + case UR_RESULT_ERROR_INVALID_CONTEXT: + return "UR_RESULT_ERROR_INVALID_CONTEXT"; + case UR_RESULT_ERROR_INVALID_PLATFORM: + return "UR_RESULT_ERROR_INVALID_PLATFORM"; + case UR_RESULT_ERROR_INVALID_BINARY: + return "UR_RESULT_ERROR_INVALID_BINARY"; + case UR_RESULT_ERROR_INVALID_PROGRAM: + return "UR_RESULT_ERROR_INVALID_PROGRAM"; + case UR_RESULT_ERROR_INVALID_SAMPLER: + return "UR_RESULT_ERROR_INVALID_SAMPLER"; + case UR_RESULT_ERROR_INVALID_BUFFER_SIZE: + return "UR_RESULT_ERROR_INVALID_BUFFER_SIZE"; + case UR_RESULT_ERROR_INVALID_MEM_OBJECT: + return "UR_RESULT_ERROR_INVALID_MEM_OBJECT"; + case UR_RESULT_ERROR_INVALID_EVENT: + return "UR_RESULT_ERROR_INVALID_EVENT"; + case UR_RESULT_ERROR_INVALID_EVENT_WAIT_LIST: + return "UR_RESULT_ERROR_INVALID_EVENT_WAIT_LIST"; + case UR_RESULT_ERROR_MISALIGNED_SUB_BUFFER_OFFSET: + return "UR_RESULT_ERROR_MISALIGNED_SUB_BUFFER_OFFSET"; + case UR_RESULT_ERROR_INVALID_WORK_GROUP_SIZE: + return "UR_RESULT_ERROR_INVALID_WORK_GROUP_SIZE"; + case UR_RESULT_ERROR_COMPILER_NOT_AVAILABLE: + return "UR_RESULT_ERROR_COMPILER_NOT_AVAILABLE"; + case UR_RESULT_ERROR_PROFILING_INFO_NOT_AVAILABLE: + return "UR_RESULT_ERROR_PROFILING_INFO_NOT_AVAILABLE"; + case UR_RESULT_ERROR_DEVICE_NOT_FOUND: + return "UR_RESULT_ERROR_DEVICE_NOT_FOUND"; + case UR_RESULT_ERROR_INVALID_DEVICE: + return "UR_RESULT_ERROR_INVALID_DEVICE"; + case UR_RESULT_ERROR_DEVICE_LOST: + return "UR_RESULT_ERROR_DEVICE_LOST"; + case UR_RESULT_ERROR_DEVICE_REQUIRES_RESET: + return "UR_RESULT_ERROR_DEVICE_REQUIRES_RESET"; + case UR_RESULT_ERROR_DEVICE_IN_LOW_POWER_STATE: + return "UR_RESULT_ERROR_DEVICE_IN_LOW_POWER_STATE"; + case UR_RESULT_ERROR_DEVICE_PARTITION_FAILED: + return "UR_RESULT_ERROR_DEVICE_PARTITION_FAILED"; + case UR_RESULT_ERROR_INVALID_DEVICE_PARTITION_COUNT: + return "UR_RESULT_ERROR_INVALID_DEVICE_PARTITION_COUNT"; + case UR_RESULT_ERROR_INVALID_WORK_ITEM_SIZE: + return "UR_RESULT_ERROR_INVALID_WORK_ITEM_SIZE"; + case UR_RESULT_ERROR_INVALID_WORK_DIMENSION: + return "UR_RESULT_ERROR_INVALID_WORK_DIMENSION"; + case UR_RESULT_ERROR_INVALID_KERNEL_ARGS: + return "UR_RESULT_ERROR_INVALID_KERNEL_ARGS"; + case UR_RESULT_ERROR_INVALID_KERNEL: + return "UR_RESULT_ERROR_INVALID_KERNEL"; + case UR_RESULT_ERROR_INVALID_KERNEL_NAME: + return "UR_RESULT_ERROR_INVALID_KERNEL_NAME"; + case UR_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_INDEX: + return "UR_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_INDEX"; + case UR_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_SIZE: + return "UR_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_SIZE"; + case UR_RESULT_ERROR_INVALID_KERNEL_ATTRIBUTE_VALUE: + return "UR_RESULT_ERROR_INVALID_KERNEL_ATTRIBUTE_VALUE"; + case UR_RESULT_ERROR_INVALID_IMAGE_SIZE: + return "UR_RESULT_ERROR_INVALID_IMAGE_SIZE"; + case UR_RESULT_ERROR_INVALID_IMAGE_FORMAT_DESCRIPTOR: + return "UR_RESULT_ERROR_INVALID_IMAGE_FORMAT_DESCRIPTOR"; + case UR_RESULT_ERROR_IMAGE_FORMAT_NOT_SUPPORTED: + return "UR_RESULT_ERROR_IMAGE_FORMAT_NOT_SUPPORTED"; + case UR_RESULT_ERROR_MEM_OBJECT_ALLOCATION_FAILURE: + return "UR_RESULT_ERROR_MEM_OBJECT_ALLOCATION_FAILURE"; + case UR_RESULT_ERROR_INVALID_PROGRAM_EXECUTABLE: + return "UR_RESULT_ERROR_INVALID_PROGRAM_EXECUTABLE"; + case UR_RESULT_ERROR_UNINITIALIZED: + return "UR_RESULT_ERROR_UNINITIALIZED"; + case UR_RESULT_ERROR_OUT_OF_HOST_MEMORY: + return "UR_RESULT_ERROR_OUT_OF_HOST_MEMORY"; + case UR_RESULT_ERROR_OUT_OF_DEVICE_MEMORY: + return "UR_RESULT_ERROR_OUT_OF_DEVICE_MEMORY"; + case UR_RESULT_ERROR_OUT_OF_RESOURCES: + return "UR_RESULT_ERROR_OUT_OF_RESOURCES"; + case UR_RESULT_ERROR_PROGRAM_BUILD_FAILURE: + return "UR_RESULT_ERROR_PROGRAM_BUILD_FAILURE"; + case UR_RESULT_ERROR_PROGRAM_LINK_FAILURE: + return "UR_RESULT_ERROR_PROGRAM_LINK_FAILURE"; + case UR_RESULT_ERROR_UNSUPPORTED_VERSION: + return "UR_RESULT_ERROR_UNSUPPORTED_VERSION"; + case UR_RESULT_ERROR_UNSUPPORTED_FEATURE: + return "UR_RESULT_ERROR_UNSUPPORTED_FEATURE"; + case UR_RESULT_ERROR_INVALID_ARGUMENT: + return "UR_RESULT_ERROR_INVALID_ARGUMENT"; + case UR_RESULT_ERROR_INVALID_NULL_HANDLE: + return "UR_RESULT_ERROR_INVALID_NULL_HANDLE"; + case UR_RESULT_ERROR_HANDLE_OBJECT_IN_USE: + return "UR_RESULT_ERROR_HANDLE_OBJECT_IN_USE"; + case UR_RESULT_ERROR_INVALID_NULL_POINTER: + return "UR_RESULT_ERROR_INVALID_NULL_POINTER"; + case UR_RESULT_ERROR_INVALID_SIZE: + return "UR_RESULT_ERROR_INVALID_SIZE"; + case UR_RESULT_ERROR_UNSUPPORTED_SIZE: + return "UR_RESULT_ERROR_UNSUPPORTED_SIZE"; + case UR_RESULT_ERROR_UNSUPPORTED_ALIGNMENT: + return "UR_RESULT_ERROR_UNSUPPORTED_ALIGNMENT"; + case UR_RESULT_ERROR_INVALID_SYNCHRONIZATION_OBJECT: + return "UR_RESULT_ERROR_INVALID_SYNCHRONIZATION_OBJECT"; + case UR_RESULT_ERROR_INVALID_ENUMERATION: + return "UR_RESULT_ERROR_INVALID_ENUMERATION"; + case UR_RESULT_ERROR_UNSUPPORTED_ENUMERATION: + return "UR_RESULT_ERROR_UNSUPPORTED_ENUMERATION"; + case UR_RESULT_ERROR_UNSUPPORTED_IMAGE_FORMAT: + return "UR_RESULT_ERROR_UNSUPPORTED_IMAGE_FORMAT"; + case UR_RESULT_ERROR_INVALID_NATIVE_BINARY: + return "UR_RESULT_ERROR_INVALID_NATIVE_BINARY"; + case UR_RESULT_ERROR_INVALID_GLOBAL_NAME: + return "UR_RESULT_ERROR_INVALID_GLOBAL_NAME"; + case UR_RESULT_ERROR_INVALID_FUNCTION_NAME: + return "UR_RESULT_ERROR_INVALID_FUNCTION_NAME"; + case UR_RESULT_ERROR_INVALID_GROUP_SIZE_DIMENSION: + return "UR_RESULT_ERROR_INVALID_GROUP_SIZE_DIMENSION"; + case UR_RESULT_ERROR_INVALID_GLOBAL_WIDTH_DIMENSION: + return "UR_RESULT_ERROR_INVALID_GLOBAL_WIDTH_DIMENSION"; + case UR_RESULT_ERROR_PROGRAM_UNLINKED: + return "UR_RESULT_ERROR_PROGRAM_UNLINKED"; + case UR_RESULT_ERROR_OVERLAPPING_REGIONS: + return "UR_RESULT_ERROR_OVERLAPPING_REGIONS"; + case UR_RESULT_ERROR_INVALID_HOST_PTR: + return "UR_RESULT_ERROR_INVALID_HOST_PTR"; + case UR_RESULT_ERROR_INVALID_USM_SIZE: + return "UR_RESULT_ERROR_INVALID_USM_SIZE"; + case UR_RESULT_ERROR_OBJECT_ALLOCATION_FAILURE: + return "UR_RESULT_ERROR_OBJECT_ALLOCATION_FAILURE"; + case UR_RESULT_ERROR_ADAPTER_SPECIFIC: + return "UR_RESULT_ERROR_ADAPTER_SPECIFIC"; + default: + return "UR_RESULT_ERROR_UNKNOWN"; + } +}; + +// Trace an internal PI call; returns in case of an error. +#define UR_CALL(Call) \ + { \ + if (PrintTrace) \ + fprintf(stderr, "UR ---> %s\n", #Call); \ + ur_result_t Result = (Call); \ + if (PrintTrace) \ + fprintf(stderr, "UR <--- %s(%s)\n", #Call, getUrResultString(Result)); \ + if (Result != UR_RESULT_SUCCESS) \ + return Result; \ + } + +// Controls UR L0 calls tracing. +enum UrDebugLevel { + UR_L0_DEBUG_NONE = 0x0, + UR_L0_DEBUG_BASIC = 0x1, + UR_L0_DEBUG_VALIDATION = 0x2, + UR_L0_DEBUG_CALL_COUNT = 0x4, + UR_L0_DEBUG_ALL = -1 +}; + +const int UrL0Debug = [] { + const char *ZeDebugMode = std::getenv("ZE_DEBUG"); + const char *UrL0DebugMode = std::getenv("UR_L0_DEBUG"); + uint32_t DebugMode = 0; + if (UrL0DebugMode) { + DebugMode = std::atoi(UrL0DebugMode); + } else if (ZeDebugMode) { + DebugMode = std::atoi(ZeDebugMode); + } + return DebugMode; +}(); + +// Controls Level Zero calls serialization to w/a Level Zero driver being not MT +// ready. Recognized values (can be used as a bit mask): +enum { + UrL0SerializeNone = + 0, // no locking or blocking (except when SYCL RT requested blocking) + UrL0SerializeLock = 1, // locking around each UR_CALL + UrL0SerializeBlock = + 2, // blocking UR calls, where supported (usually in enqueue commands) +}; + +static const uint32_t UrL0Serialize = [] { + const char *ZeSerializeMode = std::getenv("ZE_SERIALIZE"); + const char *UrL0SerializeMode = std::getenv("UR_L0_SERIALIZE"); + uint32_t SerializeModeValue = 0; + if (UrL0SerializeMode) { + SerializeModeValue = std::atoi(UrL0SerializeMode); + } else if (ZeSerializeMode) { + SerializeModeValue = std::atoi(ZeSerializeMode); + } + return SerializeModeValue; +}(); + +// This class encapsulates actions taken along with a call to Level Zero API. +class ZeCall { +private: + // The global mutex that is used for total serialization of Level Zero calls. + static std::mutex GlobalLock; + +public: + ZeCall() { + if ((UrL0Serialize & UrL0SerializeLock) != 0) { + GlobalLock.lock(); + } + } + ~ZeCall() { + if ((UrL0Serialize & UrL0SerializeLock) != 0) { + GlobalLock.unlock(); + } + } + + // The non-static version just calls static one. + ze_result_t doCall(ze_result_t ZeResult, const char *ZeName, + const char *ZeArgs, bool TraceError = true); +}; + +// This function will ensure compatibility with both Linux and Windows for +// setting environment variables. +bool setEnvVar(const char *name, const char *value); + +// Prints to stderr if UR_L0_DEBUG allows it +void urPrint(const char *Format, ...); + +// Helper for one-liner validation +#define UR_ASSERT(condition, error) \ + if (!(condition)) \ + return error; + +// Returns the ze_structure_type_t to use in .stype of a structured descriptor. +// Intentionally not defined; will give an error if no proper specialization +template ze_structure_type_t getZeStructureType(); +template zes_structure_type_t getZesStructureType(); + +// The helpers to properly default initialize Level-Zero descriptor and +// properties structures. +template struct ZeStruct : public T { + ZeStruct() : T{} { // zero initializes base struct + this->stype = getZeStructureType(); + this->pNext = nullptr; + } +}; + +template struct ZesStruct : public T { + ZesStruct() : T{} { // zero initializes base struct + this->stype = getZesStructureType(); + this->pNext = nullptr; + } +}; + // Map Level Zero runtime error code to UR error code. ur_result_t ze2urResult(ze_result_t ZeResult); + +// Trace a call to Level-Zero RT +#define ZE2UR_CALL(ZeName, ZeArgs) \ + { \ + ze_result_t ZeResult = ZeName ZeArgs; \ + if (auto Result = ZeCall().doCall(ZeResult, #ZeName, #ZeArgs, true)) \ + return ze2urResult(Result); \ + } + +// Perform traced call to L0 without checking for errors +#define ZE_CALL_NOCHECK(ZeName, ZeArgs) \ + ZeCall().doCall(ZeName ZeArgs, #ZeName, #ZeArgs, false)