diff --git a/sycl/include/sycl/handler.hpp b/sycl/include/sycl/handler.hpp index 754f24a8138a6..84834a14b86ad 100644 --- a/sycl/include/sycl/handler.hpp +++ b/sycl/include/sycl/handler.hpp @@ -421,7 +421,6 @@ template bool range_size_fits_in_size_t(const range &r) { /// \ingroup sycl_api class __SYCL_EXPORT handler { private: -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES /// Constructs SYCL handler from the pre-constructed stack-allocated /// `handler_impl` (not enforced, but meaningless to do a heap allocation /// outside handler instance). @@ -431,39 +430,6 @@ class __SYCL_EXPORT handler { // Can't provide this overload outside preview because `handler` lacks // required data members. handler(detail::handler_impl &HandlerImpl); -#else - /// Constructs SYCL handler from queue. - /// - /// \param Queue is a SYCL queue. - /// \param CallerNeedsEvent indicates if the event resulting from this handler - /// is needed by the caller. - handler(std::shared_ptr Queue, bool CallerNeedsEvent); - /// Constructs SYCL handler from the associated queue and the submission's - /// primary and secondary queue. - /// - /// \param Queue is a SYCL queue. This is equal to either PrimaryQueue or - /// SecondaryQueue. - /// \param PrimaryQueue is the primary SYCL queue of the submission. - /// \param SecondaryQueue is the secondary SYCL queue of the submission. This - /// is null if no secondary queue is associated with the submission. - /// \param CallerNeedsEvent indicates if the event resulting from this handler - /// is needed by the caller. - handler(std::shared_ptr Queue, - std::shared_ptr PrimaryQueue, - std::shared_ptr SecondaryQueue, - bool CallerNeedsEvent); - __SYCL_DLL_LOCAL handler(std::shared_ptr Queue, - detail::queue_impl *SecondaryQueue, - bool CallerNeedsEvent); - - /// Constructs SYCL handler from Graph. - /// - /// The handler will add the command-group as a node to the graph rather than - /// enqueueing it straight away. - /// - /// \param Graph is a SYCL command_graph - handler(std::shared_ptr Graph); -#endif handler(std::unique_ptr &&HandlerImpl); ~handler(); @@ -502,32 +468,9 @@ class __SYCL_EXPORT handler { void setDeviceKernelInfo(kernel &&Kernel); -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - // TODO: Those functions are not used anymore, remove it in the next - // ABI-breaking window. - void extractArgsAndReqsFromLambda( - char *LambdaPtr, - const std::vector &ParamDescs, bool IsESIMD); - void - extractArgsAndReqsFromLambda(char *LambdaPtr, size_t KernelArgsNum, - const detail::kernel_param_desc_t *KernelArgs, - bool IsESIMD); - /// Extracts and prepares kernel arguments from the lambda using information - /// from the built-ins or integration header. - void extractArgsAndReqsFromLambda( - char *LambdaPtr, detail::kernel_param_desc_t (*ParamDescGetter)(int), - size_t NumKernelParams, bool IsESIMD); -#endif /// Extracts and prepares kernel arguments set via set_arg(s). void extractArgsAndReqs(); -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - // TODO: remove in the next ABI-breaking window. - void processArg(void *Ptr, const detail::kernel_param_kind_t &Kind, - const int Size, const size_t Index, size_t &IndexShift, - bool IsKernelCreatedFromSource, bool IsESIMD); -#endif - #ifndef __INTEL_PREVIEW_BREAKING_CHANGES /// \return a string containing name of SYCL kernel. detail::ABINeutralKernelNameStrT getKernelName(); @@ -548,9 +491,7 @@ class __SYCL_EXPORT handler { /// Saves the location of user's code passed in \p CodeLoc for future usage in /// finalize() method. -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - void saveCodeLoc(detail::code_location CodeLoc); -#endif + void saveCodeLoc(detail::code_location CodeLoc, bool IsTopCodeLoc); void copyCodeLoc(const handler &other); @@ -563,11 +504,7 @@ class __SYCL_EXPORT handler { /// /// Note: in preview mode, handler.finalize() is expected to return /// nullptr if the event is not needed (discarded). -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES detail::EventImplPtr finalize(); -#else - event finalize(); -#endif /// Constructs CG object of specific type, passes it to Scheduler and /// returns sycl::event object representing the command group. @@ -716,14 +653,10 @@ class __SYCL_EXPORT handler { // Set the arg in the handler as normal setArgHelper(ArgIndex, std::move(ArgValue)); -// Register the dynamic parameter with the handler for later association -// with the node being added -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES + // Register the dynamic parameter with the handler for later association + // with the node being added registerDynamicParameter(detail::getSyclObjImpl(DynamicParam).get(), ArgIndex); -#else - registerDynamicParameter(DynamicParam, ArgIndex); -#endif } template @@ -775,14 +708,6 @@ class __SYCL_EXPORT handler { Arg.MArgSize, ArgIndex); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - // TODO: Remove in the next ABI-breaking window. - void registerDynamicParameter( - ext::oneapi::experimental::detail::dynamic_parameter_base - &DynamicParamBase, - int ArgIndex); -#endif - /// Registers a dynamic parameter with the handler for later association with /// the node being created. /// @param DynamicParamImpl The dynamic parameter impl object. @@ -802,11 +727,6 @@ class __SYCL_EXPORT handler { /// kernel bundle contains. void verifyUsedKernelBundleInternal(detail::string_view KernelName); - // TODO: Legacy symbol, remove when ABI breaking is allowed. - void verifyUsedKernelBundle(const std::string &KernelName) { - verifyUsedKernelBundleInternal(detail::string_view{KernelName}); - } - /// Stores lambda to the template-free object /// /// Also initializes the kernel name and prepares for arguments to @@ -877,51 +797,6 @@ class __SYCL_EXPORT handler { } } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - void verifyDeviceHasProgressGuarantee( - sycl::ext::oneapi::experimental::forward_progress_guarantee guarantee, - sycl::ext::oneapi::experimental::execution_scope threadScope, - sycl::ext::oneapi::experimental::execution_scope coordinationScope); - - template - void checkAndSetClusterRange(const Properties &Props) { - namespace syclex = sycl::ext::oneapi::experimental; - constexpr std::size_t ClusterDim = - syclex::detail::getClusterDim(); - if constexpr (ClusterDim > 0) { - auto ClusterSize = Props - .template get_property< - syclex::cuda::cluster_size_key>() - .get_cluster_size(); - setKernelClusterLaunch(ClusterSize); - } - } - - /// Process runtime kernel properties. - /// - /// Stores information about kernel properties into the handler. - template - void processLaunchProperties(PropertiesT Props) { - SetKernelLaunchpropertiesIfNotEmpty(detail::extractKernelProperties(Props)); - } - - /// Process kernel properties. - /// - /// Stores information about kernel properties into the handler. - /// - /// Note: it is important that this function *does not* depend on kernel - /// name or kernel type, because then it will be instantiated for every - /// kernel, even though body of those instantiated functions could be almost - /// the same, thus unnecessary increasing compilation time. - template < - bool IsESIMDKernel, - typename PropertiesT = ext::oneapi::experimental::empty_properties_t> - void processProperties(PropertiesT Props) { - SetKernelLaunchpropertiesIfNotEmpty( - detail::extractKernelProperties(Props)); - } -#endif // INTEL_PREVIEW_BREAKING_CHANGES - /// Checks whether it is possible to copy the source shape to the destination /// shape(the shapes are described by the accessor ranges) by using /// copying by regions of memory and not copying element by element @@ -1436,11 +1311,6 @@ class __SYCL_EXPORT handler { void setStateSpecConstSet(); bool isStateExplicitKernelBundle() const; -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - std::shared_ptr - getOrInsertHandlerKernelBundle(bool Insert) const; -#endif - #ifdef __INTEL_PREVIEW_BREAKING_CHANGES // Rename to just getOrInsertHandlerKernelBundle #endif @@ -1449,11 +1319,6 @@ class __SYCL_EXPORT handler { void setHandlerKernelBundle(kernel Kernel); -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - void setHandlerKernelBundle( - const std::shared_ptr &NewKernelBundleImpPtr); -#endif - template void setHandlerKernelBundle(SharedPtrT &&NewKernelBundleImpPtr); @@ -3140,15 +3005,8 @@ class __SYCL_EXPORT handler { uint64_t SignalValue); private: -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES std::unique_ptr implOwner; detail::handler_impl *impl; -#else - std::shared_ptr impl; - - // Use impl->get_queue*() instead: - std::shared_ptr MQueueDoNotUse; -#endif std::vector MLocalAccStorage; std::vector> MStreamStorage; detail::ABINeutralKernelNameStrT MKernelName; @@ -3166,11 +3024,6 @@ class __SYCL_EXPORT handler { std::unique_ptr MHostKernel; detail::code_location MCodeLoc = {}; -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - // Was used for the previous reduction implementation (via `withAuxHandler`). - bool MIsFinalizedDoNotUse = false; - event MLastEventDoNotUse; -#endif // Make queue_impl class friend to be able to call finalize method. friend class detail::queue_impl; @@ -3285,9 +3138,6 @@ class __SYCL_EXPORT handler { UserRange, KernelFunc}; } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - const std::shared_ptr &getContextImplPtr() const; -#endif detail::context_impl &getContextImpl() const; // Checks if 2D memory operations are supported by the underlying platform. @@ -3447,30 +3297,6 @@ class __SYCL_EXPORT handler { bool IsDeviceImageScoped, size_t NumBytes, size_t Offset); -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - // Modeled after ur_kernel_cache_config_t - // Used as an argument to setKernelCacheConfig that's part of the ABI. - enum class StableKernelCacheConfig : int32_t { - Default = 0, - LargeSLM = 1, - LargeData = 2 - }; - - // Set value of the gpu cache configuration for the kernel. - void setKernelCacheConfig(StableKernelCacheConfig); - // Set value of the kernel is cooperative flag - void setKernelIsCooperative(bool); - - // Set using cuda thread block cluster launch flag and set the launch bounds. - void setKernelClusterLaunch(sycl::range<3> ClusterSize, int Dims); - void setKernelClusterLaunch(sycl::range<3> ClusterSize); - void setKernelClusterLaunch(sycl::range<2> ClusterSize); - void setKernelClusterLaunch(sycl::range<1> ClusterSize); - - // Set the request work group memory size (work_group_static ext). - void setKernelWorkGroupMem(size_t Size); -#endif - void setKernelLaunchProperties( const detail::KernelPropertyHolderStructTy &KernelLaunchProperties); @@ -3520,18 +3346,6 @@ class __SYCL_EXPORT handler { } } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - // Exported on Windows for some reason, have to keep for backward ABI - // compatibility, at least formally. - void throwOnKernelParameterMisuseHelper( - int N, detail::kernel_param_desc_t (*f)(int)) const { - detail::CompileTimeKernelInfoTy Info{}; - Info.NumParams = N; - Info.ParamDescGetter = f; - throwOnKernelParameterMisuse(Info); - } -#endif - template N, bool SetNumWorkGroups, - int Dims); - void setNDRangeDescriptorPadded(sycl::range<3> NumWorkItems, - sycl::id<3> Offset, int Dims); - void setNDRangeDescriptorPadded(sycl::range<3> NumWorkItems, - sycl::range<3> LocalSize, sycl::id<3> Offset, - int Dims); -#endif - template void setNDRangeDescriptor(sycl::range N, bool SetNumWorkGroups = false) { @@ -3629,22 +3425,14 @@ class __SYCL_EXPORT handler { void setNDRangeDescriptor(sycl::range<1> NumWorkItems, sycl::id<1> Offset); void setNDRangeDescriptor(sycl::range<1> NumWorkItems, sycl::range<1> LocalSize, sycl::id<1> Offset); -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - void setKernelInfo(void *KernelFuncPtr, int KernelNumArgs, - detail::kernel_param_desc_t (*KernelParamDescGetter)(int), - bool KernelIsESIMD, bool KernelHasSpecialCaptures); -#endif void setKernelFunc(void *KernelFuncPtr); void instantiateKernelOnHost(void *InstantiateKernelOnHostPtr); friend class detail::HandlerAccess; -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES __SYCL_DLL_LOCAL detail::handler_impl *get_impl() { return impl; } -#else - __SYCL_DLL_LOCAL detail::handler_impl *get_impl() { return impl.get(); } -#endif + // Friend free-functions for asynchronous allocation and freeing. __SYCL_EXPORT friend void ext::oneapi::experimental::async_free(sycl::handler &h, void *ptr); @@ -3657,10 +3445,6 @@ class __SYCL_EXPORT handler { sycl::handler &h, size_t size, const ext::oneapi::experimental::memory_pool &pool); -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - void setKernelNameBasedCachePtr( - detail::KernelNameBasedCacheT *KernelNameBasedCachePtr); -#endif void setDeviceKernelInfoPtr(detail::DeviceKernelInfo *DeviceKernelInfoPtr); queue getQueue(); @@ -3686,13 +3470,8 @@ class HandlerAccess { } static void swap(handler &LHS, handler &RHS) { -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES std::swap(LHS.implOwner, RHS.implOwner); -#endif std::swap(LHS.impl, RHS.impl); -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - std::swap(LHS.MQueueDoNotUse, RHS.MQueueDoNotUse); -#endif std::swap(LHS.MLocalAccStorage, RHS.MLocalAccStorage); std::swap(LHS.MStreamStorage, RHS.MStreamStorage); std::swap(LHS.MKernelName, RHS.MKernelName); @@ -3703,10 +3482,6 @@ class HandlerAccess { std::swap(LHS.MPattern, RHS.MPattern); std::swap(LHS.MHostKernel, RHS.MHostKernel); std::swap(LHS.MCodeLoc, RHS.MCodeLoc); -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - std::swap(LHS.MIsFinalizedDoNotUse, RHS.MIsFinalizedDoNotUse); - std::swap(LHS.MLastEventDoNotUse, RHS.MLastEventDoNotUse); -#endif } // pre/postProcess are used only for reductions right now, but the diff --git a/sycl/source/detail/graph/dynamic_impl.cpp b/sycl/source/detail/graph/dynamic_impl.cpp index c04d424b84dbf..c8efe995aa7e5 100644 --- a/sycl/source/detail/graph/dynamic_impl.cpp +++ b/sycl/source/detail/graph/dynamic_impl.cpp @@ -309,12 +309,8 @@ void dynamic_command_group_impl::finalizeCGFList( const auto &CGF = CGFList[CGFIndex]; // Handler defined inside the loop so it doesn't appear to the runtime // as a single command-group with multiple commands inside. -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES detail::handler_impl HandlerImpl{*MGraph}; sycl::handler Handler{HandlerImpl}; -#else - sycl::handler Handler{MGraph}; -#endif CGF(Handler); if (Handler.getType() != sycl::detail::CGType::Kernel && diff --git a/sycl/source/detail/graph/graph_impl.cpp b/sycl/source/detail/graph/graph_impl.cpp index 28b66d573fa61..c84c81945fd14 100644 --- a/sycl/source/detail/graph/graph_impl.cpp +++ b/sycl/source/detail/graph/graph_impl.cpp @@ -416,12 +416,8 @@ node_impl &graph_impl::add(std::function CGF, const std::vector &Args, nodes_range Deps) { (void)Args; -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES detail::handler_impl HandlerImpl{*this}; sycl::handler Handler{HandlerImpl}; -#else - sycl::handler Handler{shared_from_this()}; -#endif // Pass the node deps to the handler so they are available when processing the // CGF, need for async_malloc nodes. diff --git a/sycl/source/detail/handler_impl.hpp b/sycl/source/detail/handler_impl.hpp index d18dc7236790b..6158f56698e40 100644 --- a/sycl/source/detail/handler_impl.hpp +++ b/sycl/source/detail/handler_impl.hpp @@ -224,15 +224,6 @@ class handler_impl { // Allocation ptr to be freed asynchronously. void *MFreePtr = nullptr; -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - // TODO: remove in the next ABI-breaking window - // Today they are used only in the handler::setKernelNameBasedCachePtr - int MKernelNumArgs = 0; - detail::kernel_param_desc_t (*MKernelParamDescGetter)(int) = nullptr; - bool MKernelIsESIMD = false; - bool MKernelHasSpecialCaptures = true; -#endif - KernelData MKernelData; }; diff --git a/sycl/source/detail/queue_impl.cpp b/sycl/source/detail/queue_impl.cpp index 4a15e767bde1a..72110faaa544e 100644 --- a/sycl/source/detail/queue_impl.cpp +++ b/sycl/source/detail/queue_impl.cpp @@ -289,12 +289,8 @@ queue_impl::submit_impl(const detail::type_erased_cgfo_ty &CGF, bool CallerNeedsEvent, const detail::code_location &Loc, bool IsTopCodeLoc, const v1::SubmissionInfo &SubmitInfo) { -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES detail::handler_impl HandlerImplVal(*this, CallerNeedsEvent); handler Handler(HandlerImplVal); -#else - handler Handler(shared_from_this(), CallerNeedsEvent); -#endif #ifdef XPTI_ENABLE_INSTRUMENTATION if (xptiTraceEnabled()) { diff --git a/sycl/source/detail/queue_impl.hpp b/sycl/source/detail/queue_impl.hpp index 953ee0b60c897..30bdf5a7bb8ab 100644 --- a/sycl/source/detail/queue_impl.hpp +++ b/sycl/source/detail/queue_impl.hpp @@ -729,18 +729,11 @@ class queue_impl : public std::enable_shared_from_this { Handler.depends_on(*ExternalEvent); } -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES inline const detail::EventImplPtr & parseEvent(const detail::EventImplPtr &Event) { assert(!Event || !Event->isDiscarded()); return Event; } -#else - inline detail::EventImplPtr parseEvent(const event &Event) { - const detail::EventImplPtr &EventImpl = getSyclObjImpl(Event); - return EventImpl->isDiscarded() ? nullptr : EventImpl; - } -#endif bool trySwitchingToNoEventsMode() { if (MNoLastEventMode.load(std::memory_order_relaxed)) diff --git a/sycl/source/handler.cpp b/sycl/source/handler.cpp index 8fd04c79cf224..82283b1e92bee 100644 --- a/sycl/source/handler.cpp +++ b/sycl/source/handler.cpp @@ -368,38 +368,10 @@ fill_copy_args(detail::handler_impl *impl, } // namespace detail -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES handler::handler(detail::handler_impl &HandlerImpl) : impl(&HandlerImpl) {} handler::handler(std::unique_ptr &&HandlerImpl) : implOwner(std::move(HandlerImpl)), impl(implOwner.get()) {} -#else -handler::handler(std::unique_ptr &&HandlerImpl) - : impl(std::move(HandlerImpl)) {} - -handler::handler(std::shared_ptr Queue, - bool CallerNeedsEvent) - : impl(std::make_shared(*Queue, CallerNeedsEvent)), - MQueueDoNotUse(std::move(Queue)) {} - -handler::handler( - std::shared_ptr Queue, - [[maybe_unused]] std::shared_ptr PrimaryQueue, - [[maybe_unused]] std::shared_ptr SecondaryQueue, - bool CallerNeedsEvent) - : impl(std::make_shared(*Queue, CallerNeedsEvent)), - MQueueDoNotUse(Queue) {} - -handler::handler(std::shared_ptr Queue, - [[maybe_unused]] detail::queue_impl *SecondaryQueue, - bool CallerNeedsEvent) - : impl(std::make_shared(*Queue, CallerNeedsEvent)), - MQueueDoNotUse(std::move(Queue)) {} - -handler::handler( - std::shared_ptr Graph) - : impl(std::make_shared(*Graph)) {} -#endif handler::~handler() = default; // Sets the submission state to indicate that an explicit kernel bundle has been @@ -420,25 +392,6 @@ bool handler::isStateExplicitKernelBundle() const { return impl->isStateExplicitKernelBundle(); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -// Returns a shared_ptr to the kernel_bundle. -// If there is no kernel_bundle created: -// returns newly created kernel_bundle if Insert is true -// returns shared_ptr(nullptr) if Insert is false -std::shared_ptr -handler::getOrInsertHandlerKernelBundle(bool Insert) const { - if (impl->MKernelBundle || !Insert) - return impl->MKernelBundle; - - context Ctx = detail::createSyclObjFromImpl(impl->get_context()); - impl->MKernelBundle = - detail::getSyclObjImpl(get_kernel_bundle( - Ctx, {detail::createSyclObjFromImpl(impl->get_device())}, - {})); - return impl->MKernelBundle; -} -#endif - // Returns a ptr to the kernel_bundle. // If there is no kernel_bundle created: // returns newly created kernel_bundle if Insert is true @@ -462,13 +415,6 @@ void handler::setHandlerKernelBundle(SharedPtrT &&NewKernelBundleImpPtr) { impl->MKernelBundle = std::forward(NewKernelBundleImpPtr); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -void handler::setHandlerKernelBundle( - const std::shared_ptr &NewKernelBundleImpPtr) { - impl->MKernelBundle = NewKernelBundleImpPtr; -} -#endif - void handler::setHandlerKernelBundle(kernel Kernel) { // Kernel may not have an associated kernel bundle if it is created from a // program. As such, apply getSyclObjImpl directly on the kernel, i.e. not @@ -478,33 +424,7 @@ void handler::setHandlerKernelBundle(kernel Kernel) { setHandlerKernelBundle(std::move(KernelBundleImpl)); } -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES detail::EventImplPtr handler::finalize() { -#else -event handler::finalize() { -#endif -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - // Old reduction implementation, prior to - // https://github.com/intel/llvm/pull/18794 - // https://github.com/intel/llvm/pull/18898 - // https://github.com/intel/llvm/pull/19203 - // relied on explicit calls to handler::finalize and those calls were inlined - // into the user applications. As such, we have to preserve the following - // behavior for ABI-compatibility purposes: - if (MIsFinalizedDoNotUse) - return MLastEventDoNotUse; - - MIsFinalizedDoNotUse = true; - // Use macros to trick clang-format: -#define WRAP_BODY_BEGIN MLastEventDoNotUse = [this]() { -#define WRAP_BODY_END \ - } \ - (); \ - return MLastEventDoNotUse; - - WRAP_BODY_BEGIN -#endif - const auto &type = getType(); detail::queue_impl *Queue = impl->get_queue_or_null(); ext::oneapi::experimental::detail::graph_impl *Graph = @@ -644,13 +564,7 @@ event handler::finalize() { impl->get_queue().submit_kernel_scheduler_bypass( impl->MKernelData, impl->CGData.MEvents, impl->MEventNeeded, MKernel.get(), KernelBundleImpPtr, MCodeLoc, impl->MIsTopCodeLoc); -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES return ResultEvent; -#else - return detail::createSyclObjFromImpl( - ResultEvent ? ResultEvent - : detail::event_impl::create_discarded_event()); -#endif } } @@ -802,14 +716,7 @@ event handler::finalize() { !impl->MExecGraph->containsHostTask(); detail::EventImplPtr GraphCompletionEvent = impl->MExecGraph->enqueue( Queue, std::move(impl->CGData), !DiscardEvent); -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES return GraphCompletionEvent; -#else - return sycl::detail::createSyclObjFromImpl( - GraphCompletionEvent - ? GraphCompletionEvent - : sycl::detail::event_impl::create_discarded_event()); -#endif } } break; case detail::CGType::CopyImage: @@ -857,11 +764,7 @@ event handler::finalize() { if (impl->get_graph_or_null()) { impl->MGraphNodeCG = std::move(CommandGroup); auto EventImpl = detail::event_impl::create_completed_host_event(); -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES return EventImpl; -#else - return detail::createSyclObjFromImpl(EventImpl); -#endif } // Because graph case is handled right above. @@ -870,14 +773,8 @@ event handler::finalize() { // If the queue has an associated graph then we need to take the CG and pass // it to the graph to create a node, rather than submit it to the scheduler. if (auto GraphImpl = Queue->getCommandGraph(); GraphImpl) { - auto EventImpl = Queue->submit_command_to_graph( - *GraphImpl, std::move(CommandGroup), type, impl->MUserFacingNodeType); - -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES - return EventImpl; -#else - return detail::createSyclObjFromImpl(EventImpl); -#endif + return Queue->submit_command_to_graph(*GraphImpl, std::move(CommandGroup), + type, impl->MUserFacingNodeType); } // For kernel submission, regardless of whether an event has been requested, @@ -896,18 +793,7 @@ event handler::finalize() { detail::EventImplPtr Event = detail::Scheduler::getInstance().addCG( std::move(CommandGroup), *Queue, !DiscardEvent); -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES return DiscardEvent ? nullptr : Event; -#else - return detail::createSyclObjFromImpl(Event); -#endif - -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES - WRAP_BODY_END - -#undef WRAP_BODY_BEGIN -#undef WRAP_BODY_END -#endif } void handler::addReduction(const std::shared_ptr &ReduObj) { @@ -957,15 +843,6 @@ void handler::associateWithHandler( static_cast(AccTarget)); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -void handler::processArg(void *Ptr, const detail::kernel_param_kind_t &Kind, - const int Size, const size_t Index, size_t &IndexShift, - bool IsKernelCreatedFromSource, bool IsESIMD) { - impl->MKernelData.processArg(Ptr, Kind, Size, Index, IndexShift, - IsKernelCreatedFromSource, IsESIMD); -} -#endif - void handler::setArgHelper(int ArgIndex, detail::work_group_memory_impl &Arg) { impl->MWorkGroupMemoryObjects.push_back( std::make_shared(Arg)); @@ -992,64 +869,6 @@ void handler::extractArgsAndReqs() { impl->MKernelData.extractArgsAndReqs(MKernel->isCreatedFromSource()); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -// TODO: Those functions are not used anymore, remove it in the next -// ABI-breaking window. -void handler::extractArgsAndReqsFromLambda( - char *LambdaPtr, detail::kernel_param_desc_t (*ParamDescGetter)(int), - size_t NumKernelParams, bool IsESIMD) { - - std::vector ParamDescs; - ParamDescs.reserve(NumKernelParams); - for (size_t i = 0; i < NumKernelParams; i++) { - ParamDescs.push_back(ParamDescGetter(i)); - } - - extractArgsAndReqsFromLambda(LambdaPtr, ParamDescs, IsESIMD); -} - -void handler::extractArgsAndReqsFromLambda( - char *LambdaPtr, const std::vector &ParamDescs, - bool IsESIMD) { - const bool IsKernelCreatedFromSource = false; - size_t IndexShift = 0; - - for (size_t I = 0; I < ParamDescs.size(); ++I) { - void *Ptr = LambdaPtr + ParamDescs[I].offset; - const detail::kernel_param_kind_t &Kind = ParamDescs[I].kind; - const int &Size = ParamDescs[I].info; - if (Kind == detail::kernel_param_kind_t::kind_accessor) { - // For args kind of accessor Size is information about accessor. - // The first 11 bits of Size encodes the accessor target. - const access::target AccTarget = - static_cast(Size & AccessTargetMask); - if ((AccTarget == access::target::device || - AccTarget == access::target::constant_buffer) || - (AccTarget == access::target::image || - AccTarget == access::target::image_array)) { - detail::AccessorBaseHost *AccBase = - static_cast(Ptr); - Ptr = detail::getSyclObjImpl(*AccBase).get(); - } else if (AccTarget == access::target::local) { - detail::LocalAccessorBaseHost *LocalAccBase = - static_cast(Ptr); - Ptr = detail::getSyclObjImpl(*LocalAccBase).get(); - } - } - impl->MKernelData.processArg(Ptr, Kind, Size, I, IndexShift, - IsKernelCreatedFromSource, IsESIMD); - } -} - -void handler::extractArgsAndReqsFromLambda( - char *LambdaPtr, size_t KernelArgsNum, - const detail::kernel_param_desc_t *KernelArgs, bool IsESIMD) { - std::vector ParamDescs( - KernelArgs, KernelArgs + KernelArgsNum); - extractArgsAndReqsFromLambda(LambdaPtr, ParamDescs, IsESIMD); -} -#endif // __INTEL_PREVIEW_BREAKING_CHANGES - #ifndef __INTEL_PREVIEW_BREAKING_CHANGES // Calling methods of kernel_impl requires knowledge of class layout. // As this is impossible in header, there's a function that calls necessary @@ -1703,55 +1522,6 @@ static bool checkContextSupports(detail::context_impl &ContextImpl, return SupportsOp; } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -void handler::verifyDeviceHasProgressGuarantee( - sycl::ext::oneapi::experimental::forward_progress_guarantee guarantee, - sycl::ext::oneapi::experimental::execution_scope threadScope, - sycl::ext::oneapi::experimental::execution_scope coordinationScope) { - - using execution_scope = sycl::ext::oneapi::experimental::execution_scope; - using forward_progress = - sycl::ext::oneapi::experimental::forward_progress_guarantee; - const bool supported = impl->get_device().supportsForwardProgress( - guarantee, threadScope, coordinationScope); - if (threadScope == execution_scope::work_group) { - if (!supported) { - throw sycl::exception( - sycl::errc::feature_not_supported, - "Required progress guarantee for work groups is not " - "supported by this device."); - } - // If we are here, the device supports the guarantee required but there is a - // caveat in that if the guarantee required is a concurrent guarantee, then - // we most likely also need to enable cooperative launch of the kernel. That - // is, although the device supports the required guarantee, some setup work - // is needed to truly make the device provide that guarantee at runtime. - // Otherwise, we will get the default guarantee which is weaker than - // concurrent. Same reasoning applies for sub_group but not for work_item. - // TODO: Further design work is probably needed to reflect this behavior in - // Unified Runtime. - if (guarantee == forward_progress::concurrent) - setKernelIsCooperative(true); - } else if (threadScope == execution_scope::sub_group) { - if (!supported) { - throw sycl::exception(sycl::errc::feature_not_supported, - "Required progress guarantee for sub groups is not " - "supported by this device."); - } - // Same reasoning as above. - if (guarantee == forward_progress::concurrent) - setKernelIsCooperative(true); - } else { // threadScope is execution_scope::work_item otherwise undefined - // behavior - if (!supported) { - throw sycl::exception(sycl::errc::feature_not_supported, - "Required progress guarantee for work items is not " - "supported by this device."); - } - } -} -#endif - bool handler::supportsUSMMemcpy2D() { if (impl->get_graph_or_null()) return true; @@ -1871,16 +1641,6 @@ void handler::setKernelLaunchProperties( impl->get_device() /*device_impl*/); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -const std::shared_ptr & -handler::getContextImplPtr() const { - if (auto *Graph = impl->get_graph_or_null()) { - return Graph->getContextImplPtr(); - } - return impl->get_queue().getContextImplPtr(); -} -#endif - detail::context_impl &handler::getContextImpl() const { if (auto *Graph = impl->get_graph_or_null()) { return Graph->getContextImpl(); @@ -1888,69 +1648,6 @@ detail::context_impl &handler::getContextImpl() const { return impl->get_queue().getContextImpl(); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -void handler::setKernelCacheConfig(handler::StableKernelCacheConfig Config) { - switch (Config) { - case handler::StableKernelCacheConfig::Default: - impl->MKernelData.setKernelCacheConfig(UR_KERNEL_CACHE_CONFIG_DEFAULT); - break; - case handler::StableKernelCacheConfig::LargeSLM: - impl->MKernelData.setKernelCacheConfig(UR_KERNEL_CACHE_CONFIG_LARGE_SLM); - break; - case handler::StableKernelCacheConfig::LargeData: - impl->MKernelData.setKernelCacheConfig(UR_KERNEL_CACHE_CONFIG_LARGE_DATA); - break; - } -} - -void handler::setKernelIsCooperative(bool KernelIsCooperative) { - impl->MKernelData.setCooperative(KernelIsCooperative); -} - -void handler::setKernelClusterLaunch(sycl::range<3> ClusterSize, int Dims) { - throwIfGraphAssociated< - syclex::detail::UnsupportedGraphFeatures:: - sycl_ext_oneapi_experimental_cuda_cluster_launch>(); - - if (Dims == 1) { - sycl::range<1> ClusterSizeTrimmed = {ClusterSize[0]}; - impl->MKernelData.setClusterDimensions(ClusterSizeTrimmed); - } else if (Dims == 2) { - sycl::range<2> ClusterSizeTrimmed = {ClusterSize[0], ClusterSize[1]}; - impl->MKernelData.setClusterDimensions(ClusterSizeTrimmed); - } else if (Dims == 3) { - impl->MKernelData.setClusterDimensions(ClusterSize); - } -} - -void handler::setKernelClusterLaunch(sycl::range<3> ClusterSize) { - throwIfGraphAssociated< - syclex::detail::UnsupportedGraphFeatures:: - sycl_ext_oneapi_experimental_cuda_cluster_launch>(); - impl->MKernelData.setClusterDimensions(ClusterSize); -} - -void handler::setKernelClusterLaunch(sycl::range<2> ClusterSize) { - throwIfGraphAssociated< - syclex::detail::UnsupportedGraphFeatures:: - sycl_ext_oneapi_experimental_cuda_cluster_launch>(); - impl->MKernelData.setClusterDimensions(ClusterSize); -} - -void handler::setKernelClusterLaunch(sycl::range<1> ClusterSize) { - throwIfGraphAssociated< - syclex::detail::UnsupportedGraphFeatures:: - sycl_ext_oneapi_experimental_cuda_cluster_launch>(); - impl->MKernelData.setClusterDimensions(ClusterSize); -} - -void handler::setKernelWorkGroupMem(size_t Size) { - throwIfGraphAssociated(); - impl->MKernelData.setKernelWorkGroupMemorySize(Size); -} -#endif // __INTEL_PREVIEW_BREAKING_CHANGES - void handler::ext_oneapi_graph( ext::oneapi::experimental::command_graph< ext::oneapi::experimental::graph_state::executable> @@ -2001,10 +1698,6 @@ std::tuple, bool> handler::getMaxWorkGroups_v2() { return {std::array{0, 0, 0}, false}; } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -void handler::setNDRangeUsed(bool Value) { (void)Value; } -#endif - void handler::registerDynamicParameter( ext::oneapi::experimental::detail::dynamic_parameter_impl *DynamicParamImpl, int ArgIndex) { @@ -2024,18 +1717,6 @@ void handler::registerDynamicParameter( impl->MKernelData.addDynamicParameter(DynamicParamImpl, ArgIndex); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -// TODO: Remove in the next ABI-breaking window. -void handler::registerDynamicParameter( - ext::oneapi::experimental::detail::dynamic_parameter_base &DynamicParamBase, - int ArgIndex) { - ext::oneapi::experimental::detail::dynamic_parameter_impl *DynParamImpl = - detail::getSyclObjImpl(DynamicParamBase).get(); - - registerDynamicParameter(DynParamImpl, ArgIndex); -} -#endif - bool handler::eventNeeded() const { return impl->MEventNeeded; } void *handler::storeRawArg(const void *Ptr, size_t Size) { @@ -2057,17 +1738,6 @@ void handler::SetHostTask(std::function Func) { setType(detail::CGType::CodeplayHostTask); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -// TODO: This function is not used anymore, remove it in the next -// ABI-breaking window. -void handler::addAccessorReq(detail::AccessorImplPtr Accessor) { - // Add accessor to the list of requirements. - impl->CGData.MRequirements.push_back(Accessor.get()); - // Store copy of the accessor. - impl->CGData.MAccStorage.push_back(std::move(Accessor)); -} -#endif - void handler::addLifetimeSharedPtrStorage(std::shared_ptr SPtr) { impl->CGData.MSharedPtrStorage.push_back(std::move(SPtr)); } @@ -2077,10 +1747,6 @@ void handler::addArg(detail::kernel_param_kind_t ArgKind, void *Req, impl->MKernelData.addArg(ArgKind, Req, AccessTarget, ArgIndex); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -void handler::clearArgs() { impl->MKernelData.clearArgs(); } -#endif - void handler::setArgsToAssociatedAccessors() { impl->MKernelData.setArgs(impl->MAssociatedAccesors); } @@ -2109,56 +1775,6 @@ void handler::setDeviceKernelInfo(kernel &&Kernel) { // `lambdaAndKernelHaveEqualName` calls can handle that. } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -void handler::setNDRangeDescriptorPadded(sycl::range<3> N, - bool SetNumWorkGroups, int Dims) { - if (Dims == 1) { - sycl::range<1> Range = {N[0]}; - impl->MKernelData.setNDRDesc(NDRDescT{Range, SetNumWorkGroups}); - } else if (Dims == 2) { - sycl::range<2> Range = {N[0], N[1]}; - impl->MKernelData.setNDRDesc(NDRDescT{Range, SetNumWorkGroups}); - } else if (Dims == 3) { - impl->MKernelData.setNDRDesc(NDRDescT{N, SetNumWorkGroups}); - } -} - -void handler::setNDRangeDescriptorPadded(sycl::range<3> NumWorkItems, - sycl::id<3> Offset, int Dims) { - if (Dims == 1) { - sycl::range<1> NumWorkItemsTrimmed = {NumWorkItems[0]}; - sycl::id<1> OffsetTrimmed = {Offset[0]}; - impl->MKernelData.setNDRDesc(NDRDescT{NumWorkItemsTrimmed, OffsetTrimmed}); - } else if (Dims == 2) { - sycl::range<2> NumWorkItemsTrimmed = {NumWorkItems[0], NumWorkItems[1]}; - sycl::id<2> OffsetTrimmed = {Offset[0], Offset[1]}; - impl->MKernelData.setNDRDesc(NDRDescT{NumWorkItemsTrimmed, OffsetTrimmed}); - } else if (Dims == 3) { - impl->MKernelData.setNDRDesc(NDRDescT{NumWorkItems, Offset}); - } -} - -void handler::setNDRangeDescriptorPadded(sycl::range<3> NumWorkItems, - sycl::range<3> LocalSize, - sycl::id<3> Offset, int Dims) { - if (Dims == 1) { - sycl::range<1> NumWorkItemsTrimmed = {NumWorkItems[0]}; - sycl::range<1> LocalSizeTrimmed = {LocalSize[0]}; - sycl::id<1> OffsetTrimmed = {Offset[0]}; - impl->MKernelData.setNDRDesc( - NDRDescT{NumWorkItemsTrimmed, LocalSizeTrimmed, OffsetTrimmed}); - } else if (Dims == 2) { - sycl::range<2> NumWorkItemsTrimmed = {NumWorkItems[0], NumWorkItems[1]}; - sycl::range<2> LocalSizeTrimmed = {LocalSize[0], LocalSize[1]}; - sycl::id<2> OffsetTrimmed = {Offset[0], Offset[1]}; - impl->MKernelData.setNDRDesc( - NDRDescT{NumWorkItemsTrimmed, LocalSizeTrimmed, OffsetTrimmed}); - } else if (Dims == 3) { - impl->MKernelData.setNDRDesc(NDRDescT{NumWorkItems, LocalSize, Offset}); - } -} -#endif - void handler::setNDRangeDescriptor(sycl::range<3> N, bool SetNumWorkGroups) { impl->MKernelData.setNDRDesc(NDRDescT{N, SetNumWorkGroups}); } @@ -2198,34 +1814,6 @@ void handler::setNDRangeDescriptor(sycl::range<1> NumWorkItems, impl->MKernelData.setNDRDesc(NDRDescT{NumWorkItems, LocalSize, Offset}); } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -void handler::setKernelNameBasedCachePtr( - sycl::detail::KernelNameBasedCacheT *KernelNameBasedCachePtr) { - assert(!impl->MKernelData.getDeviceKernelInfoPtr() && "Already set!"); - (void)KernelNameBasedCachePtr; - CompileTimeKernelInfoTy HandlerInfo; - HandlerInfo.Name = MKernelName; - HandlerInfo.NumParams = impl->MKernelNumArgs; - HandlerInfo.ParamDescGetter = impl->MKernelParamDescGetter; - HandlerInfo.IsESIMD = impl->MKernelIsESIMD; - HandlerInfo.HasSpecialCaptures = impl->MKernelHasSpecialCaptures; - impl->MKernelData.setDeviceKernelInfoPtr( - &detail::ProgramManager::getInstance().getOrCreateDeviceKernelInfo( - HandlerInfo)); -} - -void handler::setKernelInfo( - void *KernelFuncPtr, int KernelNumArgs, - detail::kernel_param_desc_t (*KernelParamDescGetter)(int), - bool KernelIsESIMD, bool KernelHasSpecialCaptures) { - impl->MKernelData.setKernelFunc(KernelFuncPtr); - impl->MKernelNumArgs = KernelNumArgs; - impl->MKernelParamDescGetter = KernelParamDescGetter; - impl->MKernelIsESIMD = KernelIsESIMD; - impl->MKernelHasSpecialCaptures = KernelHasSpecialCaptures; -} -#endif - void handler::setDeviceKernelInfoPtr( sycl::detail::DeviceKernelInfo *DeviceKernelInfoPtr) { assert(!impl->MKernelData.getDeviceKernelInfoPtr() && "Already set!"); @@ -2246,12 +1834,7 @@ void handler::saveCodeLoc(detail::code_location CodeLoc, bool IsTopCodeLoc) { MCodeLoc = CodeLoc; impl->MIsTopCodeLoc = IsTopCodeLoc; } -#ifndef __INTEL_PREVIEW_BREAKING_CHANGES -void handler::saveCodeLoc(detail::code_location CodeLoc) { - MCodeLoc = CodeLoc; - impl->MIsTopCodeLoc = true; -} -#endif + void handler::copyCodeLoc(const handler &other) { MCodeLoc = other.MCodeLoc; impl->MIsTopCodeLoc = other.impl->MIsTopCodeLoc; @@ -2265,12 +1848,8 @@ __SYCL_EXPORT void HandlerAccess::preProcess(handler &CGH, type_erased_cgfo_ty F) { queue_impl &Q = CGH.impl->get_queue(); bool EventNeeded = !Q.isInOrder(); -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES handler_impl HandlerImpl{Q, EventNeeded}; handler AuxHandler{HandlerImpl}; -#else - handler AuxHandler{Q.shared_from_this(), EventNeeded}; -#endif AuxHandler.copyCodeLoc(CGH); F(AuxHandler); auto E = AuxHandler.finalize(); diff --git a/sycl/test/abi/layout_handler.cpp b/sycl/test/abi/layout_handler.cpp index f312eb4a03485..c95e4bfd08796 100644 --- a/sycl/test/abi/layout_handler.cpp +++ b/sycl/test/abi/layout_handler.cpp @@ -11,73 +11,72 @@ void foo() { } // clang-format off - // The order of field declarations and their types are important. // CHECK: 0 | class sycl::handler -// CHECK-NEXT: 0 | class std::shared_ptr impl -// CHECK-NEXT: 0 | class std::__shared_ptr (base) -// CHECK-NEXT: 0 | class std::__shared_ptr_access (base) (empty) -// CHECK-NEXT: 0 | element_type * _M_ptr -// CHECK-NEXT: 8 | class std::__shared_count<> _M_refcount -// CHECK-NEXT: 8 | _Sp_counted_base<(enum __gnu_cxx::_Lock_policy)2U> * _M_pi -// CHECK-NEXT: 16 | class std::shared_ptr MQueue -// CHECK-NEXT: 16 | class std::__shared_ptr (base) -// CHECK-NEXT: 16 | class std::__shared_ptr_access (base) (empty) -// CHECK-NEXT: 16 | element_type * _M_ptr -// CHECK-NEXT: 24 | class std::__shared_count<> _M_refcount -// CHECK-NEXT: 24 | _Sp_counted_base<(enum __gnu_cxx::_Lock_policy)2U> * _M_pi -// CHECK-NEXT: 32 | class std::vector > MLocalAccStorage -// CHECK-NEXT: 32 | struct std::_Vector_base, class std::allocator > > (base) -// CHECK-NEXT: 32 | struct std::_Vector_base, class std::allocator > >::_Vector_impl _M_impl -// CHECK-NEXT: 32 | class std::allocator > (base) (empty) -// CHECK: 32 | pointer _M_start -// CHECK-NEXT: 40 | pointer _M_finish -// CHECK-NEXT: 48 | pointer _M_end_of_storage -// CHECK-NEXT: 56 | class std::vector > MStreamStorage -// CHECK-NEXT: 56 | struct std::_Vector_base, class std::allocator > > (base) -// CHECK-NEXT: 56 | struct std::_Vector_base, class std::allocator > >::_Vector_impl _M_impl -// CHECK-NEXT: 56 | class std::allocator > (base) (empty) -// CHECK: 56 | pointer _M_start -// CHECK-NEXT: 64 | pointer _M_finish -// CHECK-NEXT: 72 | pointer _M_end_of_storage -// CHECK-NEXT: 80 | class sycl::detail::string MKernelName -// CHECK-NEXT: 80 | char * str -// CHECK-NEXT: 88 | class std::shared_ptr MKernel -// CHECK-NEXT: 88 | class std::__shared_ptr (base) -// CHECK-NEXT: 88 | class std::__shared_ptr_access (base) (empty) -// CHECK-NEXT: 88 | element_type * _M_ptr -// CHECK-NEXT: 96 | class std::__shared_count<> _M_refcount -// CHECK-NEXT: 96 | _Sp_counted_base<(enum __gnu_cxx::_Lock_policy)2U> * _M_pi -// CHECK-NEXT: 104 | void * MSrcPtr -// CHECK-NEXT: 112 | void * MDstPtr -// CHECK-NEXT: 120 | size_t MLength -// CHECK-NEXT: 128 | class std::vector MPattern -// CHECK-NEXT: 128 | struct std::_Vector_base > (base) -// CHECK-NEXT: 128 | struct std::_Vector_base >::_Vector_impl _M_impl -// CHECK-NEXT: 128 | class std::allocator (base) (empty) -// CHECK: 128 | pointer _M_start -// CHECK-NEXT: 136 | pointer _M_finish -// CHECK-NEXT: 144 | pointer _M_end_of_storage -// CHECK-NEXT: 152 | class std::unique_ptr MHostKernel -// CHECK: 152 | class std::__uniq_ptr_impl > -// CHECK-NEXT: 152 | class std::tuple > _M_t -// CHECK-NEXT: 152 | struct std::_Tuple_impl<0, class sycl::detail::HostKernelBase *, struct std::default_delete > (base) -// CHECK-NEXT: 152 | struct std::_Tuple_impl<1, struct std::default_delete > (base) (empty) -// CHECK: 152 | struct std::_Head_base<0, class sycl::detail::HostKernelBase *> (base) -// CHECK-NEXT: 152 | class sycl::detail::HostKernelBase * _M_head_impl -// CHECK-NEXT: 160 | struct sycl::detail::code_location MCodeLoc -// CHECK-NEXT: 160 | const char * MFileName -// CHECK-NEXT: 168 | const char * MFunctionName -// CHECK-NEXT: 176 | unsigned long MLineNo -// CHECK-NEXT: 184 | unsigned long MColumnNo -// CHECK-NEXT: 192 | _Bool MIsFinalized -// CHECK-NEXT: 200 | class sycl::event MLastEvent -// CHECK-NEXT: 200 | class sycl::detail::OwnerLessBase (base) (empty) -// CHECK-NEXT: 200 | class std::shared_ptr impl -// CHECK-NEXT: 200 | class std::__shared_ptr (base) -// CHECK-NEXT: 200 | class std::__shared_ptr_access (base) (empty) -// CHECK-NEXT: 200 | element_type * _M_ptr -// CHECK-NEXT: 208 | class std::__shared_count<> _M_refcount -// CHECK-NEXT: 208 | _Sp_counted_base<(enum __gnu_cxx::_Lock_policy)2U> * _M_pi -// CHECK-NEXT: | [sizeof=216, dsize=216, align=8, -// CHECK-NEXT: | nvsize=216, nvalign=8] +// CHECK-NEXT: 0 | class std::unique_ptr implOwner +// CHECK-NEXT: 0 | struct std::__uniq_ptr_data > _M_t +// CHECK-NEXT: 0 | class std::__uniq_ptr_impl > (base) +// CHECK-NEXT: 0 | class std::tuple > _M_t +// CHECK-NEXT: 0 | struct std::_Tuple_impl<0, class sycl::detail::handler_impl *, struct std::default_delete > (base) +// CHECK-NEXT: 0 | struct std::_Tuple_impl<1, struct std::default_delete > (base) (empty) +// CHECK: 0 | struct std::_Head_base<1, struct std::default_delete > (base) (empty) +// CHECK-NEXT: 0 | struct std::default_delete _M_head_impl (empty) +// CHECK: 0 | struct std::_Head_base<0, class sycl::detail::handler_impl *> (base) +// CHECK-NEXT: 0 | class sycl::detail::handler_impl * _M_head_impl +// CHECK-NEXT: 8 | detail::handler_impl * impl +// CHECK-NEXT: 16 | class std::vector > MLocalAccStorage +// CHECK-NEXT: 16 | struct std::_Vector_base, class std::allocator > > (base) +// CHECK-NEXT: 16 | struct std::_Vector_base, class std::allocator > >::_Vector_impl _M_impl +// CHECK-NEXT: 16 | class std::allocator > (base) (empty) +// CHECK-NEXT: 16 | class std::__new_allocator > (base) (empty) +// CHECK-NEXT: 16 | struct std::_Vector_base, class std::allocator > >::_Vector_impl_data (base) +// CHECK: 16 | pointer _M_start +// CHECK-NEXT: 24 | pointer _M_finish +// CHECK-NEXT: 32 | pointer _M_end_of_storage +// CHECK-NEXT: 40 | class std::vector > MStreamStorage +// CHECK-NEXT: 40 | struct std::_Vector_base, class std::allocator > > (base) +// CHECK-NEXT: 40 | struct std::_Vector_base, class std::allocator > >::_Vector_impl _M_impl +// CHECK-NEXT: 40 | class std::allocator > (base) (empty) +// CHECK-NEXT: 40 | class std::__new_allocator > (base) (empty) +// CHECK-NEXT: 40 | struct std::_Vector_base, class std::allocator > >::_Vector_impl_data (base) +// CHECK: 40 | pointer _M_start +// CHECK-NEXT: 48 | pointer _M_finish +// CHECK-NEXT: 56 | pointer _M_end_of_storage +// CHECK-NEXT: 64 | class sycl::detail::string MKernelName +// CHECK-NEXT: 64 | char * str +// CHECK-NEXT: 72 | class std::shared_ptr MKernel +// CHECK-NEXT: 72 | class std::__shared_ptr (base) +// CHECK-NEXT: 72 | class std::__shared_ptr_access (base) (empty) +// CHECK-NEXT: 72 | element_type * _M_ptr +// CHECK-NEXT: 80 | class std::__shared_count<> _M_refcount +// CHECK-NEXT: 80 | _Sp_counted_base<(enum __gnu_cxx::_Lock_policy)2U> * _M_pi +// CHECK-NEXT: 88 | void * MSrcPtr +// CHECK-NEXT: 96 | void * MDstPtr +// CHECK-NEXT: 104 | size_t MLength +// CHECK-NEXT: 112 | class std::vector MPattern +// CHECK-NEXT: 112 | struct std::_Vector_base > (base) +// CHECK-NEXT: 112 | struct std::_Vector_base >::_Vector_impl _M_impl +// CHECK-NEXT: 112 | class std::allocator (base) (empty) +// CHECK-NEXT: 112 | class std::__new_allocator (base) (empty) +// CHECK-NEXT: 112 | struct std::_Vector_base >::_Vector_impl_data (base) +// CHECK: 112 | pointer _M_start +// CHECK-NEXT: 120 | pointer _M_finish +// CHECK-NEXT: 128 | pointer _M_end_of_storage +// CHECK-NEXT: 136 | class std::unique_ptr MHostKernel +// CHECK-NEXT: 136 | struct std::__uniq_ptr_data > _M_t +// CHECK: 136 | class std::__uniq_ptr_impl > (base) +// CHECK-NEXT: 136 | class std::tuple > _M_t +// CHECK-NEXT: 136 | struct std::_Tuple_impl<0, class sycl::detail::HostKernelBase *, struct std::default_delete > (base) +// CHECK-NEXT: 136 | struct std::_Tuple_impl<1, struct std::default_delete > (base) (empty) +// CHECK: 136 | struct std::_Head_base<1, struct std::default_delete > (base) (empty) +// CHECK-NEXT: 136 | struct std::default_delete _M_head_impl (empty) +// CHECK: 136 | struct std::_Head_base<0, class sycl::detail::HostKernelBase *> (base) +// CHECK-NEXT: 136 | class sycl::detail::HostKernelBase * _M_head_impl +// CHECK-NEXT: 144 | struct sycl::detail::code_location MCodeLoc +// CHECK-NEXT: 144 | const char * MFileName +// CHECK-NEXT: 152 | const char * MFunctionName +// CHECK-NEXT: 160 | unsigned long MLineNo +// CHECK-NEXT: 168 | unsigned long MColumnNo +// CHECK-NEXT: | [sizeof=176, dsize=176, align=8, +// CHECK-NEXT: | nvsize=176, nvalign=8] +// clang-format on diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index 8c89566db341e..06c47ba790763 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3544,11 +3544,9 @@ _ZN4sycl3_V17handler10depends_onERKSt10shared_ptrINS0_6detail10event_implEE _ZN4sycl3_V17handler10depends_onERKSt6vectorINS0_5eventESaIS3_EE _ZN4sycl3_V17handler10depends_onERKSt6vectorISt10shared_ptrINS0_6detail10event_implEESaIS6_EE _ZN4sycl3_V17handler10mem_adviseEPKvmi -_ZN4sycl3_V17handler10processArgEPvRKNS0_6detail19kernel_param_kind_tEimRmbb _ZN4sycl3_V17handler11SetHostTaskESt8functionIFvNS0_14interop_handleEEE _ZN4sycl3_V17handler11SetHostTaskESt8functionIFvvEE _ZN4sycl3_V17handler11copyCodeLocERKS1_ -_ZN4sycl3_V17handler11saveCodeLocENS0_6detail13code_locationE _ZN4sycl3_V17handler11saveCodeLocENS0_6detail13code_locationEb _ZN4sycl3_V17handler11storeRawArgEPKvm _ZN4sycl3_V17handler12addReductionERKSt10shared_ptrIKvE @@ -3556,9 +3554,6 @@ _ZN4sycl3_V17handler12setArgHelperEiONS0_6streamE _ZN4sycl3_V17handler12setArgHelperEiRNS0_6detail22work_group_memory_implE _ZN4sycl3_V17handler13getKernelNameEv _ZN4sycl3_V17handler13setKernelFuncEPv -_ZN4sycl3_V17handler13setKernelInfoEPviPFNS0_6detail19kernel_param_desc_tEiEbb -_ZN4sycl3_V17handler14addAccessorReqESt10shared_ptrINS0_6detail16AccessorImplHostEE -_ZN4sycl3_V17handler14setNDRangeUsedEb _ZN4sycl3_V17handler15ext_oneapi_copyENS0_3ext6oneapi12experimental16image_mem_handleENS0_5rangeILi3EEERKNS4_16image_descriptorEPvS7_S7_S7_ _ZN4sycl3_V17handler15ext_oneapi_copyENS0_3ext6oneapi12experimental16image_mem_handleENS0_5rangeILi3EEERKNS4_16image_descriptorEPvS7_SA_mS7_ _ZN4sycl3_V17handler15ext_oneapi_copyENS0_3ext6oneapi12experimental16image_mem_handleENS0_5rangeILi3EEERKNS4_16image_descriptorES5_S7_SA_S7_ @@ -3589,7 +3584,6 @@ _ZN4sycl3_V17handler20associateWithHandlerEPNS0_6detail16AccessorBaseHostENS0_6a _ZN4sycl3_V17handler20associateWithHandlerEPNS0_6detail28SampledImageAccessorBaseHostENS0_12image_targetE _ZN4sycl3_V17handler20associateWithHandlerEPNS0_6detail30UnsampledImageAccessorBaseHostENS0_12image_targetE _ZN4sycl3_V17handler20memcpyToDeviceGlobalEPKvS3_bmm -_ZN4sycl3_V17handler20setKernelCacheConfigENS1_23StableKernelCacheConfigE _ZN4sycl3_V17handler20setNDRangeDescriptorENS0_5rangeILi1EEENS0_2idILi1EEE _ZN4sycl3_V17handler20setNDRangeDescriptorENS0_5rangeILi1EEES3_NS0_2idILi1EEE _ZN4sycl3_V17handler20setNDRangeDescriptorENS0_5rangeILi1EEEb @@ -3600,43 +3594,27 @@ _ZN4sycl3_V17handler20setNDRangeDescriptorENS0_5rangeILi3EEENS0_2idILi3EEE _ZN4sycl3_V17handler20setNDRangeDescriptorENS0_5rangeILi3EEES3_NS0_2idILi3EEE _ZN4sycl3_V17handler20setNDRangeDescriptorENS0_5rangeILi3EEEb _ZN4sycl3_V17handler20setStateSpecConstSetEv -_ZN4sycl3_V17handler21setKernelWorkGroupMemEm _ZN4sycl3_V17handler21setUserFacingNodeTypeENS0_3ext6oneapi12experimental9node_typeE _ZN4sycl3_V17handler22ext_oneapi_fill2d_implEPvmPKvmmm _ZN4sycl3_V17handler22memcpyFromDeviceGlobalEPvPKvbmm _ZN4sycl3_V17handler22setDeviceKernelInfoPtrEPNS0_6detail16DeviceKernelInfoE _ZN4sycl3_V17handler22setHandlerKernelBundleENS0_6kernelE -_ZN4sycl3_V17handler22setHandlerKernelBundleERKSt10shared_ptrINS0_6detail18kernel_bundle_implEE -_ZN4sycl3_V17handler22setKernelClusterLaunchENS0_5rangeILi1EEE -_ZN4sycl3_V17handler22setKernelClusterLaunchENS0_5rangeILi2EEE -_ZN4sycl3_V17handler22setKernelClusterLaunchENS0_5rangeILi3EEE -_ZN4sycl3_V17handler22setKernelClusterLaunchENS0_5rangeILi3EEEi -_ZN4sycl3_V17handler22setKernelIsCooperativeEb _ZN4sycl3_V17handler23instantiateKernelOnHostEPv _ZN4sycl3_V17handler24GetRangeRoundingSettingsERmS2_S2_ _ZN4sycl3_V17handler24ext_intel_read_host_pipeENS0_6detail11string_viewEPvmb _ZN4sycl3_V17handler24ext_oneapi_memcpy2d_implEPvmPKvmmm _ZN4sycl3_V17handler24ext_oneapi_memset2d_implEPvmimm _ZN4sycl3_V17handler24registerDynamicParameterEPNS0_3ext6oneapi12experimental6detail22dynamic_parameter_implEi -_ZN4sycl3_V17handler24registerDynamicParameterERNS0_3ext6oneapi12experimental6detail22dynamic_parameter_baseEi _ZN4sycl3_V17handler25ext_intel_write_host_pipeENS0_6detail11string_viewEPvmb _ZN4sycl3_V17handler25setKernelLaunchPropertiesERKNS0_6detail27kernel_launch_properties_v111PropsHolderIJNS0_3ext6oneapi12experimental23work_group_scratch_sizeENS5_5intel12experimental12cache_configENS7_17use_root_sync_keyENS7_23work_group_progress_keyENS7_22sub_group_progress_keyENS7_22work_item_progress_keyENS7_4cuda12cluster_sizeILi1EEENSH_ILi2EEENSH_ILi3EEEEEE _ZN4sycl3_V17handler26associateWithHandlerCommonESt10shared_ptrINS0_6detail16AccessorImplHostEEi -_ZN4sycl3_V17handler26setKernelNameBasedCachePtrEPNS0_6detail21KernelNameBasedCacheTE -_ZN4sycl3_V17handler26setNDRangeDescriptorPaddedENS0_5rangeILi3EEENS0_2idILi3EEEi -_ZN4sycl3_V17handler26setNDRangeDescriptorPaddedENS0_5rangeILi3EEES3_NS0_2idILi3EEEi -_ZN4sycl3_V17handler26setNDRangeDescriptorPaddedENS0_5rangeILi3EEEbi _ZN4sycl3_V17handler27addLifetimeSharedPtrStorageESt10shared_ptrIKvE _ZN4sycl3_V17handler27computeFallbackKernelBoundsEmm -_ZN4sycl3_V17handler28extractArgsAndReqsFromLambdaEPcPFNS0_6detail19kernel_param_desc_tEiEmb -_ZN4sycl3_V17handler28extractArgsAndReqsFromLambdaEPcRKSt6vectorINS0_6detail19kernel_param_desc_tESaIS5_EEb -_ZN4sycl3_V17handler28extractArgsAndReqsFromLambdaEPcmPKNS0_6detail19kernel_param_desc_tEb _ZN4sycl3_V17handler28memcpyToHostOnlyDeviceGlobalEPKvS3_mbmm _ZN4sycl3_V17handler28setArgsToAssociatedAccessorsEv _ZN4sycl3_V17handler28setStateExplicitKernelBundleEv _ZN4sycl3_V17handler30memcpyFromHostOnlyDeviceGlobalEPvPKvbmm _ZN4sycl3_V17handler30verifyUsedKernelBundleInternalENS0_6detail11string_viewE -_ZN4sycl3_V17handler32verifyDeviceHasProgressGuaranteeENS0_3ext6oneapi12experimental26forward_progress_guaranteeENS4_15execution_scopeES6_ _ZN4sycl3_V17handler34ext_oneapi_wait_external_semaphoreENS0_3ext6oneapi12experimental18external_semaphoreE _ZN4sycl3_V17handler34ext_oneapi_wait_external_semaphoreENS0_3ext6oneapi12experimental18external_semaphoreEm _ZN4sycl3_V17handler36ext_oneapi_signal_external_semaphoreENS0_3ext6oneapi12experimental18external_semaphoreE @@ -3649,16 +3627,11 @@ _ZN4sycl3_V17handler8finalizeEv _ZN4sycl3_V17handler8getQueueEv _ZN4sycl3_V17handler8prefetchEPKvm _ZN4sycl3_V17handler8prefetchEPKvmNS0_3ext6oneapi12experimental13prefetch_typeE -_ZN4sycl3_V17handler9clearArgsEv _ZN4sycl3_V17handler9fill_implEPvPKvmm _ZN4sycl3_V17handlerC1EOSt10unique_ptrINS0_6detail12handler_implESt14default_deleteIS4_EE -_ZN4sycl3_V17handlerC1ESt10shared_ptrINS0_3ext6oneapi12experimental6detail10graph_implEE -_ZN4sycl3_V17handlerC1ESt10shared_ptrINS0_6detail10queue_implEES5_S5_b -_ZN4sycl3_V17handlerC1ESt10shared_ptrINS0_6detail10queue_implEEb +_ZN4sycl3_V17handlerC1ERNS0_6detail12handler_implE _ZN4sycl3_V17handlerC2EOSt10unique_ptrINS0_6detail12handler_implESt14default_deleteIS4_EE -_ZN4sycl3_V17handlerC2ESt10shared_ptrINS0_3ext6oneapi12experimental6detail10graph_implEE -_ZN4sycl3_V17handlerC2ESt10shared_ptrINS0_6detail10queue_implEES5_S5_b -_ZN4sycl3_V17handlerC2ESt10shared_ptrINS0_6detail10queue_implEEb +_ZN4sycl3_V17handlerC2ERNS0_6detail12handler_implE _ZN4sycl3_V17handlerD1Ev _ZN4sycl3_V17handlerD2Ev _ZN4sycl3_V17samplerC1ENS0_29coordinate_normalization_modeENS0_15addressing_modeENS0_14filtering_modeERKNS0_13property_listE @@ -4103,10 +4076,8 @@ _ZNK4sycl3_V17handler14getContextImplEv _ZNK4sycl3_V17handler15getCommandGraphEv _ZNK4sycl3_V17handler15getKernelBundleEv _ZNK4sycl3_V17handler16getDeviceBackendEv -_ZNK4sycl3_V17handler17getContextImplPtrEv _ZNK4sycl3_V17handler21HasAssociatedAccessorEPNS0_6detail16AccessorImplHostENS0_6access6targetE _ZNK4sycl3_V17handler27isStateExplicitKernelBundleEv -_ZNK4sycl3_V17handler30getOrInsertHandlerKernelBundleEb _ZNK4sycl3_V17handler33getOrInsertHandlerKernelBundlePtrEb _ZNK4sycl3_V17handler7getTypeEv _ZNK4sycl3_V17sampler11getPropListEv diff --git a/sycl/test/abi/sycl_symbols_windows.dump b/sycl/test/abi/sycl_symbols_windows.dump index 1b9be4770dd96..0c1bc8c04cfef 100644 --- a/sycl/test/abi/sycl_symbols_windows.dump +++ b/sycl/test/abi/sycl_symbols_windows.dump @@ -401,10 +401,7 @@ ??0gpu_selector@_V1@sycl@@QEAA@AEBV012@@Z ??0gpu_selector@_V1@sycl@@QEAA@XZ ??0handler@_V1@sycl@@AEAA@$$QEAV?$unique_ptr@Vhandler_impl@detail@_V1@sycl@@U?$default_delete@Vhandler_impl@detail@_V1@sycl@@@std@@@std@@@Z -??0handler@_V1@sycl@@AEAA@V?$shared_ptr@Vgraph_impl@detail@experimental@oneapi@ext@_V1@sycl@@@std@@@Z -??0handler@_V1@sycl@@AEAA@V?$shared_ptr@Vqueue_impl@detail@_V1@sycl@@@std@@00_N@Z -??0handler@_V1@sycl@@AEAA@V?$shared_ptr@Vqueue_impl@detail@_V1@sycl@@@std@@PEAVqueue_impl@detail@12@_N@Z -??0handler@_V1@sycl@@AEAA@V?$shared_ptr@Vqueue_impl@detail@_V1@sycl@@@std@@_N@Z +??0handler@_V1@sycl@@AEAA@AEAVhandler_impl@detail@12@@Z ??0image_mem@experimental@oneapi@ext@_V1@sycl@@QEAA@$$QEAV012345@@Z ??0image_mem@experimental@oneapi@ext@_V1@sycl@@QEAA@AEBUimage_descriptor@12345@AEBVdevice@45@AEBVcontext@45@@Z ??0image_mem@experimental@oneapi@ext@_V1@sycl@@QEAA@AEBUimage_descriptor@12345@AEBVqueue@45@@Z @@ -3753,7 +3750,6 @@ ?add@free_function_info_map@detail@_V1@sycl@@YAXPEBQEBDPEBII@Z ?add@host_pipe_map@detail@_V1@sycl@@YAXPEBXPEBD@Z ?add@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@QEAA?AVnode@34567@AEBVproperty_list@67@@Z -?addAccessorReq@handler@_V1@sycl@@AEAAXV?$shared_ptr@VAccessorImplHost@detail@_V1@sycl@@@std@@@Z ?addArg@handler@_V1@sycl@@AEAAXW4kernel_param_kind_t@detail@23@PEAXHH@Z ?addCounterInit@detail@_V1@sycl@@YAXAEAVhandler@23@AEAV?$shared_ptr@Vqueue_impl@detail@_V1@sycl@@@std@@AEAV?$shared_ptr@H@6@@Z ?addGraphLeafDependencies@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEAAXVnode@34567@@Z @@ -3811,7 +3807,6 @@ ?cancel_fusion@fusion_wrapper@experimental@codeplay@ext@_V1@sycl@@QEAAXXZ ?category@exception@_V1@sycl@@QEBAAEBVerror_category@std@@XZ ?checkNodePropertiesAndThrow@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@KAXAEBVproperty_list@67@@Z -?clearArgs@handler@_V1@sycl@@AEAAXXZ ?close@ipc_memory@experimental@oneapi@ext@_V1@sycl@@YAXPEAXAEBVcontext@56@@Z ?code@exception@_V1@sycl@@QEBAAEBVerror_code@std@@XZ ?compile_from_source@detail@experimental@oneapi@ext@_V1@sycl@@YA?AV?$kernel_bundle@$00@56@AEAV?$kernel_bundle@$02@56@AEBV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@std@@AEBV?$vector@Vstring_view@detail@_V1@sycl@@V?$allocator@Vstring_view@detail@_V1@sycl@@@std@@@std@@PEAVstring@156@2@Z @@ -4008,11 +4003,8 @@ ?ext_oneapi_wait_external_semaphore@queue@_V1@sycl@@QEAA?AVevent@23@Uexternal_semaphore@experimental@oneapi@ext@23@_KAEBV?$vector@Vevent@_V1@sycl@@V?$allocator@Vevent@_V1@sycl@@@std@@@std@@AEBUcode_location@detail@23@@Z ?ext_oneapi_wait_external_semaphore@queue@_V1@sycl@@QEAA?AVevent@23@Uexternal_semaphore@experimental@oneapi@ext@23@_KV423@AEBUcode_location@detail@23@@Z ?extractArgsAndReqs@handler@_V1@sycl@@AEAAXXZ -?extractArgsAndReqsFromLambda@handler@_V1@sycl@@AEAAXPEADAEBV?$vector@Ukernel_param_desc_t@detail@_V1@sycl@@V?$allocator@Ukernel_param_desc_t@detail@_V1@sycl@@@std@@@std@@_N@Z -?extractArgsAndReqsFromLambda@handler@_V1@sycl@@AEAAXPEADP6A?AUkernel_param_desc_t@detail@23@H@Z_K_N@Z -?extractArgsAndReqsFromLambda@handler@_V1@sycl@@AEAAXPEAD_KPEBUkernel_param_desc_t@detail@23@_N@Z ?fill_impl@handler@_V1@sycl@@AEAAXPEAXPEBX_K2@Z -?finalize@handler@_V1@sycl@@AEAA?AVevent@23@XZ +?finalize@handler@_V1@sycl@@AEAA?AV?$shared_ptr@Vevent_impl@detail@_V1@sycl@@@std@@XZ ?finalize@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@QEBA?AV?$command_graph@$00@34567@AEBVproperty_list@67@@Z ?finalizeImpl@executable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEAAXXZ ?find_device_intersection@detail@_V1@sycl@@YA?AV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@std@@AEBV?$vector@V?$kernel_bundle@$00@_V1@sycl@@V?$allocator@V?$kernel_bundle@$00@_V1@sycl@@@std@@@5@@Z @@ -4045,7 +4037,6 @@ ?getChannelType@image_plain@detail@_V1@sycl@@IEBA?AW4image_channel_type@34@XZ ?getCommandGraph@handler@_V1@sycl@@AEBA?AV?$shared_ptr@Vgraph_impl@detail@experimental@oneapi@ext@_V1@sycl@@@std@@XZ ?getContextImpl@handler@_V1@sycl@@AEBAAEAVcontext_impl@detail@23@XZ -?getContextImplPtr@handler@_V1@sycl@@AEBAAEBV?$shared_ptr@Vcontext_impl@detail@_V1@sycl@@@std@@XZ ?getCurrentDSODir@OSUtil@detail@_V1@sycl@@SA?AV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@XZ ?getDeviceBackend@handler@_V1@sycl@@AEBA?AW4backend@23@XZ ?getDeviceFromHandler@detail@_V1@sycl@@YA?AVdevice@23@AEAVhandler@23@@Z @@ -4087,7 +4078,6 @@ ?getOSMemSize@OSUtil@detail@_V1@sycl@@SA_KXZ ?getOffset@AccessorBaseHost@detail@_V1@sycl@@QEAAAEAV?$id@$02@34@XZ ?getOffset@AccessorBaseHost@detail@_V1@sycl@@QEBAAEBV?$id@$02@34@XZ -?getOrInsertHandlerKernelBundle@handler@_V1@sycl@@AEBA?AV?$shared_ptr@Vkernel_bundle_impl@detail@_V1@sycl@@@std@@_N@Z ?getOrInsertHandlerKernelBundlePtr@handler@_V1@sycl@@AEBAPEAVkernel_bundle_impl@detail@23@_N@Z ?getPitch@SampledImageAccessorBaseHost@detail@_V1@sycl@@QEBA?AV?$id@$02@34@XZ ?getPitch@UnsampledImageAccessorBaseHost@detail@_V1@sycl@@QEBA?AV?$id@$02@34@XZ @@ -4355,7 +4345,6 @@ ?prepare_for_device_copy@experimental@oneapi@ext@_V1@sycl@@YAXPEBX_KAEBVqueue@45@@Z ?print_graph@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEBAXVstring_view@267@_N@Z ?print_graph@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@QEBAXV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@_N@Z -?processArg@handler@_V1@sycl@@AEAAXPEAXAEBW4kernel_param_kind_t@detail@23@H_KAEA_K_N4@Z ?put@ipc_memory@experimental@oneapi@ext@_V1@sycl@@YAXAEAUhandle@123456@AEBVcontext@56@@Z ?query@tls_code_loc_t@detail@_V1@sycl@@QEAAAEBUcode_location@234@XZ ?reduComputeWGSize@detail@_V1@sycl@@YA_K_K0AEA_K@Z @@ -4365,7 +4354,6 @@ ?reduGetMaxWGSize@detail@_V1@sycl@@YA_KV?$shared_ptr@Vqueue_impl@detail@_V1@sycl@@@std@@_K@Z ?reduGetPreferredWGSize@detail@_V1@sycl@@YA_KAEAV?$shared_ptr@Vqueue_impl@detail@_V1@sycl@@@std@@_K@Z ?reduGetPreferredWGSize@detail@_V1@sycl@@YA_KAEAVhandler@23@_K@Z -?registerDynamicParameter@handler@_V1@sycl@@AEAAXAEAVdynamic_parameter_base@detail@experimental@oneapi@ext@23@H@Z ?registerDynamicParameter@handler@_V1@sycl@@AEAAXPEAVdynamic_parameter_impl@detail@experimental@oneapi@ext@23@H@Z ?release_external_memory@experimental@oneapi@ext@_V1@sycl@@YAXUexternal_mem@12345@AEBVdevice@45@AEBVcontext@45@@Z ?release_external_memory@experimental@oneapi@ext@_V1@sycl@@YAXUexternal_mem@12345@AEBVqueue@45@@Z @@ -4384,7 +4372,6 @@ ?sampledImageConstructorNotification@detail@_V1@sycl@@YAXPEAX0AEBV?$optional@W4image_target@_V1@sycl@@@std@@PEBXIAEBUcode_location@123@@Z ?sampledImageConstructorNotification@image_plain@detail@_V1@sycl@@IEAAXAEBUcode_location@234@PEAXPEBXIQEA_KW4image_format@34@AEBUimage_sampler@34@@Z ?sampledImageDestructorNotification@image_plain@detail@_V1@sycl@@IEAAXPEAX@Z -?saveCodeLoc@handler@_V1@sycl@@AEAAXUcode_location@detail@23@@Z ?saveCodeLoc@handler@_V1@sycl@@AEAAXUcode_location@detail@23@_N@Z ?select_device@detail@_V1@sycl@@YA?AVdevice@23@AEBV?$function@$$A6AHAEBVdevice@_V1@sycl@@@Z@std@@@Z ?select_device@detail@_V1@sycl@@YA?AVdevice@23@AEBV?$function@$$A6AHAEBVdevice@_V1@sycl@@@Z@std@@AEBVcontext@23@@Z @@ -4400,19 +4387,9 @@ ?setDevice@HostProfilingInfo@detail@_V1@sycl@@QEAAXPEAVdevice_impl@234@@Z ?setDeviceKernelInfo@handler@_V1@sycl@@AEAAX$$QEAVkernel@23@@Z ?setDeviceKernelInfoPtr@handler@_V1@sycl@@AEAAXPEAVDeviceKernelInfo@detail@23@@Z -?setHandlerKernelBundle@handler@_V1@sycl@@AEAAXAEBV?$shared_ptr@Vkernel_bundle_impl@detail@_V1@sycl@@@std@@@Z ?setHandlerKernelBundle@handler@_V1@sycl@@AEAAXVkernel@23@@Z -?setKernelCacheConfig@handler@_V1@sycl@@AEAAXW4StableKernelCacheConfig@123@@Z -?setKernelClusterLaunch@handler@_V1@sycl@@AEAAXV?$range@$00@23@@Z -?setKernelClusterLaunch@handler@_V1@sycl@@AEAAXV?$range@$01@23@@Z -?setKernelClusterLaunch@handler@_V1@sycl@@AEAAXV?$range@$02@23@@Z -?setKernelClusterLaunch@handler@_V1@sycl@@AEAAXV?$range@$02@23@H@Z ?setKernelFunc@handler@_V1@sycl@@AEAAXPEAX@Z -?setKernelInfo@handler@_V1@sycl@@AEAAXPEAXHP6A?AUkernel_param_desc_t@detail@23@H@Z_N2@Z -?setKernelIsCooperative@handler@_V1@sycl@@AEAAX_N@Z ?setKernelLaunchProperties@handler@_V1@sycl@@AEAAXAEBU?$PropsHolder@Uwork_group_scratch_size@experimental@oneapi@ext@_V1@sycl@@Ucache_config@2intel@456@Uuse_root_sync_key@23456@Uwork_group_progress_key@23456@Usub_group_progress_key@23456@Uwork_item_progress_key@23456@U?$cluster_size@$00@cuda@23456@U?$cluster_size@$01@cuda@23456@U?$cluster_size@$02@cuda@23456@@kernel_launch_properties_v1@detail@23@@Z -?setKernelNameBasedCachePtr@handler@_V1@sycl@@AEAAXPEAUKernelNameBasedCacheT@detail@23@@Z -?setKernelWorkGroupMem@handler@_V1@sycl@@AEAAX_K@Z ?setLocalAccessorArgHelper@handler@_V1@sycl@@AEAAXHAEAVLocalAccessorBaseHost@detail@23@@Z ?setNDRangeDescriptor@handler@_V1@sycl@@AEAAXV?$range@$00@23@0V?$id@$00@23@@Z ?setNDRangeDescriptor@handler@_V1@sycl@@AEAAXV?$range@$00@23@V?$id@$00@23@@Z @@ -4423,10 +4400,6 @@ ?setNDRangeDescriptor@handler@_V1@sycl@@AEAAXV?$range@$02@23@0V?$id@$02@23@@Z ?setNDRangeDescriptor@handler@_V1@sycl@@AEAAXV?$range@$02@23@V?$id@$02@23@@Z ?setNDRangeDescriptor@handler@_V1@sycl@@AEAAXV?$range@$02@23@_N@Z -?setNDRangeDescriptorPadded@handler@_V1@sycl@@AEAAXV?$range@$02@23@0V?$id@$02@23@H@Z -?setNDRangeDescriptorPadded@handler@_V1@sycl@@AEAAXV?$range@$02@23@V?$id@$02@23@H@Z -?setNDRangeDescriptorPadded@handler@_V1@sycl@@AEAAXV?$range@$02@23@_NH@Z -?setNDRangeUsed@handler@_V1@sycl@@AEAAX_N@Z ?setStateExplicitKernelBundle@handler@_V1@sycl@@AEAAXXZ ?setStateSpecConstSet@handler@_V1@sycl@@AEAAXXZ ?setType@handler@_V1@sycl@@AEAAXW4CGType@detail@23@@Z @@ -4482,7 +4455,6 @@ ?sycl_category@_V1@sycl@@YAAEBVerror_category@std@@XZ ?throwIfActionIsCreated@handler@_V1@sycl@@AEAAXXZ ?throwOnKernelParameterMisuse@handler@_V1@sycl@@AEBAXAEBUCompileTimeKernelInfoTy@compile_time_kernel_info_v1@detail@23@@Z -?throwOnKernelParameterMisuseHelper@handler@_V1@sycl@@AEBAXHP6A?AUkernel_param_desc_t@detail@23@H@Z@Z ?throw_asynchronous@queue@_V1@sycl@@QEAAXXZ ?unmap@experimental@oneapi@ext@_V1@sycl@@YAXPEBX_KAEBVcontext@45@@Z ?unmap_external_image_memory@experimental@oneapi@ext@_V1@sycl@@YAXUimage_mem_handle@12345@W4image_type@12345@AEBVdevice@45@AEBVcontext@45@@Z @@ -4500,10 +4472,8 @@ ?updateValue@dynamic_parameter_base@detail@experimental@oneapi@ext@_V1@sycl@@IEAAXPEBX_K@Z ?updateWorkGroupMem@dynamic_work_group_memory_base@detail@experimental@oneapi@ext@_V1@sycl@@IEAAX_K@Z ?use_kernel_bundle@handler@_V1@sycl@@QEAAXAEBV?$kernel_bundle@$01@23@@Z -?verifyDeviceHasProgressGuarantee@handler@_V1@sycl@@AEAAXW4forward_progress_guarantee@experimental@oneapi@ext@23@W4execution_scope@56723@1@Z ?verifyReductionProps@detail@_V1@sycl@@YAXAEBVproperty_list@23@@Z ?verifyUSMAllocatorProperties@_V1@sycl@@YAXAEBVproperty_list@12@@Z -?verifyUsedKernelBundle@handler@_V1@sycl@@AEAAXAEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@@Z ?verifyUsedKernelBundleInternal@handler@_V1@sycl@@AEAAXVstring_view@detail@23@@Z ?wait@event@_V1@sycl@@QEAAXXZ ?wait@event@_V1@sycl@@SAXAEBV?$vector@Vevent@_V1@sycl@@V?$allocator@Vevent@_V1@sycl@@@std@@@std@@@Z diff --git a/sycl/test/abi/symbol_size_alignment.cpp b/sycl/test/abi/symbol_size_alignment.cpp index 2a0d09351f7dc..5f0279580e657 100644 --- a/sycl/test/abi/symbol_size_alignment.cpp +++ b/sycl/test/abi/symbol_size_alignment.cpp @@ -53,9 +53,9 @@ int main() { check(); check(); #ifdef _MSC_VER - check(); + check(); #else - check(); + check(); #endif check, 16, 8>(); check(); diff --git a/sycl/unittests/scheduler/InOrderQueueSyncCheck.cpp b/sycl/unittests/scheduler/InOrderQueueSyncCheck.cpp index 04d2f81281de2..ee305793107da 100644 --- a/sycl/unittests/scheduler/InOrderQueueSyncCheck.cpp +++ b/sycl/unittests/scheduler/InOrderQueueSyncCheck.cpp @@ -42,17 +42,9 @@ class LimitedHandler { virtual void depends_on(const std::vector &Events) {} virtual void depends_on(event Event) {}; -#ifdef __INTEL_PREVIEW_BREAKING_CHANGES virtual sycl::detail::EventImplPtr finalize() { return detail::event_impl::create_default_event(); } -#else - virtual event finalize() { - sycl::detail::EventImplPtr NewEvent = - detail::event_impl::create_completed_host_event(); - return sycl::detail::createSyclObjFromImpl(NewEvent); - } -#endif sycl::detail::CGType getType() { return MCGType; }