diff --git a/sycl/include/sycl/detail/pi.def b/sycl/include/sycl/detail/pi.def index d5709aac9dbac..015eeca985248 100644 --- a/sycl/include/sycl/detail/pi.def +++ b/sycl/include/sycl/detail/pi.def @@ -103,6 +103,8 @@ _PI_API(piSamplerCreate) _PI_API(piSamplerGetInfo) _PI_API(piSamplerRetain) _PI_API(piSamplerRelease) +_PI_API(piextSamplerGetNativeHandle) +_PI_API(piextSamplerCreateWithNativeHandle) // Queue commands _PI_API(piEnqueueKernelLaunch) _PI_API(piEnqueueEventsWait) diff --git a/sycl/include/sycl/detail/pi.h b/sycl/include/sycl/detail/pi.h index 5059125da7646..6d81d9df3915b 100644 --- a/sycl/include/sycl/detail/pi.h +++ b/sycl/include/sycl/detail/pi.h @@ -1743,6 +1743,13 @@ __SYCL_EXPORT pi_result piSamplerRetain(pi_sampler sampler); __SYCL_EXPORT pi_result piSamplerRelease(pi_sampler sampler); +__SYCL_EXPORT pi_result piextSamplerCreateWithNativeHandle( + pi_native_handle nativeHandle, pi_context context, + const bool ownNativeHandle, pi_sampler *sampler); + +__SYCL_EXPORT pi_result +piextSamplerGetNativeHandle(pi_sampler sampler, pi_native_handle *nativeHandle); + // // Queue Commands // diff --git a/sycl/plugins/cuda/pi_cuda.cpp b/sycl/plugins/cuda/pi_cuda.cpp index de715de0835fd..2fd164d6de421 100644 --- a/sycl/plugins/cuda/pi_cuda.cpp +++ b/sycl/plugins/cuda/pi_cuda.cpp @@ -637,6 +637,19 @@ pi_result piSamplerRelease(pi_sampler Sampler) { return pi2ur::piSamplerRelease(Sampler); } +pi_result piextSamplerGetNativeHandle(pi_sampler sampler, + pi_native_handle *nativeHandle) { + return pi2ur::piextSamplerGetNativeHandle(sampler, nativeHandle); +} + +pi_result piextSamplerCreateWithNativeHandle(pi_native_handle nativeHandle, + pi_context context, + const bool ownNativeHandle, + pi_sampler *sampler) { + return pi2ur::piextSamplerCreateWithNativeHandle(nativeHandle, context, + ownNativeHandle, sampler); +} + pi_result piEnqueueEventsWait(pi_queue Queue, pi_uint32 NumEventsInWaitList, const pi_event *EventWaitList, pi_event *OutEvent) { diff --git a/sycl/plugins/hip/pi_hip.cpp b/sycl/plugins/hip/pi_hip.cpp index 126ada92348f6..144d85efdb8a8 100644 --- a/sycl/plugins/hip/pi_hip.cpp +++ b/sycl/plugins/hip/pi_hip.cpp @@ -640,6 +640,19 @@ pi_result piSamplerRelease(pi_sampler Sampler) { return pi2ur::piSamplerRelease(Sampler); } +pi_result piextSamplerGetNativeHandle(pi_sampler sampler, + pi_native_handle *nativeHandle) { + return pi2ur::piextSamplerGetNativeHandle(sampler, nativeHandle); +} + +pi_result piextSamplerCreateWithNativeHandle(pi_native_handle nativeHandle, + pi_context context, + const bool ownNativeHandle, + pi_sampler *sampler) { + return pi2ur::piextSamplerCreateWithNativeHandle(nativeHandle, context, + ownNativeHandle, sampler); +} + pi_result piEnqueueEventsWait(pi_queue Queue, pi_uint32 NumEventsInWaitList, const pi_event *EventWaitList, pi_event *OutEvent) { diff --git a/sycl/plugins/level_zero/pi_level_zero.cpp b/sycl/plugins/level_zero/pi_level_zero.cpp index 0fc36a231be6c..34bbfb9a66afd 100644 --- a/sycl/plugins/level_zero/pi_level_zero.cpp +++ b/sycl/plugins/level_zero/pi_level_zero.cpp @@ -659,6 +659,19 @@ pi_result piSamplerRelease(pi_sampler Sampler) { return pi2ur::piSamplerRelease(Sampler); } +pi_result piextSamplerGetNativeHandle(pi_sampler sampler, + pi_native_handle *nativeHandle) { + return pi2ur::piextSamplerGetNativeHandle(sampler, nativeHandle); +} + +pi_result piextSamplerCreateWithNativeHandle(pi_native_handle nativeHandle, + pi_context context, + const bool ownNativeHandle, + pi_sampler *sampler) { + return pi2ur::piextSamplerCreateWithNativeHandle(nativeHandle, context, + ownNativeHandle, sampler); +} + // // Queue Commands // diff --git a/sycl/plugins/native_cpu/pi_native_cpu.cpp b/sycl/plugins/native_cpu/pi_native_cpu.cpp index 48ce104a94e90..6340d7f4eb52c 100644 --- a/sycl/plugins/native_cpu/pi_native_cpu.cpp +++ b/sycl/plugins/native_cpu/pi_native_cpu.cpp @@ -654,6 +654,19 @@ pi_result piSamplerRelease(pi_sampler Sampler) { return pi2ur::piSamplerRelease(Sampler); } +pi_result piextSamplerGetNativeHandle(pi_sampler sampler, + pi_native_handle *nativeHandle) { + return pi2ur::piextSamplerGetNativeHandle(sampler, nativeHandle); +} + +pi_result piextSamplerCreateWithNativeHandle(pi_native_handle nativeHandle, + pi_context context, + const bool ownNativeHandle, + pi_sampler *sampler) { + return pi2ur::piextSamplerCreateWithNativeHandle(nativeHandle, context, + ownNativeHandle, sampler); +} + // // Queue Commands // diff --git a/sycl/plugins/opencl/pi_opencl.cpp b/sycl/plugins/opencl/pi_opencl.cpp index c09be92f89406..eab0e36afd46d 100644 --- a/sycl/plugins/opencl/pi_opencl.cpp +++ b/sycl/plugins/opencl/pi_opencl.cpp @@ -616,6 +616,19 @@ pi_result piSamplerRelease(pi_sampler Sampler) { return pi2ur::piSamplerRelease(Sampler); } +pi_result piextSamplerGetNativeHandle(pi_sampler sampler, + pi_native_handle *nativeHandle) { + return pi2ur::piextSamplerGetNativeHandle(sampler, nativeHandle); +} + +pi_result piextSamplerCreateWithNativeHandle(pi_native_handle nativeHandle, + pi_context context, + const bool ownNativeHandle, + pi_sampler *sampler) { + return pi2ur::piextSamplerCreateWithNativeHandle(nativeHandle, context, + ownNativeHandle, sampler); +} + pi_result piEnqueueEventsWait(pi_queue Queue, pi_uint32 NumEventsInWaitList, const pi_event *EventWaitList, pi_event *OutEvent) { diff --git a/sycl/plugins/unified_runtime/CMakeLists.txt b/sycl/plugins/unified_runtime/CMakeLists.txt index f71e70e8ed79c..589d40573db3d 100644 --- a/sycl/plugins/unified_runtime/CMakeLists.txt +++ b/sycl/plugins/unified_runtime/CMakeLists.txt @@ -56,16 +56,14 @@ endif() if(SYCL_PI_UR_USE_FETCH_CONTENT) include(FetchContent) - set(UNIFIED_RUNTIME_REPO "https://github.com/oneapi-src/unified-runtime.git") - # commit 69ac6c7fecac1028024ec1889d3e2cc02565c9e7 - # Merge: f11823e1 b3469f1e - # Author: Kenneth Benzie (Benie) - # Date: Thu Feb 15 15:28:12 2024 +0100 - # - # Merge pull request #1344 from aarongreig/aaron/correctlyMapBufferLocation - # - # Correct mapping of BUFFER_ALLOC_LOCATION_PROPERITES in CL adapter. - set(UNIFIED_RUNTIME_TAG 69ac6c7fecac1028024ec1889d3e2cc02565c9e7) + set(UNIFIED_RUNTIME_REPO "https://github.com/omarahmed1111/unified-runtime.git") + # commit 47af3ee296ae0517213114332ffd3ac54a456b16 + # Merge: bd76c510 f2ca7a91 + # Author: Omar Ahmed <30423288+omarahmed1111@users.noreply.github.com> + # Date: Thu Nov 30 16:11:56 2023 +0000 + # - Merge pull request #1072 from omarahmed1111/merge-some-main-changes-into-adapters-third-patch + # - Merge main into adapters branch + set(UNIFIED_RUNTIME_TAG 9bfbf6e654ed3813926e041f9470689b46569907) if(SYCL_PI_UR_OVERRIDE_FETCH_CONTENT_REPO) set(UNIFIED_RUNTIME_REPO "${SYCL_PI_UR_OVERRIDE_FETCH_CONTENT_REPO}") diff --git a/sycl/plugins/unified_runtime/pi2ur.hpp b/sycl/plugins/unified_runtime/pi2ur.hpp index e3a92022567d0..26b73029f1248 100644 --- a/sycl/plugins/unified_runtime/pi2ur.hpp +++ b/sycl/plugins/unified_runtime/pi2ur.hpp @@ -4412,6 +4412,38 @@ inline pi_result piSamplerRelease(pi_sampler Sampler) { return PI_SUCCESS; } +pi_result piextSamplerGetNativeHandle(pi_sampler sampler, + pi_native_handle *nativeHandle) { + auto UrSampler = reinterpret_cast(sampler); + + ur_native_handle_t UrNativeHandle{}; + HANDLE_ERRORS(urSamplerGetNativeHandle(UrSampler, &UrNativeHandle)); + + *nativeHandle = reinterpret_cast(UrNativeHandle); + + return PI_SUCCESS; +} + +pi_result piextSamplerCreateWithNativeHandle(pi_native_handle nativeHandle, + pi_context context, + bool ownNativeHandle, + pi_sampler *sampler) { + ur_native_handle_t UrNativeSampler = + reinterpret_cast(nativeHandle); + + ur_context_handle_t UrContext = + reinterpret_cast(context); + + ur_sampler_handle_t *URSampler = + reinterpret_cast(sampler); + ur_sampler_native_properties_t Properties{}; + Properties.isNativeHandleOwned = ownNativeHandle; + HANDLE_ERRORS(urSamplerCreateWithNativeHandle(UrNativeSampler, UrContext, + &Properties, URSampler)); + + return PI_SUCCESS; +} + // Sampler /////////////////////////////////////////////////////////////////////////////// diff --git a/sycl/plugins/unified_runtime/pi_unified_runtime.cpp b/sycl/plugins/unified_runtime/pi_unified_runtime.cpp index b9742b8697fa8..36737cd2060aa 100644 --- a/sycl/plugins/unified_runtime/pi_unified_runtime.cpp +++ b/sycl/plugins/unified_runtime/pi_unified_runtime.cpp @@ -910,6 +910,18 @@ __SYCL_EXPORT pi_result piSamplerRetain(pi_sampler Sampler) { __SYCL_EXPORT pi_result piSamplerRelease(pi_sampler Sampler) { return pi2ur::piSamplerRelease(Sampler); } +/* clang-format off */ +__SYCL_EXPORT pi_result piextSamplerGetNativeHandle( + pi_sampler sampler, pi_native_handle *nativeHandle) { + return pi2ur::piextSamplerGetNativeHandle(sampler, nativeHandle); +} +/* clang-format on */ +__SYCL_EXPORT pi_result piextSamplerCreateWithNativeHandle( + pi_native_handle nativeHandle, pi_context context, + const bool ownNativeHandle, pi_sampler *sampler) { + return pi2ur::piextSamplerCreateWithNativeHandle(nativeHandle, context, + ownNativeHandle, sampler); +} __SYCL_EXPORT pi_result piMemImageGetInfo(pi_mem Image, pi_image_info ParamName, size_t ParamValueSize, @@ -1484,6 +1496,8 @@ __SYCL_EXPORT pi_result piPluginInit(pi_plugin *PluginInit) { _PI_API(piSamplerGetInfo) _PI_API(piSamplerRetain) _PI_API(piSamplerRelease) + _PI_API(piextSamplerCreateWithNativeHandle) + _PI_API(piextSamplerGetNativeHandle) // Peer to Peer _PI_API(piextEnablePeerAccess) diff --git a/sycl/source/CMakeLists.txt b/sycl/source/CMakeLists.txt index 0da1c5fa0cea2..57cfc9d14147d 100644 --- a/sycl/source/CMakeLists.txt +++ b/sycl/source/CMakeLists.txt @@ -129,6 +129,9 @@ function(add_sycl_rt_library LIB_NAME LIB_OBJ_NAME) target_link_libraries(${LIB_OBJ_NAME} PRIVATE OpenCL-Headers ) + target_link_libraries(${LIB_NAME} + PRIVATE OpenCL::OpenCL + ) if(SYCL_ENABLE_KERNEL_FUSION) target_link_libraries(${LIB_NAME} PRIVATE sycl-fusion) diff --git a/sycl/source/backend.cpp b/sycl/source/backend.cpp index ed0539f266ee2..591c9931c096d 100644 --- a/sycl/source/backend.cpp +++ b/sycl/source/backend.cpp @@ -94,6 +94,11 @@ __SYCL_EXPORT device make_device(pi_native_handle NativeHandle, __SYCL_EXPORT context make_context(pi_native_handle NativeHandle, const async_handler &Handler, backend Backend) { + if (Backend == backend::opencl) { + sycl::context Context(detail::pi::cast(NativeHandle), Handler); + return Context; + } + const auto &Plugin = getPlugin(Backend); pi::PiContext PiContext = nullptr; @@ -127,7 +132,11 @@ __SYCL_EXPORT queue make_queue(pi_native_handle NativeHandle, make_error_code(errc::invalid), "Queue create using make_queue cannot have compute_index property."); } - + if (Backend == backend::opencl) { + sycl::queue Queue(detail::pi::cast(NativeHandle), Context, + Handler); + return Queue; + } // Create PI queue first. pi::PiQueue PiQueue = nullptr; Plugin->call( @@ -276,6 +285,19 @@ kernel make_kernel(const context &TargetContext, *KernelBundle.begin(); const auto &DeviceImageImpl = getSyclObjImpl(DeviceImage); PiProgram = DeviceImageImpl->get_program_ref(); + } else if (Backend == backend::opencl) { + cl_kernel CLKernel = detail::pi::cast(NativeHandle); + cl_program CLProgram; + size_t Ret = clGetKernelInfo(CLKernel, CL_KERNEL_PROGRAM, sizeof(CLProgram), + &CLProgram, nullptr); + if (Ret != CL_SUCCESS) { + throw runtime_error( + "Failed to retrieve program associated with the kernel", + PI_ERROR_INVALID_KERNEL); + } + Plugin->call( + detail::pi::cast(CLProgram), + ContextImpl->getHandleRef(), false, &PiProgram); } // Create PI kernel first. diff --git a/sycl/source/context.cpp b/sycl/source/context.cpp index 065c5af414b05..e4a1a4e724901 100644 --- a/sycl/source/context.cpp +++ b/sycl/source/context.cpp @@ -83,11 +83,35 @@ context::context(const std::vector &DeviceList, PropList); } } + context::context(cl_context ClContext, async_handler AsyncHandler) { const auto &Plugin = sycl::detail::pi::getPlugin(); - impl = std::make_shared( - detail::pi::cast(ClContext), AsyncHandler, - Plugin); + uint32_t DeviceCount = 0; + size_t Ret = clGetContextInfo(ClContext, CL_CONTEXT_NUM_DEVICES, + sizeof(DeviceCount), &DeviceCount, nullptr); + if (Ret != CL_SUCCESS) { + throw runtime_error( + "Failed to retrieve device count associated with the context", + PI_ERROR_INVALID_CONTEXT); + } + std::vector CLDevices(DeviceCount); + Ret = clGetContextInfo(ClContext, CL_CONTEXT_DEVICES, sizeof(CLDevices), + CLDevices.data(), nullptr); + if (Ret != CL_SUCCESS) { + throw runtime_error( + "Failed to retrieve devices associated with the context", + PI_ERROR_INVALID_CONTEXT); + } + std::vector Devices(DeviceCount); + for (uint32_t i = 0; i < DeviceCount; i++) { + Plugin->call( + detail::pi::cast(CLDevices[i]), nullptr, &Devices[i]); + } + sycl::detail::pi::PiContext Context; + Plugin->call( + detail::pi::cast(ClContext), DeviceCount, + Devices.data(), false, &Context); + impl = std::make_shared(Context, AsyncHandler, Plugin); } template @@ -122,7 +146,9 @@ context::get_info() const { #undef __SYCL_PARAM_TRAITS_SPEC -cl_context context::get() const { return impl->get(); } +cl_context context::get() const { + return detail::pi::cast(impl->getNative()); +} bool context::is_host() const { bool IsHost = impl->is_host(); diff --git a/sycl/source/detail/buffer_impl.cpp b/sycl/source/detail/buffer_impl.cpp index 835c732a40bf9..9c89dbfd1a801 100644 --- a/sycl/source/detail/buffer_impl.cpp +++ b/sycl/source/detail/buffer_impl.cpp @@ -54,7 +54,10 @@ void buffer_impl::addInteropObject( const PluginPtr &Plugin = getPlugin(); Plugin->call( pi::cast(MInteropMemObject)); - Handles.push_back(pi::cast(MInteropMemObject)); + pi_native_handle Handle; + Plugin->call(MInteropMemObject, + nullptr, &Handle); + Handles.push_back(Handle); } } } diff --git a/sycl/source/detail/context_impl.cpp b/sycl/source/detail/context_impl.cpp index 46894a6d9e650..9829ed3203284 100644 --- a/sycl/source/detail/context_impl.cpp +++ b/sycl/source/detail/context_impl.cpp @@ -126,9 +126,12 @@ cl_context context_impl::get() const { "This instance of context doesn't support OpenCL interoperability.", PI_ERROR_INVALID_CONTEXT); } - // TODO catch an exception and put it to list of asynchronous exceptions getPlugin()->call(MContext); - return pi::cast(MContext); + // TODO catch an exception and put it to list of asynchronous exceptions + pi_native_handle NativeContext; + getPlugin()->call(MContext, + &NativeContext); + return pi::cast(NativeContext); } bool context_impl::is_host() const { return MHostContext; } diff --git a/sycl/source/detail/kernel_impl.cpp b/sycl/source/detail/kernel_impl.cpp index 12ae796b2e2e2..0a29c7421cf1c 100644 --- a/sycl/source/detail/kernel_impl.cpp +++ b/sycl/source/detail/kernel_impl.cpp @@ -44,16 +44,6 @@ kernel_impl::kernel_impl(sycl::detail::pi::PiKernel Kernel, MCreatedFromSource(IsCreatedFromSource), MKernelBundleImpl(std::move(KernelBundleImpl)), MKernelArgMaskPtr{ArgMask} { - - sycl::detail::pi::PiContext Context = nullptr; - // Using the plugin from the passed ContextImpl - getPlugin()->call( - MKernel, PI_KERNEL_INFO_CONTEXT, sizeof(Context), &Context, nullptr); - if (ContextImpl->getHandleRef() != Context) - throw sycl::invalid_parameter_error( - "Input context must be the same as the context of cl_kernel", - PI_ERROR_INVALID_CONTEXT); - MIsInterop = MProgramImpl->isInterop(); } diff --git a/sycl/source/detail/kernel_impl.hpp b/sycl/source/detail/kernel_impl.hpp index 057a2dcf69e15..9dfcda8478356 100644 --- a/sycl/source/detail/kernel_impl.hpp +++ b/sycl/source/detail/kernel_impl.hpp @@ -106,7 +106,11 @@ class kernel_impl { PI_ERROR_INVALID_KERNEL); } getPlugin()->call(MKernel); - return pi::cast(MKernel); + // TODO catch an exception and put it to list of asynchronous exceptions + pi_native_handle NativeKernel; + getPlugin()->call(MKernel, + &NativeKernel); + return pi::cast(NativeKernel); } /// Check if the associated SYCL context is a SYCL host context. diff --git a/sycl/source/detail/program_impl.cpp b/sycl/source/detail/program_impl.cpp index f3c42e2adcd97..283fbf1430e84 100644 --- a/sycl/source/detail/program_impl.cpp +++ b/sycl/source/detail/program_impl.cpp @@ -222,7 +222,11 @@ cl_program program_impl::get() const { PI_ERROR_INVALID_PROGRAM); } getPlugin()->call(MProgram); - return pi::cast(MProgram); + // TODO catch an exception and put it to list of asynchronous exceptions + pi_native_handle NativeProgram; + getPlugin()->call(MProgram, + &NativeProgram); + return pi::cast(NativeProgram); } void program_impl::compile_with_kernel_name(std::string KernelName, @@ -389,23 +393,22 @@ std::pair program_impl::get_pi_kernel_arg_mask_pair(const std::string &KernelName) const { std::pair Result; - const PluginPtr &Plugin = getPlugin(); - sycl::detail::pi::PiResult Err = - Plugin->call_nocheck( - MProgram, KernelName.c_str(), &Result.first); - if (Err == PI_ERROR_INVALID_KERNEL_NAME) { - throw invalid_object_error( - "This instance of program does not contain the kernel requested", - Err); - } - Plugin->checkPiResult(Err); + const PluginPtr &Plugin = getPlugin(); + sycl::detail::pi::PiResult Err = + Plugin->call_nocheck( + MProgram, KernelName.c_str(), &Result.first); + if (Err == PI_ERROR_INVALID_KERNEL_NAME) { + throw invalid_object_error( + "This instance of program does not contain the kernel requested", Err); + } + Plugin->checkPiResult(Err); - // Some PI Plugins (like OpenCL) require this call to enable USM - // For others, PI will turn this into a NOP. - Plugin->call( - Result.first, PI_USM_INDIRECT_ACCESS, sizeof(pi_bool), &PI_TRUE); + // Some PI Plugins (like OpenCL) require this call to enable USM + // For others, PI will turn this into a NOP. + Plugin->call( + Result.first, PI_USM_INDIRECT_ACCESS, sizeof(pi_bool), &PI_TRUE); - return Result; + return Result; } std::vector diff --git a/sycl/source/detail/queue_impl.hpp b/sycl/source/detail/queue_impl.hpp index 0fe4242cc9472..8eb73fddc0e82 100644 --- a/sycl/source/detail/queue_impl.hpp +++ b/sycl/source/detail/queue_impl.hpp @@ -353,7 +353,11 @@ class queue_impl { PI_ERROR_INVALID_QUEUE); } getPlugin()->call(MQueues[0]); - return pi::cast(MQueues[0]); + // TODO catch an exception and put it to list of asynchronous exceptions + pi_native_handle NativeQueue; + getPlugin()->call( + MQueues[0], &NativeQueue, nullptr); + return pi::cast(NativeQueue); } /// \return an associated SYCL context. diff --git a/sycl/source/detail/sampler_impl.cpp b/sycl/source/detail/sampler_impl.cpp index 78473d87b3689..032336ec3d52a 100644 --- a/sycl/source/detail/sampler_impl.cpp +++ b/sycl/source/detail/sampler_impl.cpp @@ -23,11 +23,14 @@ sampler_impl::sampler_impl(coordinate_normalization_mode normalizationMode, sampler_impl::sampler_impl(cl_sampler clSampler, const context &syclContext) { - sycl::detail::pi::PiSampler Sampler = - pi::cast(clSampler); + sycl::detail::pi::PiSampler Sampler; + + auto Plugin = sycl::detail::pi::getPlugin(); + Plugin->call( + pi::cast(clSampler), + getSyclObjImpl(syclContext)->getHandleRef(), false, &Sampler); + MContextToSampler[syclContext] = Sampler; - const PluginPtr &Plugin = getSyclObjImpl(syclContext)->getPlugin(); - Plugin->call(Sampler); Plugin->call( Sampler, PI_SAMPLER_INFO_NORMALIZED_COORDS, sizeof(pi_bool), &MCoordNormMode, nullptr); diff --git a/sycl/source/detail/scheduler/commands.cpp b/sycl/source/detail/scheduler/commands.cpp index 955adae8423dc..f979d72a01472 100644 --- a/sycl/source/detail/scheduler/commands.cpp +++ b/sycl/source/detail/scheduler/commands.cpp @@ -2269,23 +2269,11 @@ void SetArgBasedOnType( getMemAllocationFunc ? (sycl::detail::pi::PiMem)getMemAllocationFunc(Req) : nullptr; - if (Context.get_backend() == backend::opencl) { - // clSetKernelArg (corresponding to piKernelSetArg) returns an error - // when MemArg is null, which is the case when zero-sized buffers are - // handled. Below assignment provides later call to clSetKernelArg with - // acceptable arguments. - if (!MemArg) - MemArg = sycl::detail::pi::PiMem(); - - Plugin->call( - Kernel, NextTrueIndex, sizeof(sycl::detail::pi::PiMem), &MemArg); - } else { - pi_mem_obj_property MemObjData{}; - MemObjData.mem_access = AccessModeToPi(Req->MAccessMode); - MemObjData.type = PI_KERNEL_ARG_MEM_OBJ_ACCESS; - Plugin->call(Kernel, NextTrueIndex, - &MemObjData, &MemArg); - } + pi_mem_obj_property MemObjData{}; + MemObjData.mem_access = AccessModeToPi(Req->MAccessMode); + MemObjData.type = PI_KERNEL_ARG_MEM_OBJ_ACCESS; + Plugin->call(Kernel, NextTrueIndex, + &MemObjData, &MemArg); break; } case kernel_param_kind_t::kind_std_layout: { diff --git a/sycl/source/device.cpp b/sycl/source/device.cpp index 73a05080c5b0d..a174beacf8e22 100644 --- a/sycl/source/device.cpp +++ b/sycl/source/device.cpp @@ -77,7 +77,9 @@ std::vector device::get_devices(info::device_type deviceType) { return devices; } -cl_device_id device::get() const { return impl->get(); } +cl_device_id device::get() const { + return detail::pi::cast(impl->getNative()); +} bool device::is_host() const { bool IsHost = impl->is_host(); diff --git a/sycl/source/event.cpp b/sycl/source/event.cpp index c8f170e68cf84..51ad3bf0f3b36 100644 --- a/sycl/source/event.cpp +++ b/sycl/source/event.cpp @@ -25,13 +25,14 @@ inline namespace _V1 { event::event() : impl(std::make_shared(std::nullopt)) {} -event::event(cl_event ClEvent, const context &SyclContext) - : impl(std::make_shared( - detail::pi::cast(ClEvent), SyclContext)) { - // This is a special interop constructor for OpenCL, so the event must be - // retained. - impl->getPlugin()->call( - detail::pi::cast(ClEvent)); +event::event(cl_event ClEvent, const context &SyclContext) { + sycl::detail::pi::PiEvent Event; + auto Plugin = sycl::detail::pi::getPlugin(); + Plugin->call( + detail::pi::cast(ClEvent), + detail::getSyclObjImpl(SyclContext)->getHandleRef(), false, &Event); + + impl = std::make_shared(Event, SyclContext); } bool event::operator==(const event &rhs) const { return rhs.impl == impl; } diff --git a/sycl/source/kernel.cpp b/sycl/source/kernel.cpp index 369b631c88464..6a2210ff580f7 100644 --- a/sycl/source/kernel.cpp +++ b/sycl/source/kernel.cpp @@ -16,19 +16,39 @@ namespace sycl { inline namespace _V1 { -kernel::kernel(cl_kernel ClKernel, const context &SyclContext) - : impl(std::make_shared( - detail::pi::cast(ClKernel), - detail::getSyclObjImpl(SyclContext), nullptr, nullptr)) { - // This is a special interop constructor for OpenCL, so the kernel must be - // retained. - if (get_backend() == backend::opencl) { - impl->getPlugin()->call( - detail::pi::cast(ClKernel)); +kernel::kernel(cl_kernel ClKernel, const context &SyclContext) { + try { + sycl::detail::pi::PiKernel Kernel; + auto Context = detail::getSyclObjImpl(SyclContext); + auto Plugin = sycl::detail::pi::getPlugin(); + cl_program CLProgram; + size_t Ret = clGetKernelInfo(ClKernel, CL_KERNEL_PROGRAM, sizeof(CLProgram), + &CLProgram, nullptr); + if (Ret != CL_SUCCESS) { + throw runtime_error( + "Failed to retrieve program associated with the kernel", + PI_ERROR_INVALID_KERNEL); + } + sycl::detail::pi::PiProgram Program; + Plugin->call( + detail::pi::cast(CLProgram), Context->getHandleRef(), + false, &Program); + + Plugin->call( + detail::pi::cast(ClKernel), Context->getHandleRef(), + Program, false, &Kernel); + impl = std::make_shared(Kernel, Context, nullptr, + nullptr); + } catch (sycl::runtime_error &) { + throw sycl::invalid_parameter_error( + "Input context must be the same as the context of cl_kernel", + PI_ERROR_INVALID_CONTEXT); } } -cl_kernel kernel::get() const { return impl->get(); } +cl_kernel kernel::get() const { + return detail::pi::cast(impl->getNative()); +} bool kernel::is_host() const { bool IsHost = impl->is_host(); diff --git a/sycl/source/platform.cpp b/sycl/source/platform.cpp index 5cc2a49801902..5cae3c8c36361 100644 --- a/sycl/source/platform.cpp +++ b/sycl/source/platform.cpp @@ -21,9 +21,12 @@ inline namespace _V1 { platform::platform() : platform(default_selector_v) {} platform::platform(cl_platform_id PlatformId) { - impl = detail::platform_impl::getOrMakePlatformImpl( - detail::pi::cast(PlatformId), - sycl::detail::pi::getPlugin()); + sycl::detail::pi::PiPlatform Platform; + auto Plugin = sycl::detail::pi::getPlugin(); + Plugin->call( + detail::pi::cast(PlatformId), &Platform); + + impl = detail::platform_impl::getOrMakePlatformImpl(Platform, Plugin); } // protected constructor for internal use @@ -33,7 +36,9 @@ platform::platform(const device_selector &dev_selector) { *this = dev_selector.select_device().get_platform(); } -cl_platform_id platform::get() const { return impl->get(); } +cl_platform_id platform::get() const { + return detail::pi::cast(impl->getNative()); +} bool platform::has_extension(const std::string &ExtensionName) const { return impl->has_extension(ExtensionName); diff --git a/sycl/source/queue.cpp b/sycl/source/queue.cpp index 1b877a31da4e0..53c38e016eded 100644 --- a/sycl/source/queue.cpp +++ b/sycl/source/queue.cpp @@ -65,12 +65,35 @@ queue::queue(const context &SyclContext, const device &SyclDevice, queue::queue(cl_command_queue clQueue, const context &SyclContext, const async_handler &AsyncHandler) { const property_list PropList{}; - impl = std::make_shared( - reinterpret_cast(clQueue), - detail::getSyclObjImpl(SyclContext), AsyncHandler, PropList); + sycl::detail::pi::PiQueue Queue; + auto Context = detail::getSyclObjImpl(SyclContext); + auto Plugin = sycl::detail::pi::getPlugin(); + + cl_device_id CLDevice; + size_t Ret = clGetCommandQueueInfo(clQueue, CL_QUEUE_DEVICE, sizeof(CLDevice), + &CLDevice, nullptr); + if (Ret) { + throw runtime_error("Failed to retrieve device associated with the queue", + PI_ERROR_INVALID_QUEUE); + } + sycl::detail::pi::PiDevice Device; + Plugin->call( + detail::pi::cast(CLDevice), nullptr, &Device); + + sycl::detail::pi::PiQueueProperties Properties[] = {PI_QUEUE_FLAGS, 0, 0, 0, + 0}; + Plugin->call( + detail::pi::cast(clQueue), 0, Context->getHandleRef(), + Device, false, Properties, &Queue); + + impl = std::make_shared(Queue, Context, AsyncHandler, + PropList); } -cl_command_queue queue::get() const { return impl->get(); } +cl_command_queue queue::get() const { + int32_t NativeHandleDesc = 0; + return detail::pi::cast(impl->getNative(NativeHandleDesc)); +} context queue::get_context() const { return impl->get_context(); } diff --git a/sycl/test-e2e/DeprecatedFeatures/set_arg_interop.cpp b/sycl/test-e2e/DeprecatedFeatures/set_arg_interop.cpp index 986f9314823ab..70e938b1a2815 100644 --- a/sycl/test-e2e/DeprecatedFeatures/set_arg_interop.cpp +++ b/sycl/test-e2e/DeprecatedFeatures/set_arg_interop.cpp @@ -6,6 +6,7 @@ #include #include +#include using namespace sycl; @@ -15,7 +16,7 @@ int main() { cl_context ClContext = Context.get(); - const size_t CountSources = 3; + const size_t CountSources = 4; const char *Sources[CountSources] = { "kernel void foo1(global float* Array, global int* Value) { *Array = " "42; *Value = 1; }\n", @@ -23,6 +24,7 @@ int main() { "Array[id] = id; }\n", "kernel void foo3(global float* Array, local float* LocalArray) { " "(void)LocalArray; (void)Array; }\n", + "kernel void foo4(global int* Value) {}\n", }; cl_int Err; @@ -42,12 +44,16 @@ int main() { cl_kernel ThirdCLKernel = clCreateKernel(ClProgram, "foo3", &Err); assert(Err == CL_SUCCESS); + cl_kernel FourthCLKernel = clCreateKernel(ClProgram, "foo4", &Err); + assert(Err == CL_SUCCESS); + const size_t Count = 100; float Array[Count]; kernel FirstKernel(FirstCLKernel, Context); kernel SecondKernel(SecondCLKernel, Context); kernel ThirdKernel(ThirdCLKernel, Context); + kernel FourthKernel(FourthCLKernel, Context); int Value; { buffer FirstBuffer(Array, range<1>(1)); @@ -110,10 +116,37 @@ int main() { } Queue.wait_and_throw(); + // Enqueuing an interop kernel while avoid calls to piKernelSetArg from + // different threads on the same kernel. + { + constexpr std::size_t NArgs = 16; + constexpr std::size_t ThreadCount = 4; + constexpr std::size_t LaunchCount = 8; + auto TestLambda = [&](int ThreadId) { + Queue + .submit([&](sycl::handler &CGH) { + for (std::size_t I = 0; I < NArgs; ++I) + CGH.set_arg(I, &ThreadId); + }) + .wait(); + }; + + std::vector threadPool; + threadPool.reserve(ThreadCount); + for (size_t tid = 0; tid < ThreadCount; ++tid) { + threadPool.push_back(std::thread(TestLambda, tid)); + } + + for (auto ¤tThread : threadPool) { + currentThread.join(); + } + } + clReleaseContext(ClContext); clReleaseKernel(FirstCLKernel); clReleaseKernel(SecondCLKernel); clReleaseKernel(ThirdCLKernel); + clReleaseKernel(FourthCLKernel); clReleaseProgram(ClProgram); return 0; } diff --git a/sycl/test/abi/pi_cuda_symbol_check.dump b/sycl/test/abi/pi_cuda_symbol_check.dump index 0aaff91e58336..0ab5b8b9d86b7 100644 --- a/sycl/test/abi/pi_cuda_symbol_check.dump +++ b/sycl/test/abi/pi_cuda_symbol_check.dump @@ -81,6 +81,8 @@ piSamplerCreate piSamplerGetInfo piSamplerRelease piSamplerRetain +piextSamplerGetNativeHandle +piextSamplerCreateWithNativeHandle piTearDown piextBindlessImageSamplerCreate piextCommandBufferAdviseUSM diff --git a/sycl/test/abi/pi_hip_symbol_check.dump b/sycl/test/abi/pi_hip_symbol_check.dump index 3bf9f9a3bf31d..b20c7afebfc18 100644 --- a/sycl/test/abi/pi_hip_symbol_check.dump +++ b/sycl/test/abi/pi_hip_symbol_check.dump @@ -81,6 +81,8 @@ piSamplerCreate piSamplerGetInfo piSamplerRelease piSamplerRetain +piextSamplerGetNativeHandle +piextSamplerCreateWithNativeHandle piTearDown piextBindlessImageSamplerCreate piextCommandBufferAdviseUSM diff --git a/sycl/test/abi/pi_level_zero_symbol_check.dump b/sycl/test/abi/pi_level_zero_symbol_check.dump index 80ba2ad78d8ee..57953216c3435 100644 --- a/sycl/test/abi/pi_level_zero_symbol_check.dump +++ b/sycl/test/abi/pi_level_zero_symbol_check.dump @@ -80,6 +80,8 @@ piSamplerCreate piSamplerGetInfo piSamplerRelease piSamplerRetain +piextSamplerGetNativeHandle +piextSamplerCreateWithNativeHandle piTearDown piextBindlessImageSamplerCreate piextCommandBufferAdviseUSM diff --git a/sycl/test/abi/pi_nativecpu_symbol_check.dump b/sycl/test/abi/pi_nativecpu_symbol_check.dump index 3a2e654187d4d..03e3d9fd60bd8 100644 --- a/sycl/test/abi/pi_nativecpu_symbol_check.dump +++ b/sycl/test/abi/pi_nativecpu_symbol_check.dump @@ -81,6 +81,8 @@ piSamplerCreate piSamplerGetInfo piSamplerRelease piSamplerRetain +piextSamplerGetNativeHandle +piextSamplerCreateWithNativeHandle piTearDown piextBindlessImageSamplerCreate piextCommandBufferAdviseUSM diff --git a/sycl/test/abi/pi_opencl_symbol_check.dump b/sycl/test/abi/pi_opencl_symbol_check.dump index b2c3e857e049a..ffbe7248b2be1 100644 --- a/sycl/test/abi/pi_opencl_symbol_check.dump +++ b/sycl/test/abi/pi_opencl_symbol_check.dump @@ -80,6 +80,8 @@ piSamplerCreate piSamplerGetInfo piSamplerRelease piSamplerRetain +piextSamplerGetNativeHandle +piextSamplerCreateWithNativeHandle piTearDown piextCommandBufferAdviseUSM piextBindlessImageSamplerCreate diff --git a/sycl/unittests/assert/assert.cpp b/sycl/unittests/assert/assert.cpp index b94d43925d374..63d0a1c8f7525 100644 --- a/sycl/unittests/assert/assert.cpp +++ b/sycl/unittests/assert/assert.cpp @@ -217,19 +217,6 @@ static pi_result redefinedEventsWaitPositive(pi_uint32 num_events, return PI_SUCCESS; } -static pi_result redefinedEventsWaitNegative(pi_uint32 num_events, - const pi_event *event_list) { - // For negative tests we do not expect the copier kernel to be used, so - // instead we accept whatever amount we get. - // This output here is to reduce amount of time requried to debug/reproduce - // a failing test upon feature break - printf("Waiting for %i events ", num_events); - for (size_t I = 0; I < num_events; ++I) - printf("%i, ", reinterpret_cast(event_list[I])[0]); - printf("\n"); - return PI_SUCCESS; -} - static pi_result redefinedEnqueueMemBufferMapAfter( pi_queue command_queue, pi_mem buffer, pi_bool blocking_map, pi_map_flags map_flags, size_t offset, size_t size, @@ -256,153 +243,6 @@ static void setupMock(sycl::unittest::PiMock &Mock) { Mock.redefineBefore(redefinedEventsWaitPositive); } -namespace TestInteropKernel { -const sycl::context *Context = nullptr; -const sycl::device *Device = nullptr; -int KernelLaunchCounter = ::KernelLaunchCounterBase; - -static pi_result redefinedKernelGetInfo(pi_kernel Kernel, - pi_kernel_info ParamName, - size_t ParamValueSize, void *ParamValue, - size_t *ParamValueSizeRet) { - if (PI_KERNEL_INFO_CONTEXT == ParamName) { - pi_context PiContext = - sycl::detail::getSyclObjImpl(*Context)->getHandleRef(); - - if (ParamValue) - memcpy(ParamValue, &PiContext, sizeof(PiContext)); - if (ParamValueSizeRet) - *ParamValueSizeRet = sizeof(PiContext); - - return PI_SUCCESS; - } - - if (PI_KERNEL_INFO_PROGRAM == ParamName) { - pi_program PIProgram = nullptr; - pi_result Res = mock_piProgramCreate(/*pi_context=*/0x0, /**il*/ nullptr, - /*length=*/0, &PIProgram); - EXPECT_TRUE(PI_SUCCESS == Res); - - if (ParamValue) - memcpy(ParamValue, &PIProgram, sizeof(PIProgram)); - if (ParamValueSizeRet) - *ParamValueSizeRet = sizeof(PIProgram); - - return PI_SUCCESS; - } - - if (PI_KERNEL_INFO_FUNCTION_NAME == ParamName) { - static const char FName[] = "TestFnName"; - if (ParamValue) { - size_t L = strlen(FName) + 1; - if (L < ParamValueSize) - L = ParamValueSize; - - memcpy(ParamValue, FName, L); - } - if (ParamValueSizeRet) - *ParamValueSizeRet = strlen(FName) + 1; - - return PI_SUCCESS; - } - - return PI_ERROR_UNKNOWN; -} - -static pi_result redefinedEnqueueKernelLaunch(pi_queue, pi_kernel, pi_uint32, - const size_t *, const size_t *, - const size_t *LocalSize, - pi_uint32 N, const pi_event *Deps, - pi_event *RetEvent) { - int Val = KernelLaunchCounter++; - // This output here is to reduce amount of time requried to debug/reproduce a - // failing test upon feature break - printf("Enqueued %i\n", Val); - - return PI_SUCCESS; -} - -static pi_result redefinedProgramGetInfo(pi_program P, - pi_program_info ParamName, - size_t ParamValueSize, - void *ParamValue, - size_t *ParamValueSizeRet) { - if (PI_PROGRAM_INFO_NUM_DEVICES == ParamName) { - static const int V = 1; - - if (ParamValue) - memcpy(ParamValue, &V, sizeof(V)); - if (ParamValueSizeRet) - *ParamValueSizeRet = sizeof(V); - - return PI_SUCCESS; - } - - if (PI_PROGRAM_INFO_DEVICES == ParamName) { - EXPECT_EQ(ParamValueSize, 1 * sizeof(pi_device)); - - pi_device Dev = sycl::detail::getSyclObjImpl(*Device)->getHandleRef(); - - if (ParamValue) - memcpy(ParamValue, &Dev, sizeof(Dev)); - if (ParamValueSizeRet) - *ParamValueSizeRet = sizeof(Dev); - - return PI_SUCCESS; - } - - return PI_ERROR_UNKNOWN; -} - -static pi_result redefinedProgramGetBuildInfo(pi_program P, pi_device D, - pi_program_build_info ParamName, - size_t ParamValueSize, - void *ParamValue, - size_t *ParamValueSizeRet) { - if (PI_PROGRAM_BUILD_INFO_BINARY_TYPE == ParamName) { - static const pi_program_binary_type T = PI_PROGRAM_BINARY_TYPE_EXECUTABLE; - if (ParamValue) - memcpy(ParamValue, &T, sizeof(T)); - if (ParamValueSizeRet) - *ParamValueSizeRet = sizeof(T); - return PI_SUCCESS; - } - - if (PI_PROGRAM_BUILD_INFO_OPTIONS == ParamName) { - if (ParamValueSizeRet) - *ParamValueSizeRet = 0; - return PI_SUCCESS; - } - - return PI_ERROR_UNKNOWN; -} - -} // namespace TestInteropKernel - -static void setupMockForInterop(sycl::unittest::PiMock &Mock, - const sycl::context &Ctx, - const sycl::device &Dev) { - using namespace sycl::detail; - - TestInteropKernel::KernelLaunchCounter = ::KernelLaunchCounterBase; - TestInteropKernel::Device = &Dev; - TestInteropKernel::Context = &Ctx; - - Mock.redefineAfter( - redefinedKernelGetGroupInfoAfter); - Mock.redefineBefore( - TestInteropKernel::redefinedEnqueueKernelLaunch); - Mock.redefineAfter( - redefinedEnqueueMemBufferMapAfter); - Mock.redefineBefore(redefinedEventsWaitNegative); - Mock.redefineBefore( - TestInteropKernel::redefinedKernelGetInfo); - Mock.redefineBefore( - TestInteropKernel::redefinedProgramGetInfo); - Mock.redefineBefore( - TestInteropKernel::redefinedProgramGetBuildInfo); -} - #ifndef _WIN32 void ChildProcess(int StdErrFD) { static constexpr int StandardStdErrFD = 2; @@ -528,54 +368,3 @@ TEST(Assert, TestAssertServiceKernelHidden) { EXPECT_TRUE(NoFoundServiceKernelID); } - -TEST(Assert, TestInteropKernelNegative) { - sycl::unittest::PiMock Mock; - sycl::platform Plt = Mock.getPlatform(); - - const sycl::device Dev = Plt.get_devices()[0]; - sycl::context Ctx{Dev}; - - setupMockForInterop(Mock, Ctx, Dev); - - sycl::queue Queue{Ctx, Dev}; - - pi_kernel PIKernel = nullptr; - - pi_result Res = mock_piKernelCreate( - /*pi_program=*/0x0, /*kernel_name=*/"dummy_kernel", &PIKernel); - EXPECT_TRUE(PI_SUCCESS == Res); - - // TODO use make_kernel. This requires a fix in backend.cpp to get plugin - // from context instead of free getPlugin to alllow for mocking of its methods - sycl::kernel KInterop((cl_kernel)PIKernel, Ctx); - - Queue.submit([&](sycl::handler &H) { H.single_task(KInterop); }); - - EXPECT_EQ(TestInteropKernel::KernelLaunchCounter, - KernelLaunchCounterBase + 1); -} - -TEST(Assert, TestInteropKernelFromProgramNegative) { - sycl::unittest::PiMock Mock; - sycl::platform Plt = Mock.getPlatform(); - - const sycl::device Dev = Plt.get_devices()[0]; - sycl::context Ctx{Dev}; - - setupMockForInterop(Mock, Ctx, Dev); - - sycl::queue Queue{Ctx, Dev}; - - sycl::kernel_bundle Bundle = - sycl::get_kernel_bundle(Ctx); - sycl::kernel KOrig = Bundle.get_kernel(sycl::get_kernel_id()); - - cl_kernel CLKernel = sycl::get_native(KOrig); - sycl::kernel KInterop{CLKernel, Ctx}; - - Queue.submit([&](sycl::handler &H) { H.single_task(KInterop); }); - - EXPECT_EQ(TestInteropKernel::KernelLaunchCounter, - KernelLaunchCounterBase + 1); -} diff --git a/sycl/unittests/handler/CMakeLists.txt b/sycl/unittests/handler/CMakeLists.txt index eb7fc559ab73c..3f90404ab35b9 100644 --- a/sycl/unittests/handler/CMakeLists.txt +++ b/sycl/unittests/handler/CMakeLists.txt @@ -1,4 +1,3 @@ add_sycl_unittest(HandlerTests OBJECT - SetArgForLocalAccessor.cpp require.cpp ) diff --git a/sycl/unittests/handler/SetArgForLocalAccessor.cpp b/sycl/unittests/handler/SetArgForLocalAccessor.cpp deleted file mode 100644 index ab7d8f387b761..0000000000000 --- a/sycl/unittests/handler/SetArgForLocalAccessor.cpp +++ /dev/null @@ -1,53 +0,0 @@ -//==------- SetArgForLocalAccessor.cpp --- Handler unit tests --------------==// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include -#include -#include - -#include - -// This test checks that we pass the correct buffer size value when setting -// local_accessor as an argument through handler::set_arg to a kernel created -// using OpenCL interoperability methods. - -namespace { - -size_t LocalBufferArgSize = 0; - -pi_result redefined_piKernelSetArg(pi_kernel kernel, pi_uint32 arg_index, - size_t arg_size, const void *arg_value) { - LocalBufferArgSize = arg_size; - - return PI_SUCCESS; -} - -TEST(HandlerSetArg, LocalAccessor) { - sycl::unittest::PiMock Mock; - redefineMockForKernelInterop(Mock); - Mock.redefine( - redefined_piKernelSetArg); - - constexpr size_t Size = 128; - sycl::queue Q; - - DummyHandleT handle; - auto KernelCL = reinterpret_cast::template input_type>(&handle); - auto Kernel = - sycl::make_kernel(KernelCL, Q.get_context()); - - Q.submit([&](sycl::handler &CGH) { - sycl::local_accessor Acc(Size, CGH); - CGH.set_arg(0, Acc); - CGH.single_task(Kernel); - }).wait(); - - ASSERT_EQ(LocalBufferArgSize, Size * sizeof(float)); -} -} // namespace diff --git a/sycl/unittests/helpers/PiMockPlugin.hpp b/sycl/unittests/helpers/PiMockPlugin.hpp index 31eac5598f588..9469e04d7fa95 100644 --- a/sycl/unittests/helpers/PiMockPlugin.hpp +++ b/sycl/unittests/helpers/PiMockPlugin.hpp @@ -958,6 +958,21 @@ inline pi_result mock_piSamplerRelease(pi_sampler sampler) { return PI_SUCCESS; } +inline pi_result +mock_piextSamplerGetNativeHandle(pi_sampler sampler, + pi_native_handle *nativeHandle) { + *nativeHandle = reinterpret_cast(sampler); + return PI_SUCCESS; +} + +inline pi_result mock_piextSamplerCreateWithNativeHandle( + pi_native_handle nativeHandle, pi_context context, + const bool ownNativeHandle, pi_sampler *sampler) { + *sampler = reinterpret_cast(nativeHandle); + retainDummyHandle(*sampler); + return PI_SUCCESS; +} + // // Queue Commands // diff --git a/sycl/unittests/thread_safety/CMakeLists.txt b/sycl/unittests/thread_safety/CMakeLists.txt index 8b725af8b4dd4..78dc6f2190178 100644 --- a/sycl/unittests/thread_safety/CMakeLists.txt +++ b/sycl/unittests/thread_safety/CMakeLists.txt @@ -1,4 +1,3 @@ add_sycl_unittest(ThreadSafetyTests OBJECT HostAccessorDeadLock.cpp - InteropKernelEnqueue.cpp ) diff --git a/sycl/unittests/thread_safety/InteropKernelEnqueue.cpp b/sycl/unittests/thread_safety/InteropKernelEnqueue.cpp deleted file mode 100644 index 79f19504abea1..0000000000000 --- a/sycl/unittests/thread_safety/InteropKernelEnqueue.cpp +++ /dev/null @@ -1,65 +0,0 @@ -//==-------- InteropKernelEnqueue.cpp --- Thread safety unit tests ---------==// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include -#include -#include -#include -#include - -#include "ThreadUtils.h" - -namespace { -using namespace sycl; - -constexpr std::size_t NArgs = 16; -constexpr std::size_t ThreadCount = 4; -constexpr std::size_t LaunchCount = 8; - -pi_uint32 LastArgSet = -1; -std::size_t LastThread = -1; -pi_result redefined_piKernelSetArg(pi_kernel kernel, pi_uint32 arg_index, - size_t arg_size, const void *arg_value) { - EXPECT_EQ((LastArgSet + 1) % NArgs, arg_index); - LastArgSet = arg_index; - std::size_t ArgValue = *static_cast(arg_value); - if (arg_index == 0) - LastThread = ArgValue; - else - EXPECT_EQ(LastThread, ArgValue); - return PI_SUCCESS; -} - -TEST(KernelEnqueue, InteropKernel) { - unittest::PiMock Mock; - redefineMockForKernelInterop(Mock); - Mock.redefine( - redefined_piKernelSetArg); - - platform Plt = Mock.getPlatform(); - queue Q; - - DummyHandleT Handle; - auto KernelCL = reinterpret_cast::template input_type>(&Handle); - auto Kernel = - sycl::make_kernel(KernelCL, Q.get_context()); - - auto TestLambda = [&](std::size_t ThreadId) { - Q.submit([&](sycl::handler &CGH) { - for (std::size_t I = 0; I < NArgs; ++I) - CGH.set_arg(I, ThreadId); - CGH.single_task(Kernel); - }).wait(); - }; - - for (std::size_t I = 0; I < LaunchCount; ++I) { - ThreadPool Pool(ThreadCount, TestLambda); - } -} -} // namespace