diff --git a/sycl/CMakeLists.txt b/sycl/CMakeLists.txt index 88697aab0a65b..6fec30df022c0 100644 --- a/sycl/CMakeLists.txt +++ b/sycl/CMakeLists.txt @@ -139,10 +139,12 @@ install(DIRECTORY "${sycl_inc_dir}/." DESTINATION "${LLVM_INST_INC_DIRECTORY}" C # SYCL runtime library add_subdirectory(source) +add_subdirectory(plugins) # SYCL toolchain builds all components: compiler, libraries, headers, etc. add_custom_target( sycl-toolchain DEPENDS sycl + pi_opencl clang clang-offload-wrapper clang-offload-bundler diff --git a/sycl/include/CL/sycl/detail/pi.hpp b/sycl/include/CL/sycl/detail/pi.hpp index 174992d86585a..db392ee853972 100644 --- a/sycl/include/CL/sycl/detail/pi.hpp +++ b/sycl/include/CL/sycl/detail/pi.hpp @@ -13,169 +13,176 @@ #include #include #include +#include +#include + +// Function to load the shared library +// Implementation is OS dependent. +void *loadOsLibrary(const std::string &library); + +// Function to get Address of a symbol defined in the shared +// library, Implementation is OS dependent. +void *getOsLibraryFuncAddress(void *Library, const std::string &FunctionName); namespace cl { namespace sycl { namespace detail { namespace pi { - // For selection of SYCL RT back-end, now manually through the "SYCL_BE" - // environment variable. - // - enum Backend { - SYCL_BE_PI_OPENCL, - SYCL_BE_PI_OTHER - }; - - // Check for manually selected BE at run-time. - bool useBackend(Backend Backend); - - using PiResult = ::pi_result; - using PiPlatform = ::pi_platform; - using PiDevice = ::pi_device; - using PiDeviceType = ::pi_device_type; - using PiDeviceInfo = ::pi_device_info; - using PiDeviceBinaryType = ::pi_device_binary_type; - using PiContext = ::pi_context; - using PiProgram = ::pi_program; - using PiKernel = ::pi_kernel; - using PiQueue = ::pi_queue; - using PiQueueProperties = ::pi_queue_properties; - using PiMem = ::pi_mem; - using PiMemFlags = ::pi_mem_flags; - using PiEvent = ::pi_event; - using PiSampler = ::pi_sampler; - using PiSamplerInfo = ::pi_sampler_info; - using PiSamplerProperties = ::pi_sampler_properties; - using PiSamplerAddressingMode = ::pi_sampler_addressing_mode; - using PiSamplerFilterMode = ::pi_sampler_filter_mode; - using PiMemImageFormat = ::pi_image_format; - using PiMemImageDesc = ::pi_image_desc; - using PiMemImageInfo = ::pi_image_info; - using PiMemObjectType = ::pi_mem_type; - using PiMemImageChannelOrder = ::pi_image_channel_order; - using PiMemImageChannelType = ::pi_image_channel_type; - - // Get a string representing a _pi_platform_info enum - std::string platformInfoToString(pi_platform_info info); - - // Report error and no return (keeps compiler happy about no return statements). - [[noreturn]] void die(const char *Message); - void assertion(bool Condition, const char *Message = nullptr); - - // Want all the needed casts be explicit, do not define conversion operators. - template - To cast(From value); - - // Forward declarations of the PI dispatch entries. +// For selection of SYCL RT back-end, now manually through the "SYCL_BE" +// environment variable. +// +enum Backend { SYCL_BE_PI_OPENCL, SYCL_BE_PI_OTHER }; + +#ifdef SYCL_RT_OS_WINDOWS +#define PLUGIN_NAME "pi_opencl.dll" +#else +#define PLUGIN_NAME "libpi_opencl.so" +#endif + +// Check for manually selected BE at run-time. +bool useBackend(Backend Backend); + +using PiResult = ::pi_result; +using PiPlatform = ::pi_platform; +using PiDevice = ::pi_device; +using PiDeviceType = ::pi_device_type; +using PiDeviceInfo = ::pi_device_info; +using PiDeviceBinaryType = ::pi_device_binary_type; +using PiContext = ::pi_context; +using PiProgram = ::pi_program; +using PiKernel = ::pi_kernel; +using PiQueue = ::pi_queue; +using PiQueueProperties = ::pi_queue_properties; +using PiMem = ::pi_mem; +using PiMemFlags = ::pi_mem_flags; +using PiEvent = ::pi_event; +using PiSampler = ::pi_sampler; +using PiSamplerInfo = ::pi_sampler_info; +using PiSamplerProperties = ::pi_sampler_properties; +using PiSamplerAddressingMode = ::pi_sampler_addressing_mode; +using PiSamplerFilterMode = ::pi_sampler_filter_mode; +using PiMemImageFormat = ::pi_image_format; +using PiMemImageDesc = ::pi_image_desc; +using PiMemImageInfo = ::pi_image_info; +using PiMemObjectType = ::pi_mem_type; +using PiMemImageChannelOrder = ::pi_image_channel_order; +using PiMemImageChannelType = ::pi_image_channel_type; + +// Get a string representing a _pi_platform_info enum +std::string platformInfoToString(pi_platform_info info); + +// Report error and no return (keeps compiler happy about no return statements). +[[noreturn]] void die(const char *Message); +void assertion(bool Condition, const char *Message = nullptr); + +// Want all the needed casts be explicit, do not define conversion operators. +template To cast(From value); + +// Forward declarations of the PI dispatch entries. #define _PI_API(api) __SYCL_EXPORTED extern decltype(::api) *(api); #include - // Performs PI one-time initialization. - void initialize(); - - // The PiCall helper structure facilitates performing a call to PI. - // It holds utilities to do the tracing and to check the returned result. - // TODO: implement a more mature and controllable tracing of PI calls. - class PiCall { - PiResult m_Result; - static bool m_TraceEnabled; - - public: - explicit PiCall(const char *Trace = nullptr); - ~PiCall(); - PiResult get(PiResult Result); - template - void check(PiResult Result); - }; - - // The run-time tracing of PI calls. - // TODO: replace PiCall completely with this one (PiTrace) - // - template inline - void print(T val) { - std::cout << " : " << val; - } +// Performs PI one-time initialization. +void initialize(); + +// The PiCall helper structure facilitates performing a call to PI. +// It holds utilities to do the tracing and to check the returned result. +// TODO: implement a more mature and controllable tracing of PI calls. +class PiCall { + PiResult m_Result; + static bool m_TraceEnabled; + +public: + explicit PiCall(const char *Trace = nullptr); + ~PiCall(); + PiResult get(PiResult Result); + template void check(PiResult Result); +}; + +// The run-time tracing of PI calls. +// TODO: replace PiCall completely with this one (PiTrace) +// +template inline void print(T val) { + std::cout << " : " << val; +} - template<> inline void print<> (PiPlatform val) { std::cout << "pi_platform : " << val; } - template<> inline void print<> (PiResult val) { - std::cout << "pi_result : "; - if (val == PI_SUCCESS) - std::cout << "PI_SUCCESS"; - else - std::cout << val; - } - - inline void printArgs(void) {} - template - void printArgs(Arg0 arg0, Args... args) { - std::cout << std::endl << " "; - print(arg0); - printArgs(std::forward(args)...); +template <> inline void print<>(PiPlatform val) { + std::cout << "pi_platform : " << val; +} +template <> inline void print<>(PiResult val) { + std::cout << "pi_result : "; + if (val == PI_SUCCESS) + std::cout << "PI_SUCCESS"; + else + std::cout << val; +} + +inline void printArgs(void) {} +template +void printArgs(Arg0 arg0, Args... args) { + std::cout << std::endl << " "; + print(arg0); + printArgs(std::forward(args)...); +} + +template class Trace { +private: + FnType m_FnPtr; + static bool m_TraceEnabled; + +public: + Trace(FnType FnPtr, const std::string &FnName) : m_FnPtr(FnPtr) { + if (m_TraceEnabled) + std::cout << "---> " << FnName << "("; } - - template - class Trace { - private: - FnType m_FnPtr; - static bool m_TraceEnabled; - public: - Trace(FnType FnPtr, const std::string &FnName) : m_FnPtr(FnPtr) { - if (m_TraceEnabled) - std::cout << "---> " << FnName << "("; - } - - template - typename std::result_of::type - operator() (Args... args) { - if (m_TraceEnabled) - printArgs(args...); - - initialize(); - auto r = m_FnPtr(args...); - - if (m_TraceEnabled) { - std::cout << ") ---> "; - std::cout << (print(r),"") << "\n"; - } - return r; + + template + typename std::result_of::type operator()(Args... args) { + if (m_TraceEnabled) + printArgs(args...); + + initialize(); + auto r = m_FnPtr(args...); + + if (m_TraceEnabled) { + std::cout << ") ---> "; + std::cout << (print(r), "") << "\n"; } - }; + return r; + } +}; - template - bool Trace::m_TraceEnabled = (std::getenv("SYCL_PI_TRACE") != nullptr); +template +bool Trace::m_TraceEnabled = (std::getenv("SYCL_PI_TRACE") != nullptr); } // namespace pi namespace RT = cl::sycl::detail::pi; -#define PI_ASSERT(cond, msg) \ - RT::assertion((cond), "assert: " msg); +#define PI_ASSERT(cond, msg) RT::assertion((cond), "assert: " msg); #define PI_TRACE(func) RT::Trace(func, #func) // This does the call, the trace and the check for no errors. -#define PI_CALL(pi) \ - RT::initialize(), \ - RT::PiCall(#pi).check( \ - RT::cast(pi)) +#define PI_CALL(pi) \ + RT::initialize(), RT::PiCall(#pi).check( \ + RT::cast(pi)) // This does the trace, the call, and returns the result -#define PI_CALL_RESULT(pi) \ - RT::PiCall(#pi).get(detail::RT::cast(pi)) +#define PI_CALL_RESULT(pi) \ + RT::PiCall(#pi).get(detail::RT::cast(pi)) // This does the check for no errors and possibly throws -#define PI_CHECK(pi) \ - RT::PiCall().check( \ - RT::cast(pi)) +#define PI_CHECK(pi) \ + RT::PiCall().check( \ + RT::cast(pi)) // This does the check for no errors and possibly throws x -#define PI_CHECK_THROW(pi, x) \ - RT::PiCall().check( \ - RT::cast(pi)) +#define PI_CHECK_THROW(pi, x) \ + RT::PiCall().check(RT::cast(pi)) // Want all the needed casts be explicit, do not define conversion operators. -template -To pi::cast(From value) { +template To pi::cast(From value) { // TODO: see if more sanity checks are possible. PI_ASSERT(sizeof(From) == sizeof(To), "cast failed size check"); return (To)(value); diff --git a/sycl/include/CL/sycl/detail/pi_offsets.h b/sycl/include/CL/sycl/detail/pi_offsets.h new file mode 100644 index 0000000000000..b33fbd4a16c03 --- /dev/null +++ b/sycl/include/CL/sycl/detail/pi_offsets.h @@ -0,0 +1,101 @@ +//==------- pi_functionoffsets.h - Plugin Interface Function Offsets ------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// +// #define NameoftheAPI Offset_to_add. +// Map the names of PI APIs to the corresponding offsets. This offset is added +// to the address of the Function Pointer list(returned by the Plugin) to get +// the location of the corresponding API Function Pointer implemented by the +// Plugin. Eg: Plugin returns address 0x0000 for the Function List. To get the +// function pointer for function piQueueCreate, we add 13*(8) = 104 to 0x0000 +// and access it. So (*0x0068) gives the function pointer to piQueueCreate +// implemented by the Plugin. Call is made using: +// (*0x0068)(context,device,properties,queue); + +#define FUNCTION_PTR_SIZE sizeof(void (*)()) + +// Platform + +#define piPlatformsGet_Offset 0 +#define piPlatformGetInfo_Offset 1 * FUNCTION_PTR_SIZE +// Device +#define piDevicesGet_Offset 2 * FUNCTION_PTR_SIZE +#define piDeviceGetInfo_Offset 3 * FUNCTION_PTR_SIZE +#define piDevicePartition_Offset 4 * FUNCTION_PTR_SIZE +#define piDeviceRetain_Offset 5 * FUNCTION_PTR_SIZE +#define piDeviceRelease_Offset 6 * FUNCTION_PTR_SIZE +#define piextDeviceSelectBinary_Offset 7 * FUNCTION_PTR_SIZE +#define piextGetDeviceFunctionPointer_Offset 8 * FUNCTION_PTR_SIZE +// Context +#define piContextCreate_Offset 9 * FUNCTION_PTR_SIZE +#define piContextGetInfo_Offset 10 * FUNCTION_PTR_SIZE +#define piContextRetain_Offset 11 * FUNCTION_PTR_SIZE +#define piContextRelease_Offset 12 * FUNCTION_PTR_SIZE +// Queue +#define piQueueCreate_Offset 13 * FUNCTION_PTR_SIZE +#define piQueueGetInfo_Offset 14 * FUNCTION_PTR_SIZE +#define piQueueFinish_Offset 15 * FUNCTION_PTR_SIZE +#define piQueueRetain_Offset 16 * FUNCTION_PTR_SIZE +#define piQueueRelease_Offset 17 * FUNCTION_PTR_SIZE +// Memory +#define piMemBufferCreate_Offset 18 * FUNCTION_PTR_SIZE +#define piMemImageCreate_Offset 19 * FUNCTION_PTR_SIZE +#define piMemGetInfo_Offset 20 * FUNCTION_PTR_SIZE +#define piMemImageGetInfo_Offset 21 * FUNCTION_PTR_SIZE +#define piMemRetain_Offset 22 * FUNCTION_PTR_SIZE +#define piMemRelease_Offset 23 * FUNCTION_PTR_SIZE +#define piMemBufferPartition_Offset 24 * FUNCTION_PTR_SIZE +// Program +#define piProgramCreate_Offset 25 * FUNCTION_PTR_SIZE +#define piclProgramCreateWithSource_Offset 26 * FUNCTION_PTR_SIZE +#define piclProgramCreateWithBinary_Offset 27 * FUNCTION_PTR_SIZE +#define piProgramGetInfo_Offset 28 * FUNCTION_PTR_SIZE +#define piProgramCompile_Offset 29 * FUNCTION_PTR_SIZE +#define piProgramBuild_Offset 30 * FUNCTION_PTR_SIZE +#define piProgramLink_Offset 31 * FUNCTION_PTR_SIZE +#define piProgramGetBuildInfo_Offset 32 * FUNCTION_PTR_SIZE +#define piProgramRetain_Offset 33 * FUNCTION_PTR_SIZE +#define piProgramRelease_Offset 34 * FUNCTION_PTR_SIZE +// Kernel +#define piKernelCreate_Offset 35 * FUNCTION_PTR_SIZE +#define piKernelSetArg_Offset 36 * FUNCTION_PTR_SIZE +#define piKernelGetInfo_Offset 37 * FUNCTION_PTR_SIZE +#define piKernelGetGroupInfo_Offset 38 * FUNCTION_PTR_SIZE +#define piKernelGetSubGroupInfo_Offset 39 * FUNCTION_PTR_SIZE +#define piKernelRetain_Offset 40 * FUNCTION_PTR_SIZE +#define piKernelRelease_Offset 41 * FUNCTION_PTR_SIZE +// Event +#define piEventCreate_Offset 42 * FUNCTION_PTR_SIZE +#define piEventGetInfo_Offset 43 * FUNCTION_PTR_SIZE +#define piEventGetProfilingInfo_Offset 44 * FUNCTION_PTR_SIZE +#define piEventsWait_Offset 45 * FUNCTION_PTR_SIZE +#define piEventSetCallback_Offset 46 * FUNCTION_PTR_SIZE +#define piEventSetStatus_Offset 47 * FUNCTION_PTR_SIZE +#define piEventRetain_Offset 48 * FUNCTION_PTR_SIZE +#define piEventRelease_Offset 49 * FUNCTION_PTR_SIZE +// Sampler +#define piSamplerCreate_Offset 50 * FUNCTION_PTR_SIZE +#define piSamplerGetInfo_Offset 51 * FUNCTION_PTR_SIZE +#define piSamplerRetain_Offset 52 * FUNCTION_PTR_SIZE +#define piSamplerRelease_Offset 53 * FUNCTION_PTR_SIZE +// Queue commands +#define piEnqueueKernelLaunch_Offset 54 * FUNCTION_PTR_SIZE +#define piEnqueueNativeKernel_Offset 55 * FUNCTION_PTR_SIZE +#define piEnqueueEventsWait_Offset 56 * FUNCTION_PTR_SIZE +#define piEnqueueMemBufferRead_Offset 57 * FUNCTION_PTR_SIZE +#define piEnqueueMemBufferReadRect_Offset 58 * FUNCTION_PTR_SIZE +#define piEnqueueMemBufferWrite_Offset 59 * FUNCTION_PTR_SIZE +#define piEnqueueMemBufferWriteRect_Offset 60 * FUNCTION_PTR_SIZE +#define piEnqueueMemBufferCopy_Offset 61 * FUNCTION_PTR_SIZE +#define piEnqueueMemBufferCopyRect_Offset 62 * FUNCTION_PTR_SIZE +#define piEnqueueMemBufferFill_Offset 63 * FUNCTION_PTR_SIZE +#define piEnqueueMemImageRead_Offset 64 * FUNCTION_PTR_SIZE +#define piEnqueueMemImageWrite_Offset 65 * FUNCTION_PTR_SIZE +#define piEnqueueMemImageCopy_Offset 66 * FUNCTION_PTR_SIZE +#define piEnqueueMemImageFill_Offset 67 * FUNCTION_PTR_SIZE +#define piEnqueueMemBufferMap_Offset 68 * FUNCTION_PTR_SIZE +#define piEnqueueMemUnmap_Offset 69 * FUNCTION_PTR_SIZE + diff --git a/sycl/plugins/CMakeLists.txt b/sycl/plugins/CMakeLists.txt new file mode 100644 index 0000000000000..ac0ced6f26bd5 --- /dev/null +++ b/sycl/plugins/CMakeLists.txt @@ -0,0 +1 @@ +add_subdirectory(opencl) diff --git a/sycl/plugins/opencl/CMakeLists.txt b/sycl/plugins/opencl/CMakeLists.txt new file mode 100644 index 0000000000000..49750db8febf2 --- /dev/null +++ b/sycl/plugins/opencl/CMakeLists.txt @@ -0,0 +1,59 @@ +#TODO: +#1. Figure out why CMP0057 has to be set. Should have been taken care of earlier in the build +#2. Use AddLLVM to modify the build and access config options +#cmake_policy(SET CMP0057 NEW) +#include(AddLLVM) + +# Plugin for OpenCL +# Create Shared library for libpi_opencl.so. +#TODO: remove dependency on pi.hpp in sycl project. +#TODO: Currently, the pi.hpp header is common between sycl and plugin library sources. +#This can be changed by copying the pi.hpp file in the plugins project. +add_library(pi_opencl SHARED + "${sycl_inc_dir}/CL/sycl/detail/pi.h" + "pi_opencl.cpp" + ) + +add_dependencies(pi_opencl + ocl-icd + ocl-headers +) + +set_target_properties(pi_opencl PROPERTIES LINKER_LANGUAGE CXX) + +#preprocessor definitions for compiling a target's sources. We do not need it for pi_opencl +target_include_directories(pi_opencl PRIVATE "${sycl_inc_dir}") + +#link pi_opencl with OpenCL headers and ICD Loader. +target_link_libraries( pi_opencl + PRIVATE OpenCL::Headers + PRIVATE ${OpenCL_LIBRARIES} +) + +if (SYCL_USE_LIBCXX) + if ((CMAKE_CXX_COMPILER_ID STREQUAL "GNU") OR + (CMAKE_CXX_COMPILER_ID STREQUAL "Clang")) + target_compile_options(pi_opencl PRIVATE -nostdinc++) + if ((NOT (DEFINED SYCL_LIBCXX_INCLUDE_PATH)) OR (NOT (DEFINED SYCL_LIBCXX_LIBRARY_PATH))) + message(FATAL_ERROR "When building with libc++ SYCL_LIBCXX_INCLUDE_PATHS and" + "SYCL_LIBCXX_LIBRARY_PATH should be set") + endif() + target_include_directories(pi_opencl PRIVATE "${SYCL_LIBCXX_INCLUDE_PATH}") + target_link_libraries(pi_opencl PRIVATE "-L${SYCL_LIBCXX_LIBRARY_PATH}" -nodefaultlibs -lc++ -lc) + else() + message(FATAL_ERROR "Build with libc++ is not yet supported for this compiler") + endif() +else() + +# Workaround for bug in GCC version 5 and higher. +# More information https://bugs.launchpad.net/ubuntu/+source/gcc-5/+bug/1568899 +if (CMAKE_CXX_COMPILER_ID STREQUAL "GNU" AND + CMAKE_CXX_COMPILER_VERSION VERSION_GREATER 5.0) + target_link_libraries(pi_opencl PRIVATE gcc_s gcc) +endif() + +endif() + +install(TARGETS pi_opencl + LIBRARY DESTINATION "lib" COMPONENT pi_opencl + RUNTIME DESTINATION "bin" COMPONENT pi_opencl) diff --git a/sycl/plugins/opencl/pi_opencl.cpp b/sycl/plugins/opencl/pi_opencl.cpp new file mode 100644 index 0000000000000..75e4ddffd12d0 --- /dev/null +++ b/sycl/plugins/opencl/pi_opencl.cpp @@ -0,0 +1,378 @@ +//==---------- pi_opencl.cpp - OpenCL Plugin -------------------------------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// +// This file implements an OpenCL Plugin, which is comformant to the Plugin +// Interface. The plugin implements a single function call which returns the +// location of the Plugin Interface Function Pointers list. The order of the +// functions is the same as the order of the functions declared in pi.h file. It +// is important to adhere to this order, as the plugin interface assumes this +// order when it computes the offset for a specific function. + +#include "CL/opencl.h" +#include +#include +#include +#include +#include + +#define CHECK_ERR_SET_NULL_RET(err, ptr, reterr) \ + if (err != CL_SUCCESS) { \ + if (ptr != nullptr) \ + *ptr = nullptr; \ + return cast(reterr); \ + } + +std::string SupportedVersion = "Version 1.1"; + +// Want all the needed casts be explicit, do not define conversion operators. +template To cast(From value) { + // TODO: see if more sanity checks are possible. + static_assert(sizeof(From) == sizeof(To) && "cast failed size check"); + return (To)(value); +} + +extern "C" { + +// Convenience macro makes source code search easier +#define OCL(pi_api) Ocl##pi_api + +// Example of a PI interface that does not map exactly to an OpenCL one. +pi_result OCL(piPlatformsGet)(pi_uint32 num_entries, pi_platform *platforms, + pi_uint32 *num_platforms) { + cl_int result = clGetPlatformIDs(cast(num_entries), + cast(platforms), + cast(num_platforms)); + + // Absorb the CL_PLATFORM_NOT_FOUND_KHR and just return 0 in num_platforms + if (result == CL_PLATFORM_NOT_FOUND_KHR) { + assert(num_platforms != 0); + *num_platforms = 0; + result = PI_SUCCESS; + } + return static_cast(result); +} + +// Example of a PI interface that does not map exactly to an OpenCL one. +pi_result OCL(piDevicesGet)(pi_platform platform, pi_device_type device_type, + pi_uint32 num_entries, pi_device *devices, + pi_uint32 *num_devices) { + cl_int result = clGetDeviceIDs( + cast(platform), cast(device_type), + cast(num_entries), cast(devices), + cast(num_devices)); + + // Absorb the CL_DEVICE_NOT_FOUND and just return 0 in num_devices + if (result == CL_DEVICE_NOT_FOUND) { + assert(num_devices != 0); + *num_devices = 0; + result = PI_SUCCESS; + } + return cast(result); +} + +pi_result OCL(piextDeviceSelectBinary)( + pi_device device, // TODO: does this need to be context? + pi_device_binary *images, pi_uint32 num_images, + pi_device_binary *selected_image) { + + // TODO dummy implementation. + // Real implementaion will use the same mechanism OpenCL ICD dispatcher + // uses. Something like: + // PI_VALIDATE_HANDLE_RETURN_HANDLE(ctx, PI_INVALID_CONTEXT); + // return context->dispatch->piextDeviceSelectIR( + // ctx, images, num_images, selected_image); + // where context->dispatch is set to the dispatch table provided by PI + // plugin for platform/device the ctx was created for. + + *selected_image = num_images > 0 ? images[0] : nullptr; + return PI_SUCCESS; +} + +pi_result OCL(piQueueCreate)(pi_context context, pi_device device, + pi_queue_properties properties, pi_queue *queue) { + assert(queue && "piQueueCreate failed, queue argument is null"); + + cl_platform_id curPlatform; + cl_int ret_err = + clGetDeviceInfo(cast(device), CL_DEVICE_PLATFORM, + sizeof(cl_platform_id), &curPlatform, NULL); + + CHECK_ERR_SET_NULL_RET(ret_err, queue, ret_err); + + size_t platVerSize; + ret_err = clGetPlatformInfo(curPlatform, CL_PLATFORM_VERSION, 0, NULL, + &platVerSize); + + CHECK_ERR_SET_NULL_RET(ret_err, queue, ret_err); + + std::string platVer(platVerSize, '\0'); + ret_err = clGetPlatformInfo(curPlatform, CL_PLATFORM_VERSION, platVerSize, + &platVer.front(), NULL); + + CHECK_ERR_SET_NULL_RET(ret_err, queue, ret_err); + + if (platVer.find("OpenCL 1.0") != std::string::npos || + platVer.find("OpenCL 1.1") != std::string::npos || + platVer.find("OpenCL 1.2") != std::string::npos) { + *queue = cast(clCreateCommandQueue( + cast(context), cast(device), + cast(properties), &ret_err)); + return cast(ret_err); + } + + cl_queue_properties CreationFlagProperties[] = { + CL_QUEUE_PROPERTIES, cast(properties), 0}; + *queue = cast(clCreateCommandQueueWithProperties( + cast(context), cast(device), + CreationFlagProperties, &ret_err)); + return cast(ret_err); +} + +pi_result OCL(piProgramCreate)(pi_context context, const void *il, + size_t length, pi_program *res_program) { + + size_t deviceCount; + + cl_int ret_err = clGetContextInfo(cast(context), + CL_CONTEXT_DEVICES, 0, NULL, &deviceCount); + + std::vector devicesInCtx(deviceCount); + + if (ret_err != CL_SUCCESS || deviceCount < 1) { + if (res_program != nullptr) + *res_program = nullptr; + return cast(CL_INVALID_CONTEXT); + } + + ret_err = clGetContextInfo(cast(context), CL_CONTEXT_DEVICES, + deviceCount * sizeof(cl_device_id), + devicesInCtx.data(), NULL); + + CHECK_ERR_SET_NULL_RET(ret_err, res_program, CL_INVALID_CONTEXT); + + cl_platform_id curPlatform; + ret_err = clGetDeviceInfo(devicesInCtx[0], CL_DEVICE_PLATFORM, + sizeof(cl_platform_id), &curPlatform, NULL); + + CHECK_ERR_SET_NULL_RET(ret_err, res_program, CL_INVALID_CONTEXT); + + size_t devVerSize; + ret_err = + clGetPlatformInfo(curPlatform, CL_PLATFORM_VERSION, 0, NULL, &devVerSize); + std::string devVer(devVerSize, '\0'); + ret_err = clGetPlatformInfo(curPlatform, CL_PLATFORM_VERSION, devVerSize, + &devVer.front(), NULL); + + CHECK_ERR_SET_NULL_RET(ret_err, res_program, CL_INVALID_CONTEXT); + + pi_result err = PI_SUCCESS; + if (devVer.find("OpenCL 1.0") == std::string::npos && + devVer.find("OpenCL 1.1") == std::string::npos && + devVer.find("OpenCL 1.2") == std::string::npos && + devVer.find("OpenCL 2.0") == std::string::npos) { + if (res_program != nullptr) + *res_program = cast(clCreateProgramWithIL( + cast(context), il, length, cast(&err))); + return err; + } + + size_t extSize; + ret_err = + clGetPlatformInfo(curPlatform, CL_PLATFORM_EXTENSIONS, 0, NULL, &extSize); + std::string extStr(extSize, '\0'); + ret_err = clGetPlatformInfo(curPlatform, CL_PLATFORM_EXTENSIONS, extSize, + &extStr.front(), NULL); + + if (ret_err != CL_SUCCESS || + extStr.find("cl_khr_il_program") == std::string::npos) { + if (res_program != nullptr) + *res_program = nullptr; + return cast(CL_INVALID_CONTEXT); + } + + using apiFuncT = + cl_program(CL_API_CALL *)(cl_context, const void *, size_t, cl_int *); + apiFuncT funcPtr = + reinterpret_cast(clGetExtensionFunctionAddressForPlatform( + curPlatform, "clCreateProgramWithILKHR")); + + assert(funcPtr != nullptr); + if (res_program != nullptr) + *res_program = cast( + funcPtr(cast(context), il, length, cast(&err))); + else + err = PI_INVALID_VALUE; + + return err; +} + +pi_result OCL(piSamplerCreate)(pi_context context, + const pi_sampler_properties *sampler_properties, + pi_sampler *result_sampler) { + // Initialize properties according to OpenCL 2.1 spec. + pi_result error_code; + pi_bool normalizedCoords = PI_TRUE; + pi_sampler_addressing_mode addressingMode = PI_SAMPLER_ADDRESSING_MODE_CLAMP; + pi_sampler_filter_mode filterMode = PI_SAMPLER_FILTER_MODE_NEAREST; + + // Unpack sampler properties + for (std::size_t i = 0; sampler_properties && sampler_properties[i] != 0; + ++i) { + if (sampler_properties[i] == PI_SAMPLER_INFO_NORMALIZED_COORDS) { + normalizedCoords = static_cast(sampler_properties[++i]); + } else if (sampler_properties[i] == PI_SAMPLER_INFO_ADDRESSING_MODE) { + addressingMode = + static_cast(sampler_properties[++i]); + } else if (sampler_properties[i] == PI_SAMPLER_INFO_FILTER_MODE) { + filterMode = static_cast(sampler_properties[++i]); + } else { + assert(false && "Cannot recognize sampler property"); + } + } + + // Always call OpenCL 1.0 API + *result_sampler = cast( + clCreateSampler(cast(context), normalizedCoords, + addressingMode, filterMode, cast(&error_code))); + return error_code; +} + +pi_result OCL(piextGetDeviceFunctionPointer)(pi_device device, + pi_program program, + const char *func_name, + pi_uint64 *function_pointer_ret) { + pi_platform platform; + cl_int ret_err = + clGetDeviceInfo(cast(device), PI_DEVICE_INFO_PLATFORM, + sizeof(platform), &platform, nullptr); + + if (ret_err != CL_SUCCESS) { + return cast(ret_err); + } + + using FuncT = + cl_int(CL_API_CALL *)(cl_device_id, cl_program, const char *, cl_ulong *); + + // TODO: add check that device supports corresponding extension + FuncT func_ptr = + reinterpret_cast(clGetExtensionFunctionAddressForPlatform( + cast(platform), "clGetDeviceFunctionPointerINTEL")); + // TODO: once we have check that device supports corresponding extension, + // we can insert an assertion that func_ptr is not nullptr. For now, let's + // just return an error if failed to query such function + // assert( + // func_ptr != nullptr && + // "Failed to get address of clGetDeviceFunctionPointerINTEL function"); + + if (!func_ptr) { + if (function_pointer_ret) + *function_pointer_ret = 0; + return PI_INVALID_DEVICE; + } + + return cast(func_ptr(cast(device), + cast(program), func_name, + function_pointer_ret)); +} + +// Plugin Interface Functions List. +struct PluginInterfaceFunctions { +// TODO: Remove the 'OclPtr' extension used with the PI_APIs. +// Forward calls to OpenCL RT. +#define _PI_CL(pi_api, ocl_api) \ + decltype(::pi_api) *pi_api##OclPtr = cast(&ocl_api); + + // Platform + _PI_CL(piPlatformsGet, OCL(piPlatformsGet)) + _PI_CL(piPlatformGetInfo, clGetPlatformInfo) + // Device + _PI_CL(piDevicesGet, OCL(piDevicesGet)) + _PI_CL(piDeviceGetInfo, clGetDeviceInfo) + _PI_CL(piDevicePartition, clCreateSubDevices) + _PI_CL(piDeviceRetain, clRetainDevice) + _PI_CL(piDeviceRelease, clReleaseDevice) + _PI_CL(piextDeviceSelectBinary, OCL(piextDeviceSelectBinary)) + _PI_CL(piextGetDeviceFunctionPointer, OCL(piextGetDeviceFunctionPointer)) + // Context + _PI_CL(piContextCreate, clCreateContext) + _PI_CL(piContextGetInfo, clGetContextInfo) + _PI_CL(piContextRetain, clRetainContext) + _PI_CL(piContextRelease, clReleaseContext) + // Queue + _PI_CL(piQueueCreate, OCL(piQueueCreate)) + _PI_CL(piQueueGetInfo, clGetCommandQueueInfo) + _PI_CL(piQueueFinish, clFinish) + _PI_CL(piQueueRetain, clRetainCommandQueue) + _PI_CL(piQueueRelease, clReleaseCommandQueue) + // Memory + _PI_CL(piMemBufferCreate, clCreateBuffer) + _PI_CL(piMemImageCreate, clCreateImage) + _PI_CL(piMemGetInfo, clGetMemObjectInfo) + _PI_CL(piMemImageGetInfo, clGetImageInfo) + _PI_CL(piMemRetain, clRetainMemObject) + _PI_CL(piMemRelease, clReleaseMemObject) + _PI_CL(piMemBufferPartition, clCreateSubBuffer) + // Program + _PI_CL(piProgramCreate, OCL(piProgramCreate)) + _PI_CL(piclProgramCreateWithSource, clCreateProgramWithSource) + _PI_CL(piclProgramCreateWithBinary, clCreateProgramWithBinary) + _PI_CL(piProgramGetInfo, clGetProgramInfo) + _PI_CL(piProgramCompile, clCompileProgram) + _PI_CL(piProgramBuild, clBuildProgram) + _PI_CL(piProgramLink, clLinkProgram) + _PI_CL(piProgramGetBuildInfo, clGetProgramBuildInfo) + _PI_CL(piProgramRetain, clRetainProgram) + _PI_CL(piProgramRelease, clReleaseProgram) + // Kernel + _PI_CL(piKernelCreate, clCreateKernel) + _PI_CL(piKernelSetArg, clSetKernelArg) + _PI_CL(piKernelGetInfo, clGetKernelInfo) + _PI_CL(piKernelGetGroupInfo, clGetKernelWorkGroupInfo) + _PI_CL(piKernelGetSubGroupInfo, clGetKernelSubGroupInfo) + _PI_CL(piKernelRetain, clRetainKernel) + _PI_CL(piKernelRelease, clReleaseKernel) + // Event + _PI_CL(piEventCreate, clCreateUserEvent) + _PI_CL(piEventGetInfo, clGetEventInfo) + _PI_CL(piEventGetProfilingInfo, clGetEventProfilingInfo) + _PI_CL(piEventsWait, clWaitForEvents) + _PI_CL(piEventSetCallback, clSetEventCallback) + _PI_CL(piEventSetStatus, clSetUserEventStatus) + _PI_CL(piEventRetain, clRetainEvent) + _PI_CL(piEventRelease, clReleaseEvent) + // Sampler + _PI_CL(piSamplerCreate, OCL(piSamplerCreate)) + _PI_CL(piSamplerGetInfo, clGetSamplerInfo) + _PI_CL(piSamplerRetain, clRetainSampler) + _PI_CL(piSamplerRelease, clReleaseSampler) + // Queue commands + _PI_CL(piEnqueueKernelLaunch, clEnqueueNDRangeKernel) + _PI_CL(piEnqueueNativeKernel, clEnqueueNativeKernel) + _PI_CL(piEnqueueEventsWait, clEnqueueMarkerWithWaitList) + _PI_CL(piEnqueueMemBufferRead, clEnqueueReadBuffer) + _PI_CL(piEnqueueMemBufferReadRect, clEnqueueReadBufferRect) + _PI_CL(piEnqueueMemBufferWrite, clEnqueueWriteBuffer) + _PI_CL(piEnqueueMemBufferWriteRect, clEnqueueWriteBufferRect) + _PI_CL(piEnqueueMemBufferCopy, clEnqueueCopyBuffer) + _PI_CL(piEnqueueMemBufferCopyRect, clEnqueueCopyBufferRect) + _PI_CL(piEnqueueMemBufferFill, clEnqueueFillBuffer) + _PI_CL(piEnqueueMemImageRead, clEnqueueReadImage) + _PI_CL(piEnqueueMemImageWrite, clEnqueueWriteImage) + _PI_CL(piEnqueueMemImageCopy, clEnqueueCopyImage) + _PI_CL(piEnqueueMemImageFill, clEnqueueFillImage) + _PI_CL(piEnqueueMemBufferMap, clEnqueueMapBuffer) + _PI_CL(piEnqueueMemUnmap, clEnqueueUnmapMemObject) +} FunctionTable; + +void *initialize_pi_opencl(char *RetSuppVersion) { + RetSuppVersion = strcpy((char *)malloc(strlen(SupportedVersion.c_str()) + 1), + SupportedVersion.c_str()); + return &FunctionTable; +} + +#undef _PI_CL +} diff --git a/sycl/source/CMakeLists.txt b/sycl/source/CMakeLists.txt index 1e0dc19f55422..076f2369f29a6 100644 --- a/sycl/source/CMakeLists.txt +++ b/sycl/source/CMakeLists.txt @@ -4,6 +4,14 @@ #cmake_policy(SET CMP0057 NEW) #include(AddLLVM) +if (WIN32) + set(OS_PLUGIN_INTERFACE + detail/windows_pi.cpp) +else () + set(OS_PLUGIN_INTERFACE + detail/linux_pi.cpp) +endif () + add_library(sycl SHARED "${sycl_inc_dir}/CL/sycl.hpp" "detail/builtins_common.cpp" @@ -12,7 +20,6 @@ add_library(sycl SHARED "detail/builtins_math.cpp" "detail/builtins_relational.cpp" "detail/pi.cpp" - "detail/pi_opencl.cpp" "detail/common.cpp" "detail/context_impl.cpp" "detail/device_impl.cpp" @@ -56,8 +63,9 @@ add_library(sycl SHARED "sampler.cpp" "stream.cpp" "spirv_ops.cpp" + "${OS_PLUGIN_INTERFACE}" ) - +#To-Do: Remove dependency on icd loader and opencl headers. add_dependencies(sycl ocl-icd ocl-headers @@ -73,7 +81,9 @@ target_include_directories(sycl PRIVATE "${sycl_inc_dir}") target_link_libraries(sycl PRIVATE OpenCL::Headers PRIVATE ${OpenCL_LIBRARIES} + PRIVATE ${CMAKE_DL_LIBS} ) + if (SYCL_USE_LIBCXX) if ((CMAKE_CXX_COMPILER_ID STREQUAL "GNU") OR (CMAKE_CXX_COMPILER_ID STREQUAL "Clang")) diff --git a/sycl/source/detail/linux_pi.cpp b/sycl/source/detail/linux_pi.cpp new file mode 100644 index 0000000000000..91dbe7736c65f --- /dev/null +++ b/sycl/source/detail/linux_pi.cpp @@ -0,0 +1,11 @@ +#include +#include + +void *loadOsLibrary(const std::string &PluginPath) { + // TODO: Check if the option RTLD_NOW is correct. + return dlopen(PluginPath.c_str(), RTLD_NOW); +} + +void *getOsLibraryFuncAddress(void *Library, const std::string &FunctionName) { + return dlsym(Library, FunctionName.c_str()); +} diff --git a/sycl/source/detail/pi.cpp b/sycl/source/detail/pi.cpp index 0ad31763b9846..1f5e29272d329 100644 --- a/sycl/source/detail/pi.cpp +++ b/sycl/source/detail/pi.cpp @@ -7,9 +7,11 @@ //===----------------------------------------------------------------------===// #include #include +#include #include #include #include +#include namespace cl { namespace sycl { @@ -37,22 +39,70 @@ std::string platformInfoToString(pi_platform_info info) { // Check for manually selected BE at run-time. bool useBackend(Backend TheBackend) { static const char *GetEnv = std::getenv("SYCL_BE"); + // Current default backend as SYCL_BE_PI_OPENCL + // Valid values of GetEnv are "PI_OPENCL" and "PI_OTHER" + std::string StringGetEnv = (GetEnv ? GetEnv : "PI_OPENCL"); static const Backend Use = - std::map{ - { "PI_OPENCL", SYCL_BE_PI_OPENCL }, - { "PI_OTHER", SYCL_BE_PI_OTHER } - // Any other value would yield PI_OPENCL (current default) - }[ GetEnv ? GetEnv : "PI_OPENCL"]; + (StringGetEnv == "PI_OTHER" ? SYCL_BE_PI_OTHER : SYCL_BE_PI_OPENCL); return TheBackend == Use; } // Definitions of the PI dispatch entries, they will be initialized // at their first use with piInitialize. -#define _PI_API(api) decltype(::api) * api = nullptr; +// ::api are defined in pi.h as Functions. +#define _PI_API(api) decltype(::api) *api = nullptr; #include -// TODO: implement real plugins (ICD-like?) -// For now this has the effect of redirecting to built-in PI OpenCL plugin. +// Find the plugin at the appropriate location and return the location in +// PluginPath +// TODO: Change the function appropriately when there are multiple plugins. +std::string findPlugin() { + // TODO: Based on final design discussions, change the location where the + // plugin must be searched; how to identify the plugins etc. Currently the + // search is done for libpi_opencl.so/pi_opencl.dll file in LD_LIBRARY_PATH + // env only. + return PLUGIN_NAME; +} + +// Load the Plugin by calling the OS dependent library loading call. +// Return the handle to the Library. +void *loadPlugin(const std::string &PluginPath) { + return loadOsLibrary(PluginPath); +} + +void *(*PluginInitFuncPtr)(char *); + +// Binds all the PI Interface APIs to Plugin Library Function Addresses. +// TODO: The plugin interface needs to setup infrastructure to route PI_CALLs to +// the appropriate plugins. Currently, we bind to a singe plugin. +bool bindPlugin(void *Library) { + decltype(PluginInitFuncPtr) InitializeFunction = + (decltype(PluginInitFuncPtr))( + getOsLibraryFuncAddress(Library, "initialize_pi_opencl")); + char *SupportedVersion; + // FuncTable is a list of all Interface Function pointers, where each + // Interface Function is located at a predetermined offset. + void *FuncTable = InitializeFunction(SupportedVersion); + +#define STRINGIZE(x) #x + +// At the predetermined "api"_Offset from the FunctionTable, the function +// pointer for "api" is stored. So we dereference the location to get the +// function pointer. +#define _PI_API(api) \ + api = *((decltype(&api))((char *)FuncTable + (api##_Offset))); + +#include + +#undef STRINGIZE +#undef _PI_API + return true; +} + +// Load the plugin based on SYCL_BE. +// TODO: Currently only accepting OpenCL plugins. Edit it to identify and load +// other kinds of plugins, do the required changes in the findPlugin, loadPlugin +// and bindPlugin functions. void initialize() { static bool Initialized = false; if (Initialized) { @@ -61,10 +111,22 @@ void initialize() { if (!useBackend(SYCL_BE_PI_OPENCL)) { die("Unknown SYCL_BE"); } - #define _PI_API(api) \ - extern decltype(::api) * api##OclPtr; \ - api = api##OclPtr; - #include + + std::string PluginPath = findPlugin(); + if (PluginPath.empty()) + die("Plugin Not Found."); + + void *Library = loadPlugin(PluginPath); + if (!Library) { + std::string Message = + "Check if plugin is present. Failed to load plugin: " + PluginPath; + die(Message.c_str()); + } + + if (!bindPlugin(Library)) { + std::string Message = "Failed to bind PI APIs to the plugin: " + PluginPath; + die(Message.c_str()); + } Initialized = true; } @@ -102,8 +164,7 @@ RT::PiResult PiCall::get(RT::PiResult Result) { m_Result = Result; return Result; } -template -void PiCall::check(RT::PiResult Result) { +template void PiCall::check(RT::PiResult Result) { m_Result = Result; // TODO: remove dependency on CHECK_OCL_CODE_THROW. CHECK_OCL_CODE_THROW(Result, Exception); diff --git a/sycl/source/detail/pi_opencl.cpp b/sycl/source/detail/pi_opencl.cpp deleted file mode 100644 index 9f21bfb3bf58a..0000000000000 --- a/sycl/source/detail/pi_opencl.cpp +++ /dev/null @@ -1,368 +0,0 @@ -//==---------- pi_opencl.cpp - OpenCL Plugin -------------------------------==// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// -#include "CL/opencl.h" -#include -#include -#include - -namespace cl { -namespace sycl { -namespace detail { -namespace pi { - -// Convinience macro makes source code search easier -#define OCL(pi_api) Ocl##pi_api - -// Example of a PI interface that does not map exactly to an OpenCL one. -pi_result OCL(piPlatformsGet)(pi_uint32 num_entries, - pi_platform * platforms, - pi_uint32 * num_platforms) { - cl_int result = - clGetPlatformIDs(cast (num_entries), - cast (platforms), - cast (num_platforms)); - - // Absorb the CL_PLATFORM_NOT_FOUND_KHR and just return 0 in num_platforms - if (result == CL_PLATFORM_NOT_FOUND_KHR) { - assertion(num_platforms != 0); - *num_platforms = 0; - result = PI_SUCCESS; - } - return cast(result); -} - - -// Example of a PI interface that does not map exactly to an OpenCL one. -pi_result OCL(piDevicesGet)(pi_platform platform, - pi_device_type device_type, - pi_uint32 num_entries, - pi_device * devices, - pi_uint32 * num_devices) { - cl_int result = - clGetDeviceIDs(cast (platform), - cast (device_type), - cast (num_entries), - cast (devices), - cast (num_devices)); - - // Absorb the CL_DEVICE_NOT_FOUND and just return 0 in num_devices - if (result == CL_DEVICE_NOT_FOUND) { - assertion(num_devices != 0); - *num_devices = 0; - result = PI_SUCCESS; - } - return cast(result); -} - -pi_result OCL(piextDeviceSelectBinary)( - pi_device device, // TODO: does this need to be context? - pi_device_binary * images, - pi_uint32 num_images, - pi_device_binary * selected_image) { - - // TODO dummy implementation. - // Real implementaion will use the same mechanism OpenCL ICD dispatcher - // uses. Somthing like: - // PI_VALIDATE_HANDLE_RETURN_HANDLE(ctx, PI_INVALID_CONTEXT); - // return context->dispatch->piextDeviceSelectIR( - // ctx, images, num_images, selected_image); - // where context->dispatch is set to the dispatch table provided by PI - // plugin for platform/device the ctx was created for. - - *selected_image = num_images > 0 ? images[0] : nullptr; - return PI_SUCCESS; -} - -pi_result OCL(piQueueCreate)(pi_context context, pi_device device, - pi_queue_properties properties, pi_queue *queue) { - PI_ASSERT(queue, "piQueueCreate failed, queue argument is null"); - - cl_platform_id curPlatform; - cl_int ret_err = clGetDeviceInfo(cast(device), - CL_DEVICE_PLATFORM, sizeof(cl_platform_id), - &curPlatform, NULL); - - if (ret_err != CL_SUCCESS) { - *queue = nullptr; - return cast(ret_err); - } - - size_t platVerSize; - ret_err = clGetPlatformInfo(curPlatform, CL_PLATFORM_VERSION, 0, NULL, - &platVerSize); - - if (ret_err != CL_SUCCESS) { - *queue = nullptr; - return cast(ret_err); - } - - std::string platVer(platVerSize, '\0'); - ret_err = clGetPlatformInfo(curPlatform, CL_PLATFORM_VERSION, platVerSize, - &platVer.front(), NULL); - - if (ret_err != CL_SUCCESS) { - *queue = nullptr; - return cast(ret_err); - } - - if (platVer.find("OpenCL 1.0") != std::string::npos || - platVer.find("OpenCL 1.1") != std::string::npos || - platVer.find("OpenCL 1.2") != std::string::npos) { - *queue = cast(clCreateCommandQueue( - cast(context), cast(device), - cast(properties), &ret_err)); - return cast(ret_err); - } - - cl_queue_properties CreationFlagProperties[] = { - CL_QUEUE_PROPERTIES, cast(properties), 0 - }; - *queue = cast(clCreateCommandQueueWithProperties( - cast(context), - cast(device), - CreationFlagProperties, &ret_err)); - return cast(ret_err); -} - -pi_result OCL(piProgramCreate)(pi_context context, const void *il, - size_t length, pi_program *res_program) { - - size_t deviceCount; - - cl_int ret_err = clGetContextInfo(cast(context), - CL_CONTEXT_DEVICES, 0, NULL, &deviceCount); - - std::vector devicesInCtx(deviceCount); - - ret_err = clGetContextInfo(cast(context), CL_CONTEXT_DEVICES, - deviceCount * sizeof(cl_device_id), - devicesInCtx.data(), NULL); - - if (ret_err != CL_SUCCESS || deviceCount < 1) { - if (res_program != nullptr) - *res_program = nullptr; - return cast(CL_INVALID_CONTEXT); - } - - cl_platform_id curPlatform; - ret_err = clGetDeviceInfo(devicesInCtx[0], CL_DEVICE_PLATFORM, - sizeof(cl_platform_id), &curPlatform, NULL); - - if (ret_err != CL_SUCCESS) { - if (res_program != nullptr) - *res_program = nullptr; - return cast(CL_INVALID_CONTEXT); - } - - size_t devVerSize; - ret_err = - clGetPlatformInfo(curPlatform, CL_PLATFORM_VERSION, 0, NULL, &devVerSize); - std::string devVer(devVerSize, '\0'); - ret_err = clGetPlatformInfo(curPlatform, CL_PLATFORM_VERSION, devVerSize, - &devVer.front(), NULL); - - if (ret_err != CL_SUCCESS) { - if (res_program != nullptr) - *res_program = nullptr; - return cast(CL_INVALID_CONTEXT); - } - - pi_result err = PI_SUCCESS; - if (devVer.find("OpenCL 1.0") == std::string::npos && - devVer.find("OpenCL 1.1") == std::string::npos && - devVer.find("OpenCL 1.2") == std::string::npos && - devVer.find("OpenCL 2.0") == std::string::npos) { - if (res_program != nullptr) - *res_program = cast(clCreateProgramWithIL( - cast(context), il, length, cast(&err))); - return err; - } - - size_t extSize; - ret_err = clGetPlatformInfo(curPlatform, CL_PLATFORM_EXTENSIONS, 0, NULL, - &extSize); - std::string extStr(extSize, '\0'); - ret_err = clGetPlatformInfo(curPlatform, CL_PLATFORM_EXTENSIONS, - extSize, &extStr.front(), NULL); - - if (ret_err != CL_SUCCESS || - extStr.find("cl_khr_il_program") == std::string::npos) { - if (res_program != nullptr) - *res_program = nullptr; - return cast(CL_INVALID_CONTEXT); - } - - using apiFuncT = - cl_program(CL_API_CALL *)(cl_context, const void *, size_t, cl_int *); - apiFuncT funcPtr = - reinterpret_cast(clGetExtensionFunctionAddressForPlatform( - curPlatform, "clCreateProgramWithILKHR")); - - assertion(funcPtr != nullptr); - if (res_program != nullptr) - *res_program = cast(funcPtr( - cast(context), il, length, cast(&err))); - - return err; -} - -pi_result OCL(piSamplerCreate)(pi_context context, - const pi_sampler_properties *sampler_properties, - pi_sampler *result_sampler) { - // Initialize properties according to OpenCL 2.1 spec. - pi_result error_code; - pi_bool normalizedCoords = PI_TRUE; - pi_sampler_addressing_mode addressingMode = PI_SAMPLER_ADDRESSING_MODE_CLAMP; - pi_sampler_filter_mode filterMode = PI_SAMPLER_FILTER_MODE_NEAREST; - - // Unpack sampler properties - for (std::size_t i = 0; sampler_properties && sampler_properties[i] != 0; - ++i) { - if (sampler_properties[i] == PI_SAMPLER_INFO_NORMALIZED_COORDS) { - normalizedCoords = static_cast(sampler_properties[++i]); - } else if (sampler_properties[i] == PI_SAMPLER_INFO_ADDRESSING_MODE) { - addressingMode = static_cast(sampler_properties[++i]); - } else if (sampler_properties[i] == PI_SAMPLER_INFO_FILTER_MODE) { - filterMode = static_cast(sampler_properties[++i]); - } else { - PI_ASSERT(false, "Cannot recognize sampler property"); - } - } - - // Always call OpenCL 1.0 API - *result_sampler = cast(clCreateSampler(cast(context), - normalizedCoords, addressingMode, filterMode, - cast(&error_code))); - return error_code; -} - -pi_result OCL(piextGetDeviceFunctionPointer)(pi_device device, - pi_program program, - const char *func_name, - pi_uint64 *function_pointer_ret) { - pi_platform platform; - PI_CALL(piDeviceGetInfo(device, PI_DEVICE_INFO_PLATFORM, sizeof(platform), - &platform, nullptr)); - using FuncT = - cl_int(CL_API_CALL *)(cl_device_id, cl_program, const char *, cl_ulong *); - - // TODO: add check that device supports corresponding extension - FuncT func_ptr = - reinterpret_cast(clGetExtensionFunctionAddressForPlatform( - cast(platform), - "clGetDeviceFunctionPointerINTEL")); - // TODO: once we have check that device supports corresponding extension, - // we can insert an assertion that func_ptr is not nullptr. For now, let's - // just return an error if failed to query such function - // PI_ASSERT( - // func_ptr != nullptr, - // "Failed to get address of clGetDeviceFunctionPointerINTEL function"); - - if (!func_ptr) { - if (function_pointer_ret) - *function_pointer_ret = 0; - return PI_INVALID_DEVICE; - } - - return PI_CALL_RESULT(func_ptr(cast(device), - cast(program), func_name, - function_pointer_ret)); -} - -// Forward calls to OpenCL RT. -#define _PI_CL(pi_api, ocl_api) \ -decltype(::pi_api) * pi_api##OclPtr = \ - detail::pi::cast(&ocl_api); - -// Platform -_PI_CL(piPlatformsGet, OCL(piPlatformsGet)) -_PI_CL(piPlatformGetInfo, clGetPlatformInfo) -// Device -_PI_CL(piDevicesGet, OCL(piDevicesGet)) -_PI_CL(piDeviceGetInfo, clGetDeviceInfo) -_PI_CL(piDevicePartition, clCreateSubDevices) -_PI_CL(piDeviceRetain, clRetainDevice) -_PI_CL(piDeviceRelease, clReleaseDevice) -_PI_CL(piextDeviceSelectBinary, OCL(piextDeviceSelectBinary)) -_PI_CL(piextGetDeviceFunctionPointer, OCL(piextGetDeviceFunctionPointer)) - // Context -_PI_CL(piContextCreate, clCreateContext) -_PI_CL(piContextGetInfo, clGetContextInfo) -_PI_CL(piContextRetain, clRetainContext) -_PI_CL(piContextRelease, clReleaseContext) -// Queue -_PI_CL(piQueueCreate, OCL(piQueueCreate)) -_PI_CL(piQueueGetInfo, clGetCommandQueueInfo) -_PI_CL(piQueueFinish, clFinish) -_PI_CL(piQueueRetain, clRetainCommandQueue) -_PI_CL(piQueueRelease, clReleaseCommandQueue) -// Memory -_PI_CL(piMemBufferCreate, clCreateBuffer) -_PI_CL(piMemImageCreate, clCreateImage) -_PI_CL(piMemGetInfo, clGetMemObjectInfo) -_PI_CL(piMemImageGetInfo, clGetImageInfo) -_PI_CL(piMemRetain, clRetainMemObject) -_PI_CL(piMemRelease, clReleaseMemObject) -_PI_CL(piMemBufferPartition, clCreateSubBuffer) -// Program -_PI_CL(piProgramCreate, OCL(piProgramCreate)) -_PI_CL(piclProgramCreateWithSource, clCreateProgramWithSource) -_PI_CL(piclProgramCreateWithBinary, clCreateProgramWithBinary) -_PI_CL(piProgramGetInfo, clGetProgramInfo) -_PI_CL(piProgramCompile, clCompileProgram) -_PI_CL(piProgramBuild, clBuildProgram) -_PI_CL(piProgramLink, clLinkProgram) -_PI_CL(piProgramGetBuildInfo, clGetProgramBuildInfo) -_PI_CL(piProgramRetain, clRetainProgram) -_PI_CL(piProgramRelease, clReleaseProgram) -// Kernel -_PI_CL(piKernelCreate, clCreateKernel) -_PI_CL(piKernelSetArg, clSetKernelArg) -_PI_CL(piKernelGetInfo, clGetKernelInfo) -_PI_CL(piKernelGetGroupInfo, clGetKernelWorkGroupInfo) -_PI_CL(piKernelGetSubGroupInfo, clGetKernelSubGroupInfo) -_PI_CL(piKernelRetain, clRetainKernel) -_PI_CL(piKernelRelease, clReleaseKernel) -// Event -_PI_CL(piEventCreate, clCreateUserEvent) -_PI_CL(piEventGetInfo, clGetEventInfo) -_PI_CL(piEventGetProfilingInfo, clGetEventProfilingInfo) -_PI_CL(piEventsWait, clWaitForEvents) -_PI_CL(piEventSetCallback, clSetEventCallback) -_PI_CL(piEventSetStatus, clSetUserEventStatus) -_PI_CL(piEventRetain, clRetainEvent) -_PI_CL(piEventRelease, clReleaseEvent) -// Sampler -_PI_CL(piSamplerCreate, OCL(piSamplerCreate)) -_PI_CL(piSamplerGetInfo, clGetSamplerInfo) -_PI_CL(piSamplerRetain, clRetainSampler) -_PI_CL(piSamplerRelease, clReleaseSampler) -// Queue commands -_PI_CL(piEnqueueKernelLaunch, clEnqueueNDRangeKernel) -_PI_CL(piEnqueueNativeKernel, clEnqueueNativeKernel) -_PI_CL(piEnqueueEventsWait, clEnqueueMarkerWithWaitList) -_PI_CL(piEnqueueMemBufferRead, clEnqueueReadBuffer) -_PI_CL(piEnqueueMemBufferReadRect, clEnqueueReadBufferRect) -_PI_CL(piEnqueueMemBufferWrite, clEnqueueWriteBuffer) -_PI_CL(piEnqueueMemBufferWriteRect, clEnqueueWriteBufferRect) -_PI_CL(piEnqueueMemBufferCopy, clEnqueueCopyBuffer) -_PI_CL(piEnqueueMemBufferCopyRect, clEnqueueCopyBufferRect) -_PI_CL(piEnqueueMemBufferFill, clEnqueueFillBuffer) -_PI_CL(piEnqueueMemImageRead, clEnqueueReadImage) -_PI_CL(piEnqueueMemImageWrite, clEnqueueWriteImage) -_PI_CL(piEnqueueMemImageCopy, clEnqueueCopyImage) -_PI_CL(piEnqueueMemImageFill, clEnqueueFillImage) -_PI_CL(piEnqueueMemBufferMap, clEnqueueMapBuffer) -_PI_CL(piEnqueueMemUnmap, clEnqueueUnmapMemObject) - -#undef _PI_CL - -} // namespace pi -} // namespace detail -} // namespace sycl -} // namespace cl diff --git a/sycl/source/detail/windows_pi.cpp b/sycl/source/detail/windows_pi.cpp new file mode 100644 index 0000000000000..90ea0812c051b --- /dev/null +++ b/sycl/source/detail/windows_pi.cpp @@ -0,0 +1,11 @@ +#include +#include +#include + +void *loadOsLibrary(const std::string &PluginPath) { + return (void *)LoadLibraryA(PluginPath.c_str()); +} + +void *getOsLibraryFuncAddress(void *Library, const std::string &FunctionName) { + return GetProcAddress((HMODULE)Library, FunctionName.c_str()); +}