[SYCL] Add code location data to XPTI trace in case of exception thrown (#8101)

Adds extra information about code location of command submission to
exception message.
Now instrumented a few sycl::queue non-variadic methods which takes code
location from parameter.
Also instrumented variadic parallel_for where code location is extracted
from kernel info.

---------

Signed-off-by: Tikhomirova, Kseniya <[email protected]>

Author: KseniyaTikhomirova

sycl/doc/design/SYCLInstrumentationUsingXPTI.md

@@ -241,7 +241,7 @@ by the SYCL runtime.
 
 ## SYCL Stream `"sycl"` Notification Signatures
 
-All trace point types in bold provide semantic information about the graph, nodes and edges and the toplogy of the asynchronous task graphs created by the runtime.
+All trace point types in bold provide semantic information about the graph, nodes and edges and the topology of the asynchronous task graphs created by the runtime.
 | Trace Point Type | Parameter Description | Metadata |
 | :--------------: | :-------------------- | :------- |
 | **`graph_create`** | <div style="text-align: left"><li>**trace_type**: `xpti::trace_point_type_t::graph_create` that marks the creation of an asynchronous graph.</li> <li> **parent**: `nullptr`</li> <li> **event**: The global asynchronous graph object ID. All other graph related events such as node and edge creation will always this ID as the parent ID. </li> <li> **instance**: Unique ID related to the event, but not a correlation ID as there are other events to correlate to. </li> <li> **user_data**: `nullptr`</li> <p></p> SYCL runtime will always have one instance of a graph object with many disjoint subgraphs that get created during the execution of an application. </div>                                                                                                                                                                                                                                                                                                                                                                                                                                                                     | None                                                                                                                                                                                                                                                     |
@@ -256,6 +256,7 @@ All trace point types in bold provide semantic information about the graph, node
 | `wait_end`         | <div style="text-align: left"><li>**trace_type**: `xpti::trace_point_type_t::wait_end` that marks the beginning of the wait on an `event`</li> <li> **parent**: `nullptr`</li> <li> **event**: The event ID will reflect the ID of the command group object submission that created this event or a new event based on the combination of the string "queue.wait" and the address of the event. </li> <li> **instance**: Unique ID to allow the correlation of the `wait_begin` event with the `wait_end` event. </li> <li> **user_data**: String indicating `queue.wait` and the address of the event  as `const char *` </li></div>                                                                                                                                                                                                                                                                                                                                                                                                                                                                                                                   | `sycl_device`, `sycl_device_type`, `sycl_device_name`, `sym_function_name`, `sym_source_file_name`, `sym_line_no`, `sym_column_no`                                                                                                                                                                            |
 | `barrier_begin`       | <div style="text-align: left"><li>**trace_type**: `xpti::trace_point_type_t::barrier_begin` that marks the beginning of a barrier while enqueuing a command group object</li> <li> **parent**: The global graph event that is created during the `graph_create` event.</li> <li> **event**: The event ID will reflect the ID of the command group object that has encountered a barrier during the enqueue operation.  </li> <li> **instance**: Unique ID to allow the correlation of the `barrier_begin` event with the `barrier_end` event. </li> <li> **user_data**: String indicating `enqueue.barrier` and the reason for the barrier as a `const char *` </li> <p></p>The reason for the barrier could be one of `Buffer locked by host accessor`, `Blocked by host task` or `Unknown reason`.</div>   | <li> Computational Kernels </li> `sycl_device`, `sycl_device_type`, `sycl_device_name`, `kernel_name`, `from_source`, `sym_function_name`, `sym_source_file_name`, `sym_line_no`, `sym_column_no` <li>Memory operations</li> `memory_object`, `offset`, `access_range_start`, `access_range_end`, `allocation_type`, `copy_from`, `copy_to` |
 | `barrier_end`         | <div style="text-align: left"><li>**trace_type**: `xpti::trace_point_type_t::barrier_end` that marks the end of the barrier that is encountered during enqueue.</li> <li> **parent**: The global graph event that is created during the `graph_create` event.</li> <li> **event**: The event ID will reflect the ID of the command group object that has encountered a barrier during the enqueue operation. </li> <li> **instance**: Unique ID to allow the correlation of the `barrier_begin` event with the `barrier_end` event. </li> <li> **user_data**: String indicating `enqueue.barrier` and the reason for the barrier as a `const char *` </li> <p></p>The reason for the barrier could be one of `Buffer locked by host accessor`, `Blocked by host task` or `Unknown reason`.</div>   | <li> Computational Kernels </li> `sycl_device`, `sycl_device_type`, `sycl_device_name`, `kernel_name`, `from_source`, `sym_function_name`, `sym_source_file_name`, `sym_line_no`, `sym_column_no` <li>Memory operations</li> `memory_object`, `offset`, `access_range_start`, `access_range_end`, `allocation_type`, `copy_from`, `copy_to` |
+| `diagnostics`         | <div style="text-align: left"><li>**trace_type**: `xpti::trace_point_type_t::diagnostics` that represents general purpose notifications. For example, it is emitted when an exception is thrown in SYCL runtime. </li> <li> **parent**: Set to NULL.</li> <li> **event**: The event ID will reflect the code location of notification origin, if available. </li> <li> **instance**: An instance ID that records the number of times this code location has been seen. </li> <li> **user_data**: String with diagnostic message as a `const char *` </li></div>   | `sym_function_name`, `sym_source_file_name`, `sym_line_no`, `sym_column_no` |
 
 ### Metadata description
 

sycl/include/sycl/handler.hpp

@@ -656,7 +656,6 @@ class __SYCL_EXPORT handler {
             typename LambdaArgType>
   void StoreLambda(KernelType KernelFunc) {
     using KI = detail::KernelInfo<KernelName>;
-
     constexpr bool IsCallableWithKernelHandler =
         detail::KernelLambdaHasKernelHandlerArgT<KernelType,
                                                  LambdaArgType>::value;
@@ -670,7 +669,6 @@ class __SYCL_EXPORT handler {
     KernelType *KernelPtr =
         ResetHostKernel<KernelType, LambdaArgType, Dims>(KernelFunc);
 
-    using KI = sycl::detail::KernelInfo<KernelName>;
     constexpr bool KernelHasName =
         KI::getName() != nullptr && KI::getName()[0] != '\0';
 

sycl/include/sycl/queue.hpp

@@ -307,7 +307,7 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
   /// \return a SYCL event object for the submitted command group.
   template <typename T> event submit(T CGF _CODELOCPARAM(&CodeLoc)) {
     _CODELOCARG(&CodeLoc);
-
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
 #if __SYCL_USE_FALLBACK_ASSERT
     auto PostProcess = [this, &CodeLoc](bool IsKernel, bool KernelUsesAssert,
                                         event &E) {
@@ -343,7 +343,7 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
   template <typename T>
   event submit(T CGF, queue &SecondaryQueue _CODELOCPARAM(&CodeLoc)) {
     _CODELOCARG(&CodeLoc);
-
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
 #if __SYCL_USE_FALLBACK_ASSERT
     auto PostProcess = [this, &SecondaryQueue, &CodeLoc](
                            bool IsKernel, bool KernelUsesAssert, event &E) {
@@ -432,7 +432,7 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
   /// @param CodeLoc is the code location of the submit call (default argument)
   void wait(_CODELOCONLYPARAM(&CodeLoc)) {
     _CODELOCARG(&CodeLoc);
-
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     wait_proxy(CodeLoc);
   }
 
@@ -446,7 +446,7 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
   /// @param CodeLoc is the code location of the submit call (default argument)
   void wait_and_throw(_CODELOCONLYPARAM(&CodeLoc)) {
     _CODELOCARG(&CodeLoc);
-
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     wait_and_throw_proxy(CodeLoc);
   }
 
@@ -1345,6 +1345,7 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
         "sycl::queue.single_task() requires a kernel instead of command group. "
         "Use queue.submit() instead");
     _CODELOCARG(&CodeLoc);
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     return submit(
         [&](handler &CGH) {
           CGH.template single_task<KernelName, KernelType, PropertiesT>(
@@ -1384,6 +1385,7 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
         "sycl::queue.single_task() requires a kernel instead of command group. "
         "Use queue.submit() instead");
     _CODELOCARG(&CodeLoc);
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     return submit(
         [&](handler &CGH) {
           CGH.depends_on(DepEvent);
@@ -1427,6 +1429,7 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
         "sycl::queue.single_task() requires a kernel instead of command group. "
         "Use queue.submit() instead");
     _CODELOCARG(&CodeLoc);
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     return submit(
         [&](handler &CGH) {
           CGH.depends_on(DepEvents);
@@ -1667,8 +1670,11 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
           ext::oneapi::experimental::is_property_list<PropertiesT>::value,
       event>
   parallel_for(nd_range<Dims> Range, PropertiesT Properties, RestT &&...Rest) {
-    // Actual code location needs to be captured from KernelInfo object.
-    const detail::code_location CodeLoc = {};
+    using KI = sycl::detail::KernelInfo<KernelName>;
+    constexpr detail::code_location CodeLoc(
+        KI::getFileName(), KI::getFunctionName(), KI::getLineNumber(),
+        KI::getColumnNumber());
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     return submit(
         [&](handler &CGH) {
           CGH.template parallel_for<KernelName>(Range, Properties, Rest...);
@@ -1701,8 +1707,11 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
   template <typename KernelName = detail::auto_name, int Dims,
             typename... RestT>
   event parallel_for(nd_range<Dims> Range, event DepEvent, RestT &&...Rest) {
-    // Actual code location needs to be captured from KernelInfo object.
-    const detail::code_location CodeLoc = {};
+    using KI = sycl::detail::KernelInfo<KernelName>;
+    constexpr detail::code_location CodeLoc(
+        KI::getFileName(), KI::getFunctionName(), KI::getLineNumber(),
+        KI::getColumnNumber());
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     return submit(
         [&](handler &CGH) {
           CGH.depends_on(DepEvent);
@@ -1723,8 +1732,11 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
             typename... RestT>
   event parallel_for(nd_range<Dims> Range, const std::vector<event> &DepEvents,
                      RestT &&...Rest) {
-    // Actual code location needs to be captured from KernelInfo object.
-    const detail::code_location CodeLoc = {};
+    using KI = sycl::detail::KernelInfo<KernelName>;
+    constexpr detail::code_location CodeLoc(
+        KI::getFileName(), KI::getFunctionName(), KI::getLineNumber(),
+        KI::getColumnNumber());
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     return submit(
         [&](handler &CGH) {
           CGH.depends_on(DepEvents);
@@ -1950,8 +1962,11 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
       event>
   parallel_for_impl(range<Dims> Range, PropertiesT Properties,
                     RestT &&...Rest) {
-    // Actual code location needs to be captured from KernelInfo object.
-    const detail::code_location CodeLoc = {};
+    using KI = sycl::detail::KernelInfo<KernelName>;
+    constexpr detail::code_location CodeLoc(
+        KI::getFileName(), KI::getFunctionName(), KI::getLineNumber(),
+        KI::getColumnNumber());
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     return submit(
         [&](handler &CGH) {
           CGH.template parallel_for<KernelName>(Range, Properties, Rest...);
@@ -1985,8 +2000,11 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
       ext::oneapi::experimental::is_property_list<PropertiesT>::value, event>
   parallel_for_impl(range<Dims> Range, event DepEvent, PropertiesT Properties,
                     RestT &&...Rest) {
-    // Actual code location needs to be captured from KernelInfo object.
-    const detail::code_location CodeLoc = {};
+    using KI = sycl::detail::KernelInfo<KernelName>;
+    constexpr detail::code_location CodeLoc(
+        KI::getFileName(), KI::getFunctionName(), KI::getLineNumber(),
+        KI::getColumnNumber());
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     return submit(
         [&](handler &CGH) {
           CGH.depends_on(DepEvent);
@@ -2022,8 +2040,11 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
       ext::oneapi::experimental::is_property_list<PropertiesT>::value, event>
   parallel_for_impl(range<Dims> Range, const std::vector<event> &DepEvents,
                     PropertiesT Properties, RestT &&...Rest) {
-    // Actual code location needs to be captured from KernelInfo object.
-    const detail::code_location CodeLoc = {};
+    using KI = sycl::detail::KernelInfo<KernelName>;
+    constexpr detail::code_location CodeLoc(
+        KI::getFileName(), KI::getFunctionName(), KI::getLineNumber(),
+        KI::getColumnNumber());
+    detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
     return submit(
         [&](handler &CGH) {
           CGH.depends_on(DepEvents);

sycl/source/CMakeLists.txt

@@ -3,7 +3,6 @@
 #2. Use AddLLVM to modify the build and access config options
 #cmake_policy(SET CMP0057 NEW)
 #include(AddLLVM)
-
 configure_file(
   ${CMAKE_CURRENT_SOURCE_DIR}/version.rc.in
   ${CMAKE_CURRENT_BINARY_DIR}/version.rc

sycl/source/detail/global_handler.cpp

@@ -71,6 +71,49 @@ std::atomic_uint ObjectUsageCounter::MCounter{0};
 GlobalHandler::GlobalHandler() = default;
 GlobalHandler::~GlobalHandler() = default;
 
+void GlobalHandler::InitXPTI() {
+#ifdef XPTI_ENABLE_INSTRUMENTATION
+  // Let subscribers know a new stream is being initialized
+  getXPTIRegistry().initializeStream(SYCL_STREAM_NAME, GMajVer, GMinVer,
+                                     GVerStr);
+  xpti::payload_t SYCLPayload("SYCL Runtime Exceptions");
+  uint64_t SYCLInstanceNo;
+  GSYCLCallEvent = xptiMakeEvent("SYCL Try-catch Exceptions", &SYCLPayload,
+                                 xpti::trace_algorithm_event, xpti_at::active,
+                                 &SYCLInstanceNo);
+#endif
+}
+
+void GlobalHandler::TraceEventXPTI(const char *Message) {
+#ifdef XPTI_ENABLE_INSTRUMENTATION
+  if (!Message)
+    return;
+  if (xptiTraceEnabled()) {
+    // We have to handle the cases where: (1) we may have just the code location
+    // set and not UID and (2) UID set
+    detail::tls_code_loc_t Tls;
+    auto CodeLocation = Tls.query();
+
+    // Creating a tracepoint will convert a CodeLocation to UID, if not set
+    xpti::framework::tracepoint_t TP(
+        CodeLocation.fileName(), CodeLocation.functionName(),
+        CodeLocation.lineNumber(), CodeLocation.columnNumber(), nullptr);
+
+    // The call to notify will have the signature of:
+    // (1) the stream defined in .stream()
+    // (2) The trace type equal to what is set by .trace_type()
+    // (3) Parent event set to NULL
+    // (4) Current event set to one created from CodeLocation and UID
+    // (5) An instance ID that records the number of times this code location
+    // has been seen (6) The message generated by the exception handler
+    TP.stream(SYCL_STREAM_NAME)
+        .trace_type(xpti::trace_point_type_t::diagnostics)
+        .notify(static_cast<const void *>(Message));
+  }
+
+#endif
+}
+
 GlobalHandler *&GlobalHandler::getInstancePtr() {
   static GlobalHandler *RTGlobalObjHandler = new GlobalHandler();
   return RTGlobalObjHandler;
@@ -79,6 +122,13 @@ GlobalHandler *&GlobalHandler::getInstancePtr() {
 GlobalHandler &GlobalHandler::instance() {
   GlobalHandler *RTGlobalObjHandler = GlobalHandler::getInstancePtr();
   assert(RTGlobalObjHandler && "Handler must not be deallocated earlier");
+
+#ifdef XPTI_ENABLE_INSTRUMENTATION
+  static std::once_flag InitXPTIFlag;
+  if (xptiTraceEnabled()) {
+    std::call_once(InitXPTIFlag, [&]() { RTGlobalObjHandler->InitXPTI(); });
+  }
+#endif
   return *RTGlobalObjHandler;
 }
 

sycl/source/detail/global_handler.hpp

@@ -79,10 +79,17 @@ class GlobalHandler {
   void drainThreadPool();
   void prepareSchedulerToRelease();
 
+  void InitXPTI();
+  void TraceEventXPTI(const char *Message);
+
   // For testing purposes only
   void attachScheduler(Scheduler *Scheduler);
 
 private:
+#ifdef XPTI_ENABLE_INSTRUMENTATION
+  void *GSYCLCallEvent = nullptr;
+#endif
+
   friend void shutdown();
   friend class ObjectUsageCounter;
   static GlobalHandler *&getInstancePtr();

sycl/source/detail/pi.cpp

@@ -38,10 +38,6 @@
 #include "xpti/xpti_trace_framework.h"
 #endif
 
-#define STR(x) #x
-#define SYCL_VERSION_STR                                                       \
-  "sycl " STR(__LIBSYCL_MAJOR_VERSION) "." STR(__LIBSYCL_MINOR_VERSION)
-
 namespace sycl {
 __SYCL_INLINE_VER_NAMESPACE(_V1) {
 namespace detail {
@@ -54,11 +50,6 @@ xpti_td *GSYCLGraphEvent = nullptr;
 xpti_td *GPICallEvent = nullptr;
 /// Event to be used by PI layer calls with arguments
 xpti_td *GPIArgCallEvent = nullptr;
-/// Constants being used as placeholder until one is able to reliably get the
-/// version of the SYCL runtime
-constexpr uint32_t GMajVer = __LIBSYCL_MAJOR_VERSION;
-constexpr uint32_t GMinVer = __LIBSYCL_MINOR_VERSION;
-constexpr const char *GVerStr = SYCL_VERSION_STR;
 #endif // XPTI_ENABLE_INSTRUMENTATION
 
 template <sycl::backend BE> void *getPluginOpaqueData(void *OpaqueDataParam) {