Add support for events, programs, and kernel submission to dpctl. (#80)

* Add support for events, programs, and kernel submission to dpctl.

    * New dpctl C API wrappers for sycl::event, sycl::kernel and sycl::program.
    * Wrappers to create an OpenCL interoperability program from either string
      or SPIR-V IL files.
    * Support for submitting an interoperability kernel to a sycl::queue. We
      support using either a range or an nd_range to submit the kernel.
    * Event wait function is now exposed via the C API.
    * New test cases for the C API.
    * The dpctl Cython wrapper now supports program creation and kernel
      submission.
    * Other refactoring to standardize function names, attribute names, comments.

* Fix windows build for dpctl.

* Update .gitignore

Add new line at end of file.

* Minor cosmetic changes

    * Remove empty doxygen comment blocks
    * Fix indentation
    * Fix grammatical typos

* Additional information is now printed when dpctl.dump is called.

* Fix test cases so that kernels are not submitted to default queue.

    - The default queue points to a Level-0 device is Level-0 drivers
      are installed. However, we always compile our program for OpenCL
      and this causes a crash if the test kernels are submitted to the
      default queue.

      Fixing it by selecting the GPU:0 queue that is known to be an
      OpenCL device. A follow up commit will fix this in a proper way.

* Add test files to package

* Update dpctl/tests/__init__.py

* Update dpctl/tests/test_sycl_kernel_submit.py

* Update dpctl/tests/test_sycl_program.py

* Fix comment.

* Another comment was fixed.

Co-authored-by: Sergey Pokhodenko <[email protected]>
Co-authored-by: Sergey Pokhodenko <[email protected]>

Author: diptorupd

MANIFEST.in

@@ -2,4 +2,5 @@ include versioneer.py
 recursive-include dpctl/include *.h *.hpp
 include dpctl/*.pxd
 include dpctl/*DPPL*Interface.*
+include dpctl/tests/input_files/*
 global-exclude *.cpp

backends/CMakeLists.txt

@@ -98,7 +98,10 @@ add_library(
   SHARED
   source/dppl_sycl_context_interface.cpp
   source/dppl_sycl_device_interface.cpp
+  source/dppl_sycl_event_interface.cpp
+  source/dppl_sycl_kernel_interface.cpp
   source/dppl_sycl_platform_interface.cpp
+  source/dppl_sycl_program_interface.cpp
   source/dppl_sycl_queue_interface.cpp
   source/dppl_sycl_queue_manager.cpp
   source/dppl_sycl_usm_interface.cpp
@@ -146,13 +149,12 @@ if(WIN32)
     )
     target_link_libraries(
         DPPLSyclInterface
-        PRIVATE
-        ${DPCPP_ROOT}/lib/sycl.lib
+        PRIVATE ${DPCPP_ROOT}/lib/sycl.lib
+        PRIVATE ${DPCPP_ROOT}/lib/OpenCL.lib
     )
     target_link_libraries(
         DPPLOpenCLInterface
-        PRIVATE
-        ${DPCPP_ROOT}/lib/OpenCL.lib
+        PRIVATE ${DPCPP_ROOT}/lib/OpenCL.lib
     )
 endif()
 

backends/dbg_build.sh

@@ -5,6 +5,7 @@ mkdir build
 pushd build
 
 INSTALL_PREFIX=`pwd`/../install
+rm -rf ${INSTALL_PREFIX}
 export ONEAPI_ROOT=/opt/intel/oneapi
 DPCPP_ROOT=${ONEAPI_ROOT}/compiler/latest/linux
 PYTHON_INC=`python -c "import distutils.sysconfig;                  \
@@ -24,9 +25,6 @@ cmake                                                       \
     -DGTEST_LIB_DIR=${CONDA_PREFIX}/lib                     \
     ..
 
-make V=1 -n -j 4
-make check
-make install
-
+make V=1 -n -j 4 && make check && make install
 
 popd

backends/include/dppl_sycl_event_interface.h

@@ -0,0 +1,55 @@
+//===--- dppl_sycl_event_interface.h - DPPL-SYCL interface ---*---C++ -*---===//
+//
+//               Python Data Parallel Processing Library (PyDPPL)
+//
+// Copyright 2020 Intel Corporation
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//    http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+//
+//===----------------------------------------------------------------------===//
+///
+/// \file
+/// This header declares a C API to a sub-set of the sycl::event interface.
+///
+//===----------------------------------------------------------------------===//
+
+#pragma once
+
+#include "dppl_data_types.h"
+#include "dppl_sycl_types.h"
+#include "Support/DllExport.h"
+#include "Support/ExternC.h"
+#include "Support/MemOwnershipAttrs.h"
+
+
+DPPL_C_EXTERN_C_BEGIN
+
+/*!
+ * @brief C-API wrapper for sycl::event.wait.
+ *
+ * @param    ERef           An opaque DPPLSyclEventRef pointer on which to wait.
+ */
+DPPL_API
+void DPPLEvent_Wait (__dppl_keep DPPLSyclEventRef ERef);
+
+/*!
+ * @brief Deletes the DPPLSyclEventRef after casting it to a sycl::event.
+ *
+ * @param    ERef           An opaque DPPLSyclEventRef pointer that would be
+ *                          freed.
+ */
+DPPL_API
+void
+DPPLEvent_Delete (__dppl_take DPPLSyclEventRef ERef);
+
+DPPL_C_EXTERN_C_END

backends/include/dppl_sycl_kernel_interface.h

@@ -0,0 +1,74 @@
+//===---- dppl_sycl_kernel_interface.h - DPPL-SYCL interface --*--C++ --*--===//
+//
+//               Python Data Parallel Processing Library (PyDPPL)
+//
+// Copyright 2020 Intel Corporation
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//    http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+//
+//===----------------------------------------------------------------------===//
+///
+/// \file
+/// This header declares a C API to create Sycl kernels from OpenCL kernels. In
+/// future, API to create interoperability kernels from other languages such as
+/// Level-0 driver API may be added here.
+///
+/// \todo Investigate what we should do when we add support for Level-0 API.
+///
+//===----------------------------------------------------------------------===//
+
+#pragma once
+
+#include "dppl_data_types.h"
+#include "dppl_sycl_types.h"
+#include "Support/DllExport.h"
+#include "Support/ExternC.h"
+#include "Support/MemOwnershipAttrs.h"
+
+DPPL_C_EXTERN_C_BEGIN
+
+/*!
+ * @brief Returns a C string for the kernel name.
+ *
+ * @param    KRef           DPPLSyclKernelRef pointer to an OpenCL
+ *                          interoperability kernel.
+ * @return   If a kernel name exists then returns it as a C string, else
+ *           returns a nullptr.
+ */
+DPPL_API
+__dppl_give const char*
+DPPLKernel_GetFunctionName (__dppl_keep const DPPLSyclKernelRef KRef);
+
+/*!
+ * @brief Returns the number of arguments for the OpenCL kernel.
+ *
+ * @param    KRef           DPPLSyclKernelRef pointer to an OpenCL
+ *                          interoperability kernel.
+ * @return   Returns the number of arguments for the OpenCL interoperability
+ *           kernel.
+ */
+DPPL_API
+size_t
+DPPLKernel_GetNumArgs (__dppl_keep const DPPLSyclKernelRef KRef);
+
+/*!
+ * @brief Deletes the DPPLSyclKernelRef after casting it to a sycl::kernel.
+ *
+ * @param    KRef           DPPLSyclKernelRef pointer to an OpenCL
+ *                          interoperability kernel.
+ */
+DPPL_API
+void
+DPPLKernel_Delete (__dppl_take DPPLSyclKernelRef KRef);
+
+DPPL_C_EXTERN_C_END

backends/include/dppl_sycl_program_interface.h

@@ -0,0 +1,114 @@
+//===---- dppl_sycl_program_interface.h - DPPL-SYCL interface --*--C++ --*--===//
+//
+//               Python Data Parallel Processing Library (PyDPPL)
+//
+// Copyright 2020 Intel Corporation
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//    http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+//
+//===----------------------------------------------------------------------===//
+///
+/// \file
+/// This header declares a C API to create Sycl program an interoperability
+/// program defined in OpenCL. In future, API to create interoperability
+/// kernels from other languages such as Level-0 driver API may be added here.
+///
+/// \todo Investigate what we should do when we add support for Level-0 API.
+///
+//===----------------------------------------------------------------------===//
+
+#pragma once
+
+#include "dppl_data_types.h"
+#include "dppl_sycl_types.h"
+#include "Support/DllExport.h"
+#include "Support/ExternC.h"
+#include "Support/MemOwnershipAttrs.h"
+
+DPPL_C_EXTERN_C_BEGIN
+
+/*!
+ * @brief Create a Sycl program from an OpenCL SPIR-V binary file.
+ *
+ * Sycl 1.2 does not expose any method to create a sycl::program from a SPIR-V
+ * IL file. To get around this limitation, we need to use the Sycl feature to
+ * create an interoperability kernel from an OpenCL kernel. This function first
+ * creates an OpenCL program and kernel from the SPIR-V binary and then using
+ * the Sycl-OpenCL interoperability feature creates a Sycl kernel from the
+ * OpenCL kernel.
+ *
+ * The feature to create a Sycl kernel from a SPIR-V IL binary will be available
+ * in Sycl 2.0 at which point this function may become deprecated.
+ *
+ * @param    Ctx            An opaque pointer to a sycl::context
+ * @param    IL             SPIR-V binary
+ * @return   A new SyclProgramRef pointer if the program creation succeeded,
+ *           else returns NULL.
+ */
+DPPL_API
+__dppl_give DPPLSyclProgramRef
+DPPLProgram_CreateFromOCLSpirv (__dppl_keep const DPPLSyclContextRef Ctx,
+                                __dppl_keep const void *IL,
+                                size_t Length);
+
+/*!
+ * @brief Create a Sycl program from an OpenCL kernel source string.
+ *
+ * @param    Ctx            An opaque pointer to a sycl::context
+ * @param    Source         OpenCL source string
+ * @param    CompileOptions Extra compiler flags (refer Sycl spec.)
+ * @return   A new SyclProgramRef pointer if the program creation succeeded,
+ *           else returns NULL.
+ */
+DPPL_API
+__dppl_give DPPLSyclProgramRef
+DPPLProgram_CreateFromOCLSource (__dppl_keep const DPPLSyclContextRef Ctx,
+                                 __dppl_keep const char *Source,
+                                 __dppl_keep const char *CompileOpts = nullptr);
+
+/*!
+ * @brief Returns the SyclKernel with given name from the program, if not found
+ * then return NULL.
+ *
+ * @param    PRef           Opaque pointer to a sycl::program
+ * @param    KernelName     Name of kernel
+ * @return   A SyclKernel reference if the kernel exists, else NULL
+ */
+DPPL_API
+__dppl_give DPPLSyclKernelRef
+DPPLProgram_GetKernel (__dppl_keep DPPLSyclProgramRef PRef,
+                       __dppl_keep const char *KernelName);
+
+/*!
+ * @brief Return True if a SyclKernel with given name exists in the program, if
+ * not found then returns False.
+ *
+ * @param    PRef           Opaque pointer to a sycl::program
+ * @param    KernelName     Name of kernel
+ * @return   True if the kernel exists, else False
+ */
+DPPL_API
+bool
+DPPLProgram_HasKernel (__dppl_keep DPPLSyclProgramRef PRef,
+                       __dppl_keep const char *KernelName);
+
+/*!
+ * @brief Frees the DPPLSyclProgramRef pointer.
+ *
+ * @param    PRef           Opaque pointer to a sycl::program
+ */
+DPPL_API
+void
+DPPLProgram_Delete (__dppl_take DPPLSyclProgramRef PRef);
+
+DPPL_C_EXTERN_C_END