Enable conditionally and make API C-compatible

Enable SYCL source compilation only for DPC++ versions that actually
support the compilation, based on the __SYCL_COMPILER_VERSION reported.
Use the correct naming for the property based on DPC++ version.

Remove all mentions of `std::vector` and other STL types from the header
and use opaque pointers instead.

Signed-off-by: Lukas Sommer <lukas.sommer@codeplay.com>

Author: sommerlukas

dpctl/_backend.pxd

@@ -454,18 +454,18 @@ cdef extern from "syclinterface/dpctl_sycl_kernel_bundle_interface.h":
     cdef DPCTLBuildOptionListRef DPCTLBuildOptionList_Create()
     cdef void DPCTLBuildOptionList_Delete(DPCTLBuildOptionListRef Ref)
     cdef void DPCTLBuildOptionList_Append(DPCTLBuildOptionListRef Ref,
-                                 const char *Option)
+                                          const char *Option)
 
     cdef DPCTLKernelNameListRef DPCTLKernelNameList_Create()
     cdef void DPCTLKernelNameList_Delete(DPCTLKernelNameListRef Ref)
     cdef void DPCTLKernelNameList_Append(DPCTLKernelNameListRef Ref,
-                                const char *Option)
+                                         const char *Option)
 
     cdef DPCTLVirtualHeaderListRef DPCTLVirtualHeaderList_Create()
     cdef void DPCTLVirtualHeaderList_Delete(DPCTLVirtualHeaderListRef Ref)
     cdef void DPCTLVirtualHeaderList_Append(DPCTLVirtualHeaderListRef Ref,
-                                   const char *Name,
-                                   const char *Content)
+                                            const char *Name,
+                                            const char *Content)
 
     cdef DPCTLSyclKernelBundleRef DPCTLKernelBundle_CreateFromSYCLSource(
         const DPCTLSyclContextRef Ctx,
@@ -475,11 +475,12 @@ cdef extern from "syclinterface/dpctl_sycl_kernel_bundle_interface.h":
         DPCTLKernelNameListRef Names,
         DPCTLBuildOptionListRef BuildOptions)
 
-    cdef DPCTLSyclKernelRef DPCTLKernelBundle_GetSyclKernel(DPCTLSyclKernelBundleRef KBRef,
-                                const char *KernelName)
+    cdef DPCTLSyclKernelRef DPCTLKernelBundle_GetSyclKernel(
+                                                DPCTLSyclKernelBundleRef KBRef,
+                                                const char *KernelName)
 
     cdef bool DPCTLKernelBundle_HasSyclKernel(DPCTLSyclKernelBundleRef KBRef,
-                                     const char *KernelName);
+                                              const char *KernelName)
 
 
 cdef extern from "syclinterface/dpctl_sycl_queue_interface.h":

dpctl/_sycl_device.pyx

@@ -26,7 +26,6 @@ from ._backend cimport (  # noqa: E211
     DPCTLDefaultSelector_Create,
     DPCTLDevice_AreEq,
     DPCTLDevice_CanCompileOpenCL,
-    DPCTLDevice_CanCompileSPIRV,
     DPCTLDevice_CanCompileSYCL,
     DPCTLDevice_Copy,
     DPCTLDevice_CreateFromSelector,
@@ -2192,7 +2191,6 @@ cdef class SyclDevice(_SyclDevice):
         raise ValueError(f"Unknown source language {language}")
 
 
-
 cdef api DPCTLSyclDeviceRef SyclDevice_GetDeviceRef(SyclDevice dev):
     """
     C-API function to get opaque device reference from

dpctl/program/_program.pxd

@@ -52,7 +52,8 @@ cdef api class SyclProgram [object PySyclProgramObject, type PySyclProgramType]:
     cdef bint _is_sycl_source
 
     @staticmethod
-    cdef  SyclProgram _create (DPCTLSyclKernelBundleRef pref, bint _is_sycl_source)
+    cdef  SyclProgram _create (DPCTLSyclKernelBundleRef pref,
+                               bint _is_sycl_source)
     cdef  DPCTLSyclKernelBundleRef get_program_ref (self)
     cpdef SyclKernel get_sycl_kernel(self, str kernel_name)
 

dpctl/program/_program.pyx

@@ -211,7 +211,8 @@ cdef class SyclProgram:
     """
 
     @staticmethod
-    cdef SyclProgram _create(DPCTLSyclKernelBundleRef KBRef, bint is_sycl_source):
+    cdef SyclProgram _create(DPCTLSyclKernelBundleRef KBRef,
+                             bint is_sycl_source):
         cdef SyclProgram ret = SyclProgram.__new__(SyclProgram)
         ret._program_ref = KBRef
         ret._is_sycl_source = is_sycl_source
@@ -342,7 +343,10 @@ cpdef create_program_from_spirv(SyclQueue q, const unsigned char[:] IL,
     return SyclProgram._create(KBref, False)
 
 
-cpdef create_program_from_sycl_source(SyclQueue q, unicode source, list headers=[], list registered_names=[], list copts=[]):
+cpdef create_program_from_sycl_source(SyclQueue q, unicode source,
+                                      list headers=None,
+                                      list registered_names=None,
+                                      list copts=None):
     """
         Creates an executable SYCL kernel_bundle from SYCL source code.
 
@@ -384,7 +388,7 @@ cpdef create_program_from_sycl_source(SyclQueue q, unicode source, list headers=
     cdef DPCTLSyclKernelBundleRef KBref
     cdef DPCTLSyclContextRef CRef = q.get_sycl_context().get_context_ref()
     cdef DPCTLSyclDeviceRef DRef = q.get_sycl_device().get_device_ref()
-    cdef bytes bSrc = source.encode('utf8')
+    cdef bytes bSrc = source.encode("utf8")
     cdef const char *Src = <const char*>bSrc
     cdef DPCTLBuildOptionListRef BuildOpts = DPCTLBuildOptionList_Create()
     cdef bytes bOpt
@@ -397,7 +401,7 @@ cpdef create_program_from_sycl_source(SyclQueue q, unicode source, list headers=
         if not isinstance(opt, unicode):
             DPCTLBuildOptionList_Delete(BuildOpts)
             raise SyclProgramCompilationError()
-        bOpt = opt.encode('utf8')
+        bOpt = opt.encode("utf8")
         sOpt = <const char*>bOpt
         DPCTLBuildOptionList_Append(BuildOpts, sOpt)
 
@@ -407,21 +411,22 @@ cpdef create_program_from_sycl_source(SyclQueue q, unicode source, list headers=
             DPCTLBuildOptionList_Delete(BuildOpts)
             DPCTLKernelNameList_Delete(KernelNames)
             raise SyclProgramCompilationError()
-        bName = name.encode('utf8')
+        bName = name.encode("utf8")
         sName = <const char*>bName
         DPCTLKernelNameList_Append(KernelNames, sName)
 
+    cdef DPCTLVirtualHeaderListRef VirtualHeaders
+    VirtualHeaders = DPCTLVirtualHeaderList_Create()
 
-    cdef DPCTLVirtualHeaderListRef VirtualHeaders = DPCTLVirtualHeaderList_Create()
     for name, content in headers:
         if not isinstance(name, unicode) or not isinstance(content, unicode):
             DPCTLBuildOptionList_Delete(BuildOpts)
             DPCTLKernelNameList_Delete(KernelNames)
             DPCTLVirtualHeaderList_Delete(VirtualHeaders)
             raise SyclProgramCompilationError()
-        bName = name.encode('utf8')
+        bName = name.encode("utf8")
         sName = <const char*>bName
-        bContent = content.encode('utf8')
+        bContent = content.encode("utf8")
         sContent = <const char*>bContent
         DPCTLVirtualHeaderList_Append(VirtualHeaders, sName, sContent)
 

libsyclinterface/include/syclinterface/dpctl_sycl_kernel_bundle_interface.h

@@ -32,8 +32,6 @@
 #include "Support/MemOwnershipAttrs.h"
 #include "dpctl_data_types.h"
 #include "dpctl_sycl_types.h"
-#include <string>
-#include <vector>
 
 DPCTL_C_EXTERN_C_BEGIN
 
@@ -131,24 +129,9 @@ DPCTL_API
 __dpctl_give DPCTLSyclKernelBundleRef
 DPCTLKernelBundle_Copy(__dpctl_keep const DPCTLSyclKernelBundleRef KBRef);
 
-struct DPCTLBuildOptionList
-{
-    std::vector<std::string> options;
-};
-
-struct DPCTLKernelNameList
-{
-    std::vector<std::string> names;
-};
-
-struct DPCTLVirtualHeaderList
-{
-    std::vector<std::pair<std::string, std::string>> headers;
-};
-
-using DPCTLBuildOptionListRef = DPCTLBuildOptionList *;
-using DPCTLKernelNameListRef = DPCTLKernelNameList *;
-using DPCTLVirtualHeaderListRef = DPCTLVirtualHeaderList *;
+typedef struct DPCTLBuildOptionList *DPCTLBuildOptionListRef;
+typedef struct DPCTLKernelNameList *DPCTLKernelNameListRef;
+typedef struct DPCTLVirtualHeaderList *DPCTLVirtualHeaderListRef;
 
 /*!
  * @brief Create an empty list of build options.

libsyclinterface/source/dpctl_sycl_kernel_bundle_interface.cpp

@@ -762,58 +762,102 @@ DPCTLKernelBundle_Copy(__dpctl_keep const DPCTLSyclKernelBundleRef KBRef)
     }
 }
 
+using build_option_list_t = std::vector<std::string>;
+
 __dpctl_give DPCTLBuildOptionListRef DPCTLBuildOptionList_Create()
 {
-    return new DPCTLBuildOptionList;
+    auto BuildOptionList =
+        std::unique_ptr<build_option_list_t>(new build_option_list_t());
+    auto *RetVal =
+        reinterpret_cast<DPCTLBuildOptionListRef>(BuildOptionList.get());
+    BuildOptionList.release();
+    return RetVal;
 }
 
 void DPCTLBuildOptionList_Delete(__dpctl_take DPCTLBuildOptionListRef Ref)
 {
-    delete Ref;
+    delete reinterpret_cast<build_option_list_t *>(Ref);
 }
 
 void DPCTLBuildOptionList_Append(__dpctl_keep DPCTLBuildOptionListRef Ref,
                                  __dpctl_keep const char *Option)
 {
-    Ref->options.emplace_back(Option);
+    reinterpret_cast<build_option_list_t *>(Ref)->emplace_back(Option);
 }
 
+using kernel_name_list_t = std::vector<std::string>;
+
 __dpctl_give DPCTLKernelNameListRef DPCTLKernelNameList_Create()
 {
-    return new DPCTLKernelNameList;
+    auto KernelNameList =
+        std::unique_ptr<kernel_name_list_t>(new kernel_name_list_t());
+    auto *RetVal =
+        reinterpret_cast<DPCTLKernelNameListRef>(KernelNameList.get());
+    KernelNameList.release();
+    return RetVal;
 }
 
 void DPCTLKernelNameList_Delete(__dpctl_take DPCTLKernelNameListRef Ref)
 {
-    delete Ref;
+    delete reinterpret_cast<kernel_name_list_t *>(Ref);
 }
 
 void DPCTLKernelNameList_Append(__dpctl_keep DPCTLKernelNameListRef Ref,
                                 __dpctl_keep const char *Option)
 {
-    Ref->names.emplace_back(Option);
+    reinterpret_cast<kernel_name_list_t *>(Ref)->emplace_back(Option);
 }
 
+using virtual_header_list_t = std::vector<std::pair<std::string, std::string>>;
+
 __dpctl_give DPCTLVirtualHeaderListRef DPCTLVirtualHeaderList_Create()
 {
-    return new DPCTLVirtualHeaderList;
+    auto HeaderList =
+        std::unique_ptr<virtual_header_list_t>(new virtual_header_list_t());
+    auto *RetVal =
+        reinterpret_cast<DPCTLVirtualHeaderListRef>(HeaderList.get());
+    HeaderList.release();
+    return RetVal;
 }
 
 void DPCTLVirtualHeaderList_Delete(__dpctl_take DPCTLVirtualHeaderListRef Ref)
 {
-    delete Ref;
+    delete reinterpret_cast<virtual_header_list_t *>(Ref);
 }
 
 void DPCTLVirtualHeaderList_Append(__dpctl_keep DPCTLVirtualHeaderListRef Ref,
                                    __dpctl_keep const char *Name,
                                    __dpctl_keep const char *Content)
 {
     auto Header = std::make_pair<std::string, std::string>(Name, Content);
-    Ref->headers.push_back(Header);
+    reinterpret_cast<virtual_header_list_t *>(Ref)->push_back(Header);
 }
 
 namespace syclex = sycl::ext::oneapi::experimental;
 
+#if defined(SYCL_EXT_ONEAPI_KERNEL_COMPILER) &&                                \
+    defined(__SYCL_COMPILER_VERSION) && !defined(SUPPORTS_SYCL_COMPILATION)
+// SYCL source code compilation is supported from 2025.1 onwards.
+#if __SYCL_COMPILER_VERSION >= 20250317u
+#define SUPPORTS_SYCL_COMPILATION 1
+#else
+#define SUPPORTS_SYCL_COMPILATION 0
+#endif
+#endif
+
+#if (SUPPORTS_SYCL_COMPILATION > 0)
+#ifndef __SYCL_COMPILER_VERSION
+#error SYCL compiler version not defined
+#else
+// The property was renamed to `registered_names` after 2025.1
+#if __SYCL_COMPILER_VERSION > 20250317u
+using registered_names_property_t = syclex::registered_names;
+#else
+using registered_names_property_t = syclex::registered_kernel_names;
+#endif
+#endif
+#endif
+
 __dpctl_give DPCTLSyclKernelBundleRef DPCTLKernelBundle_CreateFromSYCLSource(
     __dpctl_keep const DPCTLSyclContextRef Ctx,
     __dpctl_keep const DPCTLSyclDeviceRef Dev,
@@ -822,15 +866,17 @@ __dpctl_give DPCTLSyclKernelBundleRef DPCTLKernelBundle_CreateFromSYCLSource(
     __dpctl_keep DPCTLKernelNameListRef Names,
     __dpctl_keep DPCTLBuildOptionListRef BuildOptions)
 {
-#ifdef SYCL_EXT_ONEAPI_KERNEL_COMPILER
+#if (SUPPORTS_SYCL_COMPILATION > 0)
     context *SyclCtx = unwrap<context>(Ctx);
     device *SyclDev = unwrap<device>(Dev);
     if (!SyclDev->ext_oneapi_can_compile(syclex::source_language::sycl)) {
         return nullptr;
     }
     try {
         syclex::include_files IncludeFiles;
-        for (auto &Include : Headers->headers) {
+        for (auto &Include :
+             *reinterpret_cast<virtual_header_list_t *>(Headers))
+        {
             const auto &[Name, Content] = Include;
             IncludeFiles.add(Name, Content);
         }
@@ -840,12 +886,15 @@ __dpctl_give DPCTLSyclKernelBundleRef DPCTLKernelBundle_CreateFromSYCLSource(
             *SyclCtx, syclex::source_language::sycl, Src,
             syclex::properties{IncludeFiles});
 
-        syclex::registered_names RegisteredNames;
-        for (const std::string &Name : Names->names) {
+        registered_names_property_t RegisteredNames;
+        for (const std::string &Name :
+             *reinterpret_cast<kernel_name_list_t *>(Names))
+        {
             RegisteredNames.add(Name);
         }
 
-        syclex::build_options Opts{BuildOptions->options};
+        syclex::build_options Opts{
+            *reinterpret_cast<build_option_list_t *>(BuildOptions)};
 
         std::vector<sycl::device> Devices({*SyclDev});
 
@@ -869,7 +918,7 @@ __dpctl_give DPCTLSyclKernelRef
 DPCTLKernelBundle_GetSyclKernel(__dpctl_keep DPCTLSyclKernelBundleRef KBRef,
                                 __dpctl_keep const char *KernelName)
 {
-#ifdef SYCL_EXT_ONEAPI_KERNEL_COMPILER
+#if (SUPPORTS_SYCL_COMPILATION > 0)
     try {
         auto KernelBundle =
             unwrap<sycl::kernel_bundle<bundle_state::executable>>(KBRef);
@@ -888,7 +937,7 @@ bool DPCTLKernelBundle_HasSyclKernel(__dpctl_keep DPCTLSyclKernelBundleRef
                                          KBRef,
                                      __dpctl_keep const char *KernelName)
 {
-#ifdef SYCL_EXT_ONEAPI_KERNEL_COMPILER
+#if (SUPPORTS_SYCL_COMPILATION > 0)
     try {
         auto KernelBundle =
             unwrap<sycl::kernel_bundle<bundle_state::executable>>(KBRef);