From 7d98ed82b46f3f9e8b4ff974f28520973f6312cb Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Wed, 10 Mar 2021 09:10:02 +0300 Subject: [PATCH 01/32] Initial implementation --- sycl/doc/EnvironmentVariables.md | 7 +++++++ sycl/doc/KernelProgramCache.md | 3 ++- sycl/source/detail/config.def | 6 ++++++ sycl/source/detail/program_manager/program_manager.cpp | 2 +- 4 files changed, 16 insertions(+), 2 deletions(-) diff --git a/sycl/doc/EnvironmentVariables.md b/sycl/doc/EnvironmentVariables.md index 2b68844498c91..136e4ff66ff9a 100644 --- a/sycl/doc/EnvironmentVariables.md +++ b/sycl/doc/EnvironmentVariables.md @@ -34,6 +34,13 @@ subject to change. Do not rely on these variables in production code. | SYCL_DISABLE_PARALLEL_FOR_RANGE_ROUNDING | Any(\*) | Disables automatic rounding-up of parallel_for invocation ranges. | | SYCL_ENABLE_PCI | Integer | When set to 1, enables obtaining the GPU PCI address when using the Level Zero backend. The default is 0. | | SYCL_HOST_UNIFIED_MEMORY | Integer | Enforce host unified memory support or lack of it for the execution graph builder. If set to 0, it is enforced as not supported by all devices. If set to 1, it is enforced as supported by all devices. | +| SYCL_CACHE_DIR | Path | Path to persistent cache root directory. Default values are `%AppData%\Intel\sycl_program_cache` for Windows and `$HOME/intel/sycl_program_cache` on Linux. | +| SYCL_CACHE_ENABLED | ON, OFF | Switches persistent cache switch on or off. Default value is ON. | +| SYCL_CACHE_EVICTION | ON, OFF | Switches cache eviction on or off. Default value is ON. | +| SYCL_CACHE_MAX_SIZE | Positive integer | Cache eviction is triggered once total size of cached images exceeds the value in megabytes (default - 8 192 for 8 GB). Set to 0 to disable size-based cache eviction. | +| SYCL_CACHE_THRESHOLD | Positive integer | Cache eviction threshold in days (default value is 7 for 1 week). Set to 0 for disabling time-based cache eviction. | +| SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE | Positive integer | Minimum size of device code image in kilobytes which is reasonable to cache on disk because disk access operation may take more time than do JIT compilation for it. Default value is 0 to cache all images. | +| SYCL_CACHE_MAX_DEVICE_IMAGE_SIZE | Positive integer | Maximum size of device image in megabytes which is cached. Too big kernels may overload disk too fast. Default value is 0 to cache all images. | `(*) Note: Any means this environment variable is effective when set to any non-null value.` diff --git a/sycl/doc/KernelProgramCache.md b/sycl/doc/KernelProgramCache.md index b75f26b507fa7..3b634b709c61d 100644 --- a/sycl/doc/KernelProgramCache.md +++ b/sycl/doc/KernelProgramCache.md @@ -176,7 +176,8 @@ There is set of configuration parameters which can be set as environment variabl | Environment variable | Values | Description | | -------------------- | ------ | ----------- | | `SYCL_CACHE_DIR`| Path | Path to persistent cache root directory. Default values are `%AppData%\Intel\sycl_program_cache` for Windows and `$HOME/intel/sycl_program_cache` on Linux. | -| `SYCL_CACHE_ENABLED` | ON, OFF | Switches persistent cache switch on or off. Default value is ON. | +| `SYCL_CACHE_ENABLED` | ON, OFF | Switches persistent cache switch on or off. Default value is ON. | +| `SYCL_CACHE_EVICTION` | ON, OFF | Switches cache eviction on or off. Default value is ON. | | `SYCL_CACHE_MAX_SIZE` | Positive integer | Cache eviction is triggered once total size of cached images exceeds the value in megabytes (default - 8 192 for 8 GB). Set to 0 to disable size-based cache eviction. | | `SYCL_CACHE_THRESHOLD` | Positive integer | Cache eviction threshold in days (default value is 7 for 1 week). Set to 0 for disabling time-based cache eviction. | | `SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE` | Positive integer | Minimum size of device code image in kilobytes which is reasonable to cache on disk because disk access operation may take more time than do JIT compilation for it. Default value is 0 to cache all images. | diff --git a/sycl/source/detail/config.def b/sycl/source/detail/config.def index 6404f6508360a..725dedf20f099 100644 --- a/sycl/source/detail/config.def +++ b/sycl/source/detail/config.def @@ -20,3 +20,9 @@ CONFIG(SYCL_DEVICE_FILTER, 1024, __SYCL_DEVICE_FILTER) CONFIG(SYCL_PROGRAM_LINK_OPTIONS, 64, __SYCL_PROGRAM_LINK_OPTIONS) CONFIG(SYCL_PROGRAM_COMPILE_OPTIONS, 64, __SYCL_PROGRAM_COMPILE_OPTIONS) CONFIG(SYCL_HOST_UNIFIED_MEMORY, 1, __SYCL_HOST_UNIFIED_MEMORY) +CONFIG(SYCL_CACHE_DIR, 1024, __SYCL_CACHE_DIR) +CONFIG(SYCL_CACHE_ENABLED, 3, __SYCL_CACHE_ENABLED) +CONFIG(SYCL_CACHE_MAX_SIZE, 16, __SYCL_CACHE_MAX_SIZE) +CONFIG(SYCL_CACHE_THRESHOLD, 16, __SYCL_CACHE_THRESHOLD) +CONFIG(SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE, 16, __SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE) +CONFIG(SYCL_CACHE_MAX_DEVICE_IMAGE_SIZE, 16, __SYCL_CACHE_MAX_DEVICE_IMAGE_SIZE) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index e653156697464..feb59b7aaf277 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -41,7 +41,7 @@ namespace detail { using ContextImplPtr = std::shared_ptr; -static constexpr int DbgProgMgr = 0; +static constexpr int DbgProgMgr = 1; enum BuildState { BS_InProgress, BS_Done, BS_Failed }; From 6854e842e9a53c5c2467713ce33f40011bae86a2 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Thu, 11 Mar 2021 11:06:56 +0300 Subject: [PATCH 02/32] [SYCL] Remove redundant build options processing After 86b0e8d5 patch extra operations with device images happen before check if it is present in in-memory cache. For applications with small kernels which are executed multiple times noticeable performance degradation is seen for host code. That was done to get build options stored in the kernel image and use them as in-memory cache key. At the same time kernel image (where these options are taken from) is used in cache key so it is reasonable to use only build options which are specified in SYCL API and/or environment variables as separate cache key. Getting build options from kernel image is moved back to build operation which happens only if built program is missed in in-memory cache. --- .../program_manager/program_manager.cpp | 82 ++++++++++--------- 1 file changed, 43 insertions(+), 39 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index feb59b7aaf277..6fd38fb71ec8e 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -367,35 +367,50 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, auto GetF = [](const Locked &LockedCache) -> ProgramCacheT & { return LockedCache.get(); }; - std::string BuildOptions; - if (Prg) - BuildOptions = Prg->get_build_options(); - const RTDeviceBinaryImage &Img = - getDeviceImage(M, KSId, Context, Device, JITCompilationIsRequired); - std::string CompileOpts = Img.getCompileOptions(); - std::string LinkOpts = Img.getLinkOptions(); - pi_device_binary_property isEsimdImage = Img.getProperty("isEsimdImage"); - if (!BuildOptions.empty()) { - CompileOpts += " "; - CompileOpts += BuildOptions; - } - if (isEsimdImage && pi::DeviceBinaryProperty(isEsimdImage).asUint32()) { - if (!CompileOpts.empty()) - CompileOpts += " "; - CompileOpts += "-vc-codegen"; - } - // Build options are overridden if environment variables are present - const char *CompileOptsEnv = SYCLConfig::get(); + std::string CompileOpts; + std::string LinkOpts; + // Build options are overridden if environment variables are present. + // Environment variables are not changed during program lifecycle so it + // is reasonable to use static here to read them only once. + static const char *CompileOptsEnv = + SYCLConfig::get(); if (CompileOptsEnv) { CompileOpts = CompileOptsEnv; + } else { // Use build options only when the environment variable is missed + if (Prg) { + std::string BuildOptions = Prg->get_build_options(); + if (!BuildOptions.empty()) { + CompileOpts += " "; + CompileOpts += BuildOptions; + } + } } - const char *LinkOptsEnv = SYCLConfig::get(); + static const char *LinkOptsEnv = SYCLConfig::get(); if (LinkOptsEnv) { LinkOpts = LinkOptsEnv; } - auto BuildF = [this, &Context, &Device, Prg, &Img, &CompileOpts, &LinkOpts] { + auto BuildF = [this, &M, &KSId, &Context, &Device, Prg, &CompileOpts, + &LinkOpts, &JITCompilationIsRequired] { + const RTDeviceBinaryImage &Img = + getDeviceImage(M, KSId, Context, Device, JITCompilationIsRequired); + // Update only if compile options are not overwritten by environment + // variable + if (!CompileOptsEnv) { + CompileOpts += Img.getCompileOptions(); + pi_device_binary_property isEsimdImage = Img.getProperty("isEsimdImage"); + + if (isEsimdImage && pi::DeviceBinaryProperty(isEsimdImage).asUint32()) { + if (!CompileOpts.empty()) + CompileOpts += " "; + CompileOpts += "-vc-codegen"; + } + } + + // Update only if link options are not overwritten by environment variable + if (!LinkOptsEnv) + LinkOpts += Img.getLinkOptions(); ContextImplPtr ContextImpl = getSyclObjImpl(Context); const detail::plugin &Plugin = ContextImpl->getPlugin(); RT::PiProgram NativePrg = createPIProgram(Img, Context, Device); @@ -820,14 +835,6 @@ ProgramManager::ProgramPtr ProgramManager::build( } bool LinkDeviceLibs = (DeviceLibReqMask != 0); - const char *CompileOpts = std::getenv("SYCL_PROGRAM_COMPILE_OPTIONS"); - if (!CompileOpts) { - CompileOpts = CompileOptions.c_str(); - } - const char *LinkOpts = std::getenv("SYCL_PROGRAM_LINK_OPTIONS"); - if (!LinkOpts) { - LinkOpts = LinkOptions.c_str(); - } // TODO: Currently, online linking isn't implemented yet on Level Zero. // To enable device libraries and unify the behaviors on all backends, @@ -839,9 +846,8 @@ ProgramManager::ProgramPtr ProgramManager::build( // TODO: this is a temporary workaround for GPU tests for ESIMD compiler. // We do not link with other device libraries, because it may fail // due to unrecognized SPIR-V format of those libraries. - if (std::string(CompileOpts).find(std::string("-cmc")) != std::string::npos || - std::string(CompileOpts).find(std::string("-vc-codegen")) != - std::string::npos) + if (CompileOptions.find(std::string("-cmc")) != std::string::npos || + CompileOptions.find(std::string("-vc-codegen")) != std::string::npos) LinkDeviceLibs = false; std::vector LinkPrograms; @@ -852,11 +858,9 @@ ProgramManager::ProgramPtr ProgramManager::build( const detail::plugin &Plugin = Context->getPlugin(); if (LinkPrograms.empty()) { - std::string Opts(CompileOpts); - RT::PiResult Error = Plugin.call_nocheck( - Program.get(), /*num devices =*/1, &Device, Opts.c_str(), nullptr, - nullptr); + Program.get(), /*num devices =*/1, &Device, CompileOptions.c_str(), + nullptr, nullptr); if (Error != PI_SUCCESS) throw compile_program_error(getProgramBuildLog(Program.get(), Context), Error); @@ -865,13 +869,13 @@ ProgramManager::ProgramPtr ProgramManager::build( // Include the main program and compile/link everything together Plugin.call(Program.get(), /*num devices =*/1, - &Device, CompileOpts, 0, nullptr, - nullptr, nullptr, nullptr); + &Device, CompileOptions.c_str(), 0, + nullptr, nullptr, nullptr, nullptr); LinkPrograms.push_back(Program.get()); RT::PiProgram LinkedProg = nullptr; RT::PiResult Error = Plugin.call_nocheck( - Context->getHandleRef(), /*num devices =*/1, &Device, LinkOpts, + Context->getHandleRef(), /*num devices =*/1, &Device, LinkOptions.c_str(), LinkPrograms.size(), LinkPrograms.data(), nullptr, nullptr, &LinkedProg); // Link program call returns a new program object if all parameters are valid, From 4db44a5eebf6d5df3f2a545787bb22aff436ff1a Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Fri, 19 Mar 2021 14:25:02 +0300 Subject: [PATCH 03/32] Raw implementation --- sycl/doc/EnvironmentVariables.md | 4 +- sycl/include/CL/sycl/detail/os_util.hpp | 3 + sycl/source/detail/config.def | 3 +- sycl/source/detail/os_util.cpp | 21 +- .../program_manager/program_manager.cpp | 240 +++++++++++++++++- .../program_manager/program_manager.hpp | 12 + 6 files changed, 267 insertions(+), 16 deletions(-) diff --git a/sycl/doc/EnvironmentVariables.md b/sycl/doc/EnvironmentVariables.md index 136e4ff66ff9a..89086205d0938 100644 --- a/sycl/doc/EnvironmentVariables.md +++ b/sycl/doc/EnvironmentVariables.md @@ -35,8 +35,8 @@ subject to change. Do not rely on these variables in production code. | SYCL_ENABLE_PCI | Integer | When set to 1, enables obtaining the GPU PCI address when using the Level Zero backend. The default is 0. | | SYCL_HOST_UNIFIED_MEMORY | Integer | Enforce host unified memory support or lack of it for the execution graph builder. If set to 0, it is enforced as not supported by all devices. If set to 1, it is enforced as supported by all devices. | | SYCL_CACHE_DIR | Path | Path to persistent cache root directory. Default values are `%AppData%\Intel\sycl_program_cache` for Windows and `$HOME/intel/sycl_program_cache` on Linux. | -| SYCL_CACHE_ENABLED | ON, OFF | Switches persistent cache switch on or off. Default value is ON. | -| SYCL_CACHE_EVICTION | ON, OFF | Switches cache eviction on or off. Default value is ON. | +| SYCL_CACHE_DISABLE_PERSISTENT | Any(\*) | Switches persistent cache switch off. Default value is ON. | +| SYCL_CACHE_EVICTION_DISABLE | Any(\*) | Switches cache eviction off. Default value is ON. | | SYCL_CACHE_MAX_SIZE | Positive integer | Cache eviction is triggered once total size of cached images exceeds the value in megabytes (default - 8 192 for 8 GB). Set to 0 to disable size-based cache eviction. | | SYCL_CACHE_THRESHOLD | Positive integer | Cache eviction threshold in days (default value is 7 for 1 week). Set to 0 for disabling time-based cache eviction. | | SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE | Positive integer | Minimum size of device code image in kilobytes which is reasonable to cache on disk because disk access operation may take more time than do JIT compilation for it. Default value is 0 to cache all images. | diff --git a/sycl/include/CL/sycl/detail/os_util.hpp b/sycl/include/CL/sycl/detail/os_util.hpp index 271943484f85b..f0fd5cfd7ad43 100644 --- a/sycl/include/CL/sycl/detail/os_util.hpp +++ b/sycl/include/CL/sycl/detail/os_util.hpp @@ -80,6 +80,9 @@ class __SYCL_EXPORT OSUtil { /// Deallocates the memory referenced by \p Ptr. static void alignedFree(void *Ptr); + + /// Returns the path to directory storing on-disk SYCL program cache. + static std::string getCacheRoot(); }; } // namespace detail diff --git a/sycl/source/detail/config.def b/sycl/source/detail/config.def index 725dedf20f099..9d6dd937e2bff 100644 --- a/sycl/source/detail/config.def +++ b/sycl/source/detail/config.def @@ -21,7 +21,8 @@ CONFIG(SYCL_PROGRAM_LINK_OPTIONS, 64, __SYCL_PROGRAM_LINK_OPTIONS) CONFIG(SYCL_PROGRAM_COMPILE_OPTIONS, 64, __SYCL_PROGRAM_COMPILE_OPTIONS) CONFIG(SYCL_HOST_UNIFIED_MEMORY, 1, __SYCL_HOST_UNIFIED_MEMORY) CONFIG(SYCL_CACHE_DIR, 1024, __SYCL_CACHE_DIR) -CONFIG(SYCL_CACHE_ENABLED, 3, __SYCL_CACHE_ENABLED) +CONFIG(SYCL_CACHE_DISABLE_PERSISTENT, 1, __SYCL_CACHE_DISABLE_PERSISTENT) +CONFIG(SYCL_CACHE_EVICTION_DISABLE, 1, __SYCL_CACHE_EVICTION_DISABLE) CONFIG(SYCL_CACHE_MAX_SIZE, 16, __SYCL_CACHE_MAX_SIZE) CONFIG(SYCL_CACHE_THRESHOLD, 16, __SYCL_CACHE_THRESHOLD) CONFIG(SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE, 16, __SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE) diff --git a/sycl/source/detail/os_util.cpp b/sycl/source/detail/os_util.cpp index d49fcb037d786..9685b74d1847e 100644 --- a/sycl/source/detail/os_util.cpp +++ b/sycl/source/detail/os_util.cpp @@ -8,6 +8,7 @@ #include #include +#include #include @@ -121,7 +122,7 @@ std::string OSUtil::getCurrentDSODir() { // // 4) Extract an absolute path to a filename and get a dirname from it. // - uintptr_t CurrentFunc = (uintptr_t) &getCurrentDSODir; + uintptr_t CurrentFunc = (uintptr_t)&getCurrentDSODir; std::ifstream Stream("/proc/self/maps"); Stream >> std::hex; while (!Stream.eof()) { @@ -166,7 +167,7 @@ std::string OSUtil::getCurrentDSODir() { return ""; } -std::string OSUtil::getDirName(const char* Path) { +std::string OSUtil::getDirName(const char *Path) { std::string Tmp(Path); // dirname(3) needs a writable C string: a null-terminator is written where a // path should split. @@ -258,6 +259,22 @@ void OSUtil::alignedFree(void *Ptr) { #endif } +std::string OSUtil::getCacheRoot() { + static const char *PersistenCacheRoot = SYCLConfig::get(); + if (PersistenCacheRoot) + return PersistenCacheRoot; + +#if defined(__SYCL_RT_OS_LINUX) + static const char *RootDir = std::getenv("HOME"); +#else + static const char *RootDir = std::getenv("AppData"); +#endif + std::string Root{RootDir ? RootDir : "."}; + + Root += "/intel/sycl_cache"; + return Root; +} + } // namespace detail } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 6fd38fb71ec8e..fbe24507b7fc9 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -30,10 +30,18 @@ #include #include #include +#include +#include #include #include #include +#include +#include +#include #include +#include +#include +#include __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { @@ -41,7 +49,7 @@ namespace detail { using ContextImplPtr = std::shared_ptr; -static constexpr int DbgProgMgr = 1; +static constexpr int DbgProgMgr = 2; enum BuildState { BS_InProgress, BS_Done, BS_Failed }; @@ -182,6 +190,7 @@ getOrBuild(KernelProgramCache &KPCache, KeyT &&CacheKey, AcquireFT &&Acquire, // only the building thread will run this try { + RetT *Desired = Build(); #ifndef NDEBUG @@ -346,6 +355,204 @@ RT::PiProgram ProgramManager::createPIProgram(const RTDeviceBinaryImage &Img, return Res; } +long GetFileSize(const char *FileName) { + struct stat Stat; + if (!stat(FileName, &Stat)) + return Stat.st_size; + return -1; +} + +inline bool IsFSEntryPresent(const char *Path) { + struct stat Stat; + return !stat(Path, &Stat); +} + +int MakePathRecur(const char *Dir, mode_t Mode) { + assert((Dir != nullptr) && "Passed null-pointer as directory name."); + + // Directory is present - do nothing + if (IsFSEntryPresent(Dir)) + return 0; + + char *CurDir = strdup(Dir); + MakePathRecur(dirname(CurDir), Mode); + if (DbgProgMgr > 1) + std::cerr << "Created directory: " << CurDir << std::endl; + + free(CurDir); + return mkdir(Dir, Mode); +} + +void WriteCacheItem(const std::string &FileName, + const std::vector> &Data) { + std::ofstream FileStream{FileName, std::ios::binary}; + if (DbgProgMgr > 1) { + std::cerr << "####Writing programs built for " << std::dec << Data.size() + << " devices:\n"; + } + + size_t Size = Data.size(); + FileStream.write((char *)&Size, sizeof(Size)); + for (size_t i = 0; i < Data.size(); ++i) { + if (DbgProgMgr > 1) { + std::cerr << "\tWrite " << i << "-th image of size " << std::dec + << Data[i].size() << "\n"; + } + Size = Data[i].size(); + FileStream.write((char *)&Size, sizeof(Size)); + FileStream.write(Data[i].data(), Size); + } + FileStream.close(); +} + +std::vector> ReadCacheItem(const std::string &FileName) { + std::vector> Res; + std::ifstream FileStream{FileName, std::ios::binary}; + size_t ImgNum, ImgSize; + FileStream.read((char *)&ImgNum, sizeof(ImgNum)); + if (DbgProgMgr > 1) { + std::cerr << "####Reading programs built for " << std::dec << ImgNum + << " devices:\n"; + } + + Res.resize(ImgNum); + + for (size_t i = 0; i < ImgNum; ++i) { + FileStream.read((char *)&ImgSize, sizeof(ImgSize)); + if (DbgProgMgr > 1) { + std::cerr << "\tRead " << i << "-th image of size " << std::dec << ImgSize + << "\n"; + } + + Res[i].resize(ImgSize); + FileStream.read(Res[i].data(), ImgSize); + } + + return Res; +} + +std::string getDeviceString(const device &Device) { + return {Device.get_platform().get_info() + + Device.get_info() + + Device.get_info() + + Device.get_info()}; +} + +std::string DumpBinData(const unsigned char *Data, size_t Size) { + if (!Size) + return "NONE"; + std::stringstream ss; + for (size_t i = 0; i < Size; i++) { + ss << std::hex << (int)Data[i]; + } + return ss.str(); +} + +std::string GetCacheItemDirName(const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj SpecConsts, + const std::string &BuildOptionsString) { + static std::string cache_root{detail::OSUtil::getCacheRoot()}; + + std::string ImgString{ + DumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + std::string DeviceString{getDeviceString(Device)}; + std::string SpecConstsString{ + DumpBinData(SpecConsts.data(), SpecConsts.size())}; + std::hash StringHasher{}; + return {cache_root + "/" + std::to_string(StringHasher(DeviceString)) + "/" + + std::to_string(StringHasher(ImgString)) + "/" + + std::to_string(StringHasher(SpecConstsString)) + "/" + + std::to_string(StringHasher(BuildOptionsString))}; +} + +bool IsPersistentCacheEnabled() { + static const char *PersistenCacheDisabled = + SYCLConfig::get(); + + if (DbgProgMgr > 0) + std::cerr << "Persistent cache " + << (PersistenCacheDisabled ? "disabled." : "enabled.") + << std::endl; + return !PersistenCacheDisabled; +} + +void ProgramManager::putPIProgramToDisc(const detail::plugin &Plugin, + const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj SpecConsts, + const std::string &BuildOptionsString, + const RT::PiProgram &Program) { + if (!IsPersistentCacheEnabled()) { + return; + } + + static std::string DirName = + GetCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); + + size_t i = 0; + std::string FileName; + do { + FileName = DirName + "/" + std::to_string(i++) + ".bin"; + } while (IsFSEntryPresent(FileName.c_str())); + + size_t DeviceNum; + Plugin.call(Program, PI_PROGRAM_INFO_NUM_DEVICES, + sizeof(DeviceNum), &DeviceNum, + nullptr); + std::vector BinarySizes(DeviceNum); + Plugin.call( + Program, PI_PROGRAM_INFO_BINARY_SIZES, + sizeof(size_t) * BinarySizes.size(), BinarySizes.data(), nullptr); + + std::vector> Result; + std::vector Pointers; + for (size_t I = 0; I < BinarySizes.size(); ++I) { + Result.emplace_back(BinarySizes[I]); + Pointers.push_back(Result[I].data()); + } + + Plugin.call(Program, PI_PROGRAM_INFO_BINARIES, + sizeof(char *) * Pointers.size(), + Pointers.data(), nullptr); + + MakePathRecur(DirName.c_str(), 0777); + WriteCacheItem(FileName, Result); +} + +bool ProgramManager::getPIProgramFromDisc(ContextImplPtr ContextImpl, + const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj SpecConsts, + const std::string &BuildOptionsString, + RT::PiProgram &NativePrg) { + + if (!IsPersistentCacheEnabled()) + return false; + + std::string Path{ + GetCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString)}; + + if (!IsFSEntryPresent(Path.c_str())) + return false; + + int i = 0; + std::string BinFileName{Path + "/" + std::to_string(i) + ".bin"}; + while (IsFSEntryPresent(BinFileName.c_str())) { + auto BinDataItem = ReadCacheItem(BinFileName); + if (BinDataItem.size()) { + // TODO: Build for multiple devices once supported by program manager + NativePrg = createBinaryProgram( + ContextImpl, Device, (const unsigned char *)BinDataItem[0].data(), + BinDataItem[0].size()); + return true; + } + BinFileName = Path + "/" + std::to_string(++i) + ".bin"; + } + + return false; +} + RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, const context &Context, const device &Device, @@ -390,9 +597,12 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, if (LinkOptsEnv) { LinkOpts = LinkOptsEnv; } + SerializedObj SpecConsts; + if (Prg) + Prg->stableSerializeSpecConstRegistry(SpecConsts); auto BuildF = [this, &M, &KSId, &Context, &Device, Prg, &CompileOpts, - &LinkOpts, &JITCompilationIsRequired] { + &LinkOpts, &JITCompilationIsRequired, SpecConsts] { const RTDeviceBinaryImage &Img = getDeviceImage(M, KSId, Context, Device, JITCompilationIsRequired); // Update only if compile options are not overwritten by environment @@ -413,19 +623,28 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, LinkOpts += Img.getLinkOptions(); ContextImplPtr ContextImpl = getSyclObjImpl(Context); const detail::plugin &Plugin = ContextImpl->getPlugin(); - RT::PiProgram NativePrg = createPIProgram(Img, Context, Device); - if (Prg) - flushSpecConstants(*Prg, NativePrg, &Img); + RT::PiProgram NativePrg; + bool LoadedFromDiskCache = + getPIProgramFromDisc(ContextImpl, Device, Img, SpecConsts, + CompileOpts + LinkOpts, NativePrg); + if (!LoadedFromDiskCache) { + NativePrg = createPIProgram(Img, Context, Device); + if (Prg) + flushSpecConstants(*Prg, NativePrg, &Img); + } + ProgramPtr ProgramManaged( NativePrg, Plugin.getPiPlugin().PiFunctionTable.piProgramRelease); // Link a fallback implementation of device libraries if they are not // supported by a device compiler. - // Pre-compiled programs are supposed to be already linked. + // Pre-compiled programs (after AOT compilation or read from persitent + // cache) are supposed to be already linked. // If device image is not SPIR-V, DeviceLibReqMask will be 0 which means // no fallback device library will be linked. uint32_t DeviceLibReqMask = 0; - if (Img.getFormat() == PI_DEVICE_BINARY_TYPE_SPIRV && + if (!LoadedFromDiskCache && + Img.getFormat() == PI_DEVICE_BINARY_TYPE_SPIRV && !SYCLConfig::get()) DeviceLibReqMask = getDeviceLibReqMask(Img); @@ -438,13 +657,12 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, std::lock_guard Lock(MNativeProgramsMutex); NativePrograms[BuiltProgram.get()] = &Img; } + if (!LoadedFromDiskCache) + putPIProgramToDisc(Plugin, Device, Img, SpecConsts, + CompileOpts + LinkOpts, BuiltProgram.get()); return BuiltProgram.release(); }; - SerializedObj SpecConsts; - if (Prg) - Prg->stableSerializeSpecConstRegistry(SpecConsts); - const RT::PiDevice PiDevice = getRawSyclObjImpl(Device)->getHandleRef(); auto BuildResult = getOrBuild( Cache, diff --git a/sycl/source/detail/program_manager/program_manager.hpp b/sycl/source/detail/program_manager/program_manager.hpp index fa32965ee8878..3fb8abd6e1986 100644 --- a/sycl/source/detail/program_manager/program_manager.hpp +++ b/sycl/source/detail/program_manager/program_manager.hpp @@ -147,6 +147,18 @@ class ProgramManager { bool JITCompilationIsRequired = false); using ProgramPtr = unique_ptr_class, decltype(&::piProgramRelease)>; + bool getPIProgramFromDisc(ContextImplPtr ContextImpl, const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj SpecConsts, + const std::string &BuildOptions, + RT::PiProgram &NativePrg); + + void putPIProgramToDisc(const detail::plugin &Plugin, const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj SpecConsts, + const std::string &BuildOptionsString, + const RT::PiProgram &Program); + ProgramPtr build(ProgramPtr Program, const ContextImplPtr Context, const string_class &CompileOptions, const string_class &LinkOptions, const RT::PiDevice &Device, From 95f4e99c153bd0b3b2881fc4dd090d47e07eb882 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Mon, 22 Mar 2021 15:03:56 +0300 Subject: [PATCH 04/32] Fix LIT tests --- sycl/doc/KernelProgramCache.md | 6 +- .../program_manager/program_manager.cpp | 192 ++++++++++++++---- .../program_manager/program_manager.hpp | 4 +- sycl/test/abi/sycl_symbols_linux.dump | 1 + sycl/test/on-device/cache/basic.cpp | 26 +++ sycl/test/on-device/cache/basic.hpp | 68 +++++++ sycl/test/on-device/cache/spec_consts.cpp | 28 +++ sycl/test/on-device/cache/spec_consts.hpp | 178 ++++++++++++++++ sycl/unittests/kernel-and-program/Cache.cpp | 40 ++++ 9 files changed, 497 insertions(+), 46 deletions(-) create mode 100644 sycl/test/on-device/cache/basic.cpp create mode 100644 sycl/test/on-device/cache/basic.hpp create mode 100644 sycl/test/on-device/cache/spec_consts.cpp create mode 100644 sycl/test/on-device/cache/spec_consts.hpp diff --git a/sycl/doc/KernelProgramCache.md b/sycl/doc/KernelProgramCache.md index 3b634b709c61d..d5313c02acaa3 100644 --- a/sycl/doc/KernelProgramCache.md +++ b/sycl/doc/KernelProgramCache.md @@ -175,9 +175,9 @@ three sources of build options: There is set of configuration parameters which can be set as environment variables or parameters in `sycl.conf` and affect cache behavior: | Environment variable | Values | Description | | -------------------- | ------ | ----------- | -| `SYCL_CACHE_DIR`| Path | Path to persistent cache root directory. Default values are `%AppData%\Intel\sycl_program_cache` for Windows and `$HOME/intel/sycl_program_cache` on Linux. | -| `SYCL_CACHE_ENABLED` | ON, OFF | Switches persistent cache switch on or off. Default value is ON. | -| `SYCL_CACHE_EVICTION` | ON, OFF | Switches cache eviction on or off. Default value is ON. | +| `SYCL_CACHE_DIR` | Path | Path to persistent cache root directory. Default values are `%AppData%\Intel\sycl_program_cache` for Windows and `$HOME/intel/sycl_program_cache` on Linux. | +| `SYCL_CACHE_DISABLE_PERSISTENT` | Any(\*) | Switches persistent cache switch off. Default value is ON. | +| `SYCL_CACHE_EVICTION_DISABLE` | Any(\*) | Switches cache eviction off. Default value is ON. | | `SYCL_CACHE_MAX_SIZE` | Positive integer | Cache eviction is triggered once total size of cached images exceeds the value in megabytes (default - 8 192 for 8 GB). Set to 0 to disable size-based cache eviction. | | `SYCL_CACHE_THRESHOLD` | Positive integer | Cache eviction threshold in days (default value is 7 for 1 week). Set to 0 for disabling time-based cache eviction. | | `SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE` | Positive integer | Minimum size of device code image in kilobytes which is reasonable to cache on disk because disk access operation may take more time than do JIT compilation for it. Default value is 0 to cache all images. | diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index ca591eec54b3f..cb4a0266c7750 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -356,16 +356,26 @@ RT::PiProgram ProgramManager::createPIProgram(const RTDeviceBinaryImage &Img, return Res; } -long GetFileSize(const char *FileName) { - struct stat Stat; - if (!stat(FileName, &Stat)) - return Stat.st_size; - return -1; +std::string getDeviceString(const device &Device) { + return {Device.get_platform().get_info() +"/"+ + Device.get_info() + "/"+ + Device.get_info() + "/"+ + Device.get_info()}; } -inline bool IsFSEntryPresent(const char *Path) { +std::string DumpBinData(const unsigned char *Data, size_t Size) { + if (!Size) + return "NONE"; + std::stringstream ss; + for (size_t i = 0; i < Size; i++) { + ss << std::hex << (int)Data[i]; + } + return ss.str(); +} + +inline bool IsFSEntryPresent(std::string Path) { struct stat Stat; - return !stat(Path, &Stat); + return !stat(Path.c_str(), &Stat); } int MakePathRecur(const char *Dir, mode_t Mode) { @@ -378,14 +388,14 @@ int MakePathRecur(const char *Dir, mode_t Mode) { char *CurDir = strdup(Dir); MakePathRecur(dirname(CurDir), Mode); if (DbgProgMgr > 1) - std::cerr << "Created directory: " << CurDir << std::endl; + std::cerr << "####Created directory: " << CurDir << std::endl; free(CurDir); return mkdir(Dir, Mode); } -void WriteCacheItem(const std::string &FileName, - const std::vector> &Data) { +void WriteCacheItemBin(const std::string &FileName, + const std::vector> &Data) { std::ofstream FileStream{FileName, std::ios::binary}; if (DbgProgMgr > 1) { std::cerr << "####Writing programs built for " << std::dec << Data.size() @@ -396,7 +406,7 @@ void WriteCacheItem(const std::string &FileName, FileStream.write((char *)&Size, sizeof(Size)); for (size_t i = 0; i < Data.size(); ++i) { if (DbgProgMgr > 1) { - std::cerr << "\tWrite " << i << "-th image of size " << std::dec + std::cerr << "####\tWrite " << i << "-th image of size " << std::dec << Data[i].size() << "\n"; } Size = Data[i].size(); @@ -406,6 +416,36 @@ void WriteCacheItem(const std::string &FileName, FileStream.close(); } +void WriteCacheItemSrc(const std::string &FileName, const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString) { + std::ofstream FileStream{FileName, std::ios::binary}; + std::string ImgString{ + DumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + std::string DeviceString{getDeviceString(Device)}; + std::string SpecConstsString{ + DumpBinData(SpecConsts.data(), SpecConsts.size())}; + if (DbgProgMgr > 1) { + std::cerr << "####Writing source for cache item.\n"; + std::cerr << "####'"<> ReadCacheItem(const std::string &FileName) { std::vector> Res; std::ifstream FileStream{FileName, std::ios::binary}; @@ -421,8 +461,8 @@ std::vector> ReadCacheItem(const std::string &FileName) { for (size_t i = 0; i < ImgNum; ++i) { FileStream.read((char *)&ImgSize, sizeof(ImgSize)); if (DbgProgMgr > 1) { - std::cerr << "\tRead " << i << "-th image of size " << std::dec << ImgSize - << "\n"; + std::cerr << "####\tRead " << i << "-th image of size " << std::dec + << ImgSize << "\n"; } Res[i].resize(ImgSize); @@ -432,26 +472,88 @@ std::vector> ReadCacheItem(const std::string &FileName) { return Res; } -std::string getDeviceString(const device &Device) { - return {Device.get_platform().get_info() + - Device.get_info() + - Device.get_info() + - Device.get_info()}; -} +bool IsCacheItemSrcEqual(const std::string &FileName, const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString) { + std::ifstream FileStream{FileName, std::ios::binary}; + std::string ImgString{ + DumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + std::string DeviceString{getDeviceString(Device)}; + std::string SpecConstsString{ + DumpBinData(SpecConsts.data(), SpecConsts.size())}; -std::string DumpBinData(const unsigned char *Data, size_t Size) { - if (!Size) - return "NONE"; - std::stringstream ss; - for (size_t i = 0; i < Size; i++) { - ss << std::hex << (int)Data[i]; + size_t Size; + std::string res; + + FileStream.read((char *)&Size, sizeof(Size)); + res.resize(Size); + FileStream.read(&res[0], Size); + if (DeviceString.compare(res)) { + if (DbgProgMgr > 1) { + std::cerr << "####Devices differ:"< 1) { + std::cerr << "####Build options differ:\n"; + std::cerr << "####'" < 1) { + std::cerr << "####Specialization constants differ\n"; + std::cerr << "####'" < 1) { + std::cerr << "####Images differ\n"; + std::cerr << "####'" < 1) + std::cerr << "####Cache item sources are equal\n"; + return true; } std::string GetCacheItemDirName(const device &Device, const RTDeviceBinaryImage &Img, - const SerializedObj SpecConsts, + const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { static std::string cache_root{detail::OSUtil::getCacheRoot()}; @@ -461,6 +563,7 @@ std::string GetCacheItemDirName(const device &Device, std::string SpecConstsString{ DumpBinData(SpecConsts.data(), SpecConsts.size())}; std::hash StringHasher{}; + return {cache_root + "/" + std::to_string(StringHasher(DeviceString)) + "/" + std::to_string(StringHasher(ImgString)) + "/" + std::to_string(StringHasher(SpecConstsString)) + "/" + @@ -472,7 +575,7 @@ bool IsPersistentCacheEnabled() { SYCLConfig::get(); if (DbgProgMgr > 0) - std::cerr << "Persistent cache " + std::cerr << "####Persistent cache " << (PersistenCacheDisabled ? "disabled." : "enabled.") << std::endl; return !PersistenCacheDisabled; @@ -481,26 +584,28 @@ bool IsPersistentCacheEnabled() { void ProgramManager::putPIProgramToDisc(const detail::plugin &Plugin, const device &Device, const RTDeviceBinaryImage &Img, - const SerializedObj SpecConsts, + const SerializedObj &SpecConsts, const std::string &BuildOptionsString, const RT::PiProgram &Program) { if (!IsPersistentCacheEnabled()) { return; } - static std::string DirName = + std::string DirName = GetCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); size_t i = 0; std::string FileName; do { - FileName = DirName + "/" + std::to_string(i++) + ".bin"; - } while (IsFSEntryPresent(FileName.c_str())); + FileName = DirName + "/" + std::to_string(i++); + } while (IsFSEntryPresent(FileName + ".bin")); + + unsigned int DeviceNum=0; - size_t DeviceNum; Plugin.call(Program, PI_PROGRAM_INFO_NUM_DEVICES, sizeof(DeviceNum), &DeviceNum, - nullptr); + nullptr); + std::vector BinarySizes(DeviceNum); Plugin.call( Program, PI_PROGRAM_INFO_BINARY_SIZES, @@ -518,13 +623,15 @@ void ProgramManager::putPIProgramToDisc(const detail::plugin &Plugin, Pointers.data(), nullptr); MakePathRecur(DirName.c_str(), 0777); - WriteCacheItem(FileName, Result); + WriteCacheItemBin(FileName + ".bin", Result); + WriteCacheItemSrc(FileName + ".src", Device, Img, SpecConsts, + BuildOptionsString); } bool ProgramManager::getPIProgramFromDisc(ContextImplPtr ContextImpl, const device &Device, const RTDeviceBinaryImage &Img, - const SerializedObj SpecConsts, + const SerializedObj &SpecConsts, const std::string &BuildOptionsString, RT::PiProgram &NativePrg) { @@ -538,17 +645,20 @@ bool ProgramManager::getPIProgramFromDisc(ContextImplPtr ContextImpl, return false; int i = 0; - std::string BinFileName{Path + "/" + std::to_string(i) + ".bin"}; - while (IsFSEntryPresent(BinFileName.c_str())) { - auto BinDataItem = ReadCacheItem(BinFileName); - if (BinDataItem.size()) { + std::string FileName{Path + "/" + std::to_string(i)}; + while (IsFSEntryPresent(FileName + ".bin") && + IsFSEntryPresent(FileName + ".src")) { + auto BinDataItem = ReadCacheItem(FileName + ".bin"); + if (BinDataItem.size() && + IsCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, + BuildOptionsString)) { // TODO: Build for multiple devices once supported by program manager NativePrg = createBinaryProgram( ContextImpl, Device, (const unsigned char *)BinDataItem[0].data(), BinDataItem[0].size()); return true; } - BinFileName = Path + "/" + std::to_string(++i) + ".bin"; + FileName = Path + "/" + std::to_string(++i); } return false; diff --git a/sycl/source/detail/program_manager/program_manager.hpp b/sycl/source/detail/program_manager/program_manager.hpp index 4f1ac38441556..e04d0125fdc03 100644 --- a/sycl/source/detail/program_manager/program_manager.hpp +++ b/sycl/source/detail/program_manager/program_manager.hpp @@ -165,13 +165,13 @@ class ProgramManager { decltype(&::piProgramRelease)>; bool getPIProgramFromDisc(ContextImplPtr ContextImpl, const device &Device, const RTDeviceBinaryImage &Img, - const SerializedObj SpecConsts, + const SerializedObj &SpecConsts, const std::string &BuildOptions, RT::PiProgram &NativePrg); void putPIProgramToDisc(const detail::plugin &Plugin, const device &Device, const RTDeviceBinaryImage &Img, - const SerializedObj SpecConsts, + const SerializedObj &SpecConsts, const std::string &BuildOptionsString, const RT::PiProgram &Program); diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index f9540b8637696..65d8b362ce4ac 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3801,6 +3801,7 @@ _ZN2cl4sycl6detail2pi9assertionEbPKc _ZN2cl4sycl6detail6OSUtil10getDirNameB5cxx11EPKc _ZN2cl4sycl6detail6OSUtil11alignedFreeEPv _ZN2cl4sycl6detail6OSUtil12alignedAllocEmm +_ZN2cl4sycl6detail6OSUtil12getCacheRootB5cxx11Ev _ZN2cl4sycl6detail6OSUtil12getOSMemSizeEv _ZN2cl4sycl6detail6OSUtil16getCurrentDSODirB5cxx11Ev _ZN2cl4sycl6detail6OSUtil17getOSModuleHandleEPKv diff --git a/sycl/test/on-device/cache/basic.cpp b/sycl/test/on-device/cache/basic.cpp new file mode 100644 index 0000000000000..7f6974e5c86c5 --- /dev/null +++ b/sycl/test/on-device/cache/basic.cpp @@ -0,0 +1,26 @@ +// No JITing for host devices. +// REQUIRES: opencl || level_zero || cuda +// RUN: rm -rf %T/cache_dir +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// +//==----------- basic.cpp --------------------------------------------------==// +// +// 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 +// +//===----------------------------------------------------------------------===// +// The test checks that caching works properly. +#include "basic.hpp" + +// CHECK-BUILD: piProgramBuild +// CHECK-BUILD: piProgramCreateWithBinary + +// CHECK-CACHE-NOT: piProgramBuild +// CHECK-CACHE: piProgramCreateWithBinary diff --git a/sycl/test/on-device/cache/basic.hpp b/sycl/test/on-device/cache/basic.hpp new file mode 100644 index 0000000000000..50b92c710d2db --- /dev/null +++ b/sycl/test/on-device/cache/basic.hpp @@ -0,0 +1,68 @@ +//==-------------- basic.hpp -----------------------------------------------==// +// +// 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 + +int main(int argc, char **argv) { + // Test program and kernel APIs when building a kernel. + { + cl::sycl::queue q; + int data = 0; + { + cl::sycl::buffer buf(&data, cl::sycl::range<1>(1)); + cl::sycl::program prg1(q.get_context()); + cl::sycl::program prg2(q.get_context()); + cl::sycl::program prg3(q.get_context()); + cl::sycl::program prg4(q.get_context()); + cl::sycl::program prg5(q.get_context()); + + prg1.build_with_kernel_type(); // 1 cache item + prg2.build_with_kernel_type( + "-cl-fast-relaxed-math"); // +1 cache item due to build options + prg3.build_with_kernel_type(); // program binary is + // equal to prg1 + prg4.build_with_kernel_type(); // program binary is + // equal to prg1 + cl::sycl::kernel krn = prg2.get_kernel(); + + q.submit([&](cl::sycl::handler &cgh) { + auto acc = buf.get_access(cgh); + cgh.single_task(krn, [=]() { acc[0] = acc[0] + 1; }); + }); + } + assert(data == 1); + } + + // Test program and kernel APIs when compiling / linking a kernel. + { + cl::sycl::queue q; + int data = 0; + { + cl::sycl::buffer buf(&data, cl::sycl::range<1>(1)); + cl::sycl::program prg6(q.get_context()); + cl::sycl::program prg7(q.get_context()); + cl::sycl::program prg8(q.get_context()); + prg6.compile_with_kernel_type(); + prg6.link(); // The binary is not cached for separate compile/link + prg7.build_with_kernel_type( + "-cl-fast-relaxed-math"); // program binary is equal to prg2 + prg8.build_with_kernel_type( + "-g"); // +1 cache item due to build options + + cl::sycl::kernel krn = prg6.get_kernel(); + + q.submit([&](cl::sycl::handler &cgh) { + auto acc = buf.get_access(cgh); + cgh.single_task(krn, + [=]() { acc[0] = acc[0] + 1; }); + }); + } + assert(data == 1); + } + return 0; +} diff --git a/sycl/test/on-device/cache/spec_consts.cpp b/sycl/test/on-device/cache/spec_consts.cpp new file mode 100644 index 0000000000000..0e2793801e3fb --- /dev/null +++ b/sycl/test/on-device/cache/spec_consts.cpp @@ -0,0 +1,28 @@ +// No JITing for host devices. +// Specialization constant values are not supported on CUDA +// REQUIRES: opencl || level_zero +// RUN: rm -rf %T/cache_dir +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// +//==----------- spec_consts.cpp --------------------------------------------==// +// +// 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 +// +//===----------------------------------------------------------------------===// +// The test checks that caching works properly for SYCL application containing +// specialization constant values. +#include "spec_consts.hpp" + +// CHECK-BUILD: piProgramBuild +// CHECK-BUILD-NOT: piProgramCreateWithBinary + +// CHECK-CACHE-NOT: piProgramBuild +// CHECK-CACHE: piProgramCreateWithBinary diff --git a/sycl/test/on-device/cache/spec_consts.hpp b/sycl/test/on-device/cache/spec_consts.hpp new file mode 100644 index 0000000000000..7f6142fb55f0f --- /dev/null +++ b/sycl/test/on-device/cache/spec_consts.hpp @@ -0,0 +1,178 @@ +//==-------------- spec_const.hpp ------------------------------------------==// +// +// 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 + +#include +#include + +class MyInt32Const; +class MyFloatConst; +class MyConst; + +using namespace sycl; + +class KernelAAAi; +class KernelBBBf; + +int global_val = 10; + +// Fetch a value at runtime. +int get_value() { return global_val; } + +float foo( + const cl::sycl::ONEAPI::experimental::spec_constant + &f32) { + return f32; +} + +struct SCWrapper { + SCWrapper(cl::sycl::program &p) + : SC1(p.set_spec_constant(4)), + SC2(p.set_spec_constant(2)) {} + + cl::sycl::ONEAPI::experimental::spec_constant SC1; + cl::sycl::ONEAPI::experimental::spec_constant SC2; +}; + +// MyKernel is used to test default constructor +using AccT = sycl::accessor; +using ScT = sycl::ONEAPI::experimental::spec_constant; + +struct MyKernel { + MyKernel(AccT &Acc) : Acc(Acc) {} + + void setConst(ScT Sc) { this->Sc = Sc; } + + void operator()() const { Acc[0] = Sc.get(); } + AccT Acc; + ScT Sc; +}; + +int main(int argc, char **argv) { + global_val = argc + 16; + + cl::sycl::queue q(default_selector{}, [](exception_list l) { + for (auto ep : l) { + try { + std::rethrow_exception(ep); + } catch (cl::sycl::exception &e0) { + std::cout << e0.what(); + } catch (std::exception &e1) { + std::cout << e1.what(); + } catch (...) { + std::cout << "*** catch (...)\n"; + } + } + }); + + std::cout << "Running on " << q.get_device().get_info() + << "\n"; + std::cout << "global_val = " << global_val << "\n"; + cl::sycl::program program1(q.get_context()); + cl::sycl::program program2(q.get_context()); + cl::sycl::program program3(q.get_context()); + cl::sycl::program program4(q.get_context()); + + int goldi = (int)get_value(); + // TODO make this floating point once supported by the compiler + float goldf = (float)get_value(); + + cl::sycl::ONEAPI::experimental::spec_constant i32 = + program1.set_spec_constant(goldi); + + cl::sycl::ONEAPI::experimental::spec_constant f32 = + program2.set_spec_constant(goldf); + + cl::sycl::ONEAPI::experimental::spec_constant sc = + program4.set_spec_constant(goldi); + + program1.build_with_kernel_type(); + // Use an option (does not matter which exactly) to test different internal + // SYCL RT execution path + program2.build_with_kernel_type("-cl-fast-relaxed-math"); + + SCWrapper W(program3); + program3.build_with_kernel_type(); + + program4.build_with_kernel_type(); + + int goldw = 6; + + std::vector veci(1); + std::vector vecf(1); + std::vector vecw(1); + std::vector vec(1); + try { + cl::sycl::buffer bufi(veci.data(), veci.size()); + cl::sycl::buffer buff(vecf.data(), vecf.size()); + cl::sycl::buffer bufw(vecw.data(), vecw.size()); + cl::sycl::buffer buf(vec.data(), vec.size()); + + q.submit([&](cl::sycl::handler &cgh) { + auto acci = bufi.get_access(cgh); + cgh.single_task(program1.get_kernel(), + [=]() { acci[0] = i32.get(); }); + }); + q.submit([&](cl::sycl::handler &cgh) { + auto accf = buff.get_access(cgh); + cgh.single_task(program2.get_kernel(), + [=]() { accf[0] = foo(f32); }); + }); + + q.submit([&](cl::sycl::handler &cgh) { + auto accw = bufw.get_access(cgh); + cgh.single_task( + program3.get_kernel(), + [=]() { accw[0] = W.SC1.get() + W.SC2.get(); }); + }); + // Check spec_constant default construction with subsequent initialization + q.submit([&](cl::sycl::handler &cgh) { + auto acc = buf.get_access(cgh); + // Specialization constants specification says: + // cl::sycl::experimental::spec_constant is default constructible, + // although the object is not considered initialized until the result of + // the call to cl::sycl::program::set_spec_constant is assigned to it. + MyKernel Kernel(acc); // default construct inside MyKernel instance + Kernel.setConst(sc); // initialize to sc, returned by set_spec_constant + + cgh.single_task(program4.get_kernel(), Kernel); + }); + + } catch (cl::sycl::exception &e) { + std::cout << "*** Exception caught: " << e.what() << "\n"; + return 1; + } + bool passed = true; + int vali = veci[0]; + + if (vali != goldi) { + std::cout << "*** ERROR: " << vali << " != " << goldi << "(gold)\n"; + passed = false; + } + int valf = vecf[0]; + + if (valf != goldf) { + std::cout << "*** ERROR: " << valf << " != " << goldf << "(gold)\n"; + passed = false; + } + int valw = vecw[0]; + + if (valw != goldw) { + std::cout << "*** ERROR: " << valw << " != " << goldw << "(gold)\n"; + passed = false; + } + int val = vec[0]; + + if (val != goldi) { + std::cout << "*** ERROR: " << val << " != " << goldi << "(gold)\n"; + passed = false; + } + std::cout << (passed ? "passed\n" : "FAILED\n"); + return passed ? 0 : 1; +} diff --git a/sycl/unittests/kernel-and-program/Cache.cpp b/sycl/unittests/kernel-and-program/Cache.cpp index aabdf86d01aa6..016da417585cb 100644 --- a/sycl/unittests/kernel-and-program/Cache.cpp +++ b/sycl/unittests/kernel-and-program/Cache.cpp @@ -44,6 +44,13 @@ static pi_result redefinedProgramCreateWithSource(pi_context context, return PI_SUCCESS; } +static pi_result redefinedProgramCreateWithBinary( + pi_context context, pi_uint32 num_devices, const pi_device *device_list, + const size_t *lengths, const unsigned char **binaries, + pi_int32 *binary_status, pi_program *ret_program) { + return PI_SUCCESS; +} + static pi_result redefinedProgramBuild(pi_program program, pi_uint32 num_devices, const pi_device *device_list, const char *options, @@ -70,6 +77,33 @@ redefinedProgramLink(pi_context context, pi_uint32 num_devices, return PI_SUCCESS; } +static pi_result redefinedProgramGetInfo(pi_program program, + pi_program_info param_name, + size_t param_value_size, + void *param_value, + size_t *param_value_size_ret) { + if (param_name == PI_PROGRAM_INFO_NUM_DEVICES) { + auto value = reinterpret_cast(param_value); + *value = 1; + } + + if (param_name == PI_PROGRAM_INFO_BINARY_SIZES) { + auto value = reinterpret_cast(param_value); + value[0] = 1; + } + + if (param_name == PI_PROGRAM_INFO_BINARIES) { + auto value = reinterpret_cast(param_value); + value[0] = 1; + } + + return PI_SUCCESS; +} + +static pi_result redefinedProgramRetain(pi_program program) { return PI_SUCCESS; } + +static pi_result redefinedProgramRelease(pi_program program) { return PI_SUCCESS; } + static pi_result redefinedKernelCreate(pi_program program, const char *kernel_name, pi_kernel *ret_kernel) { @@ -117,10 +151,16 @@ class KernelAndProgramCacheTest : public ::testing::Test { Mock->redefine( redefinedProgramCreateWithSource); + Mock->redefine( + redefinedProgramCreateWithBinary); Mock->redefine( redefinedProgramCompile); Mock->redefine(redefinedProgramLink); Mock->redefine(redefinedProgramBuild); + Mock->redefine( + redefinedProgramGetInfo); + Mock->redefine(redefinedProgramRetain); + Mock->redefine(redefinedProgramRelease); Mock->redefine(redefinedKernelCreate); Mock->redefine(redefinedKernelRetain); Mock->redefine(redefinedKernelRelease); From 3f1ab43634ba6c6acce2b27bf1abf90f2869ac55 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Tue, 23 Mar 2021 10:37:37 +0300 Subject: [PATCH 05/32] Fix CI issues --- .../program_manager/program_manager.cpp | 41 ++++++++++--------- sycl/unittests/kernel-and-program/Cache.cpp | 11 +++-- 2 files changed, 29 insertions(+), 23 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index cb4a0266c7750..0327e7964a3f4 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -357,9 +357,9 @@ RT::PiProgram ProgramManager::createPIProgram(const RTDeviceBinaryImage &Img, } std::string getDeviceString(const device &Device) { - return {Device.get_platform().get_info() +"/"+ - Device.get_info() + "/"+ - Device.get_info() + "/"+ + return {Device.get_platform().get_info() + "/" + + Device.get_info() + "/" + + Device.get_info() + "/" + Device.get_info()}; } @@ -428,7 +428,7 @@ void WriteCacheItemSrc(const std::string &FileName, const device &Device, DumpBinData(SpecConsts.data(), SpecConsts.size())}; if (DbgProgMgr > 1) { std::cerr << "####Writing source for cache item.\n"; - std::cerr << "####'"< 1) { - std::cerr << "####Devices differ:"< 1) { std::cerr << "####Build options differ:\n"; - std::cerr << "####'" < 1) { std::cerr << "####Specialization constants differ\n"; - std::cerr << "####'" < 1) { std::cerr << "####Images differ\n"; - std::cerr << "####'" <(Program, PI_PROGRAM_INFO_NUM_DEVICES, sizeof(DeviceNum), &DeviceNum, - nullptr); + nullptr); std::vector BinarySizes(DeviceNum); Plugin.call( diff --git a/sycl/unittests/kernel-and-program/Cache.cpp b/sycl/unittests/kernel-and-program/Cache.cpp index 016da417585cb..357d1bf456e16 100644 --- a/sycl/unittests/kernel-and-program/Cache.cpp +++ b/sycl/unittests/kernel-and-program/Cache.cpp @@ -100,9 +100,13 @@ static pi_result redefinedProgramGetInfo(pi_program program, return PI_SUCCESS; } -static pi_result redefinedProgramRetain(pi_program program) { return PI_SUCCESS; } +static pi_result redefinedProgramRetain(pi_program program) { + return PI_SUCCESS; +} -static pi_result redefinedProgramRelease(pi_program program) { return PI_SUCCESS; } +static pi_result redefinedProgramRelease(pi_program program) { + return PI_SUCCESS; +} static pi_result redefinedKernelCreate(pi_program program, const char *kernel_name, @@ -160,7 +164,8 @@ class KernelAndProgramCacheTest : public ::testing::Test { Mock->redefine( redefinedProgramGetInfo); Mock->redefine(redefinedProgramRetain); - Mock->redefine(redefinedProgramRelease); + Mock->redefine( + redefinedProgramRelease); Mock->redefine(redefinedKernelCreate); Mock->redefine(redefinedKernelRetain); Mock->redefine(redefinedKernelRelease); From 4d847f39c309f5a7aaa5b9705923017a65fce1b4 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Tue, 23 Mar 2021 13:09:56 +0300 Subject: [PATCH 06/32] Enable Windows and brush up code --- sycl/include/CL/sycl/detail/os_util.hpp | 10 +++++ sycl/source/detail/os_util.cpp | 32 +++++++++++++++ .../program_manager/program_manager.cpp | 41 +++---------------- sycl/test/abi/sycl_symbols_linux.dump | 1 + 4 files changed, 49 insertions(+), 35 deletions(-) diff --git a/sycl/include/CL/sycl/detail/os_util.hpp b/sycl/include/CL/sycl/detail/os_util.hpp index f0fd5cfd7ad43..e74cafe319443 100644 --- a/sycl/include/CL/sycl/detail/os_util.hpp +++ b/sycl/include/CL/sycl/detail/os_util.hpp @@ -16,6 +16,7 @@ #include #include #include +#include #ifdef _WIN32 #define __SYCL_RT_OS_WINDOWS @@ -57,6 +58,15 @@ class __SYCL_EXPORT OSUtil { /// Returns a directory component of a path. static std::string getDirName(const char *Path); + /// Checks if specified path is present + static inline bool isPathPresent(const std::string &Path) { + struct stat Stat; + return !stat(Path.c_str(), &Stat); + } + + /// Creates directory recursively + static int makeDir(const char *Dir, mode_t Mode); + /// Module handle for the executable module - it is assumed there is always /// single one at most. static constexpr OSModuleHandle ExeModuleHandle = -1; diff --git a/sycl/source/detail/os_util.cpp b/sycl/source/detail/os_util.cpp index 9685b74d1847e..9d7749204ee9d 100644 --- a/sycl/source/detail/os_util.cpp +++ b/sycl/source/detail/os_util.cpp @@ -212,6 +212,19 @@ std::string OSUtil::getCurrentDSODir() { return Path; } +std::string OSUtil::getDirName(const char *Path) { + std::string Tmp(Path); + // Remove trailing directory separators + Tmp.erase(str.find_last_not_of("/\\") + 1, std::string::npos); + + int pos = Tmp.find_last_of("/\\"); + if (pos != std::string::npos) + return Tmp.substr(0, pos); + + // If no directory separator is present return initial path like dirname does + return Tmp; +} + #elif defined(__SYCL_RT_OS_DARWIN) OSModuleHandle OSUtil::getOSModuleHandle(const void *VirtAddr) { Dl_info Res; @@ -275,6 +288,25 @@ std::string OSUtil::getCacheRoot() { return Root; } +int OSUtil::makeDir(const char *Dir, mode_t Mode) { + assert((Dir != nullptr) && "Passed null-pointer as directory name."); + + // Directory is present - do nothing + if (isPathPresent(Dir)) + return 0; + + char *CurDir = strdup(Dir); + makeDir(getDirName(CurDir).c_str(), Mode); + + free(CurDir); + +#if defined(__SYCL_RT_OS_WINDOWS) + return _mkdir(Dir); +#else + return mkdir(Dir, Mode); +#endif +} + } // namespace detail } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 0327e7964a3f4..35d38802d0a78 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -31,18 +31,10 @@ #include #include #include -#include -#include #include #include #include -#include -#include -#include #include -#include -#include -#include __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { @@ -50,7 +42,7 @@ namespace detail { using ContextImplPtr = std::shared_ptr; -static constexpr int DbgProgMgr = 2; +static constexpr int DbgProgMgr = 0; enum BuildState { BS_InProgress, BS_Done, BS_Failed }; @@ -373,27 +365,6 @@ std::string DumpBinData(const unsigned char *Data, size_t Size) { return ss.str(); } -inline bool IsFSEntryPresent(std::string Path) { - struct stat Stat; - return !stat(Path.c_str(), &Stat); -} - -int MakePathRecur(const char *Dir, mode_t Mode) { - assert((Dir != nullptr) && "Passed null-pointer as directory name."); - - // Directory is present - do nothing - if (IsFSEntryPresent(Dir)) - return 0; - - char *CurDir = strdup(Dir); - MakePathRecur(dirname(CurDir), Mode); - if (DbgProgMgr > 1) - std::cerr << "####Created directory: " << CurDir << std::endl; - - free(CurDir); - return mkdir(Dir, Mode); -} - void WriteCacheItemBin(const std::string &FileName, const std::vector> &Data) { std::ofstream FileStream{FileName, std::ios::binary}; @@ -599,7 +570,7 @@ void ProgramManager::putPIProgramToDisc(const detail::plugin &Plugin, std::string FileName; do { FileName = DirName + "/" + std::to_string(i++); - } while (IsFSEntryPresent(FileName + ".bin")); + } while (OSUtil::isPathPresent(FileName + ".bin")); unsigned int DeviceNum = 0; @@ -623,7 +594,7 @@ void ProgramManager::putPIProgramToDisc(const detail::plugin &Plugin, sizeof(char *) * Pointers.size(), Pointers.data(), nullptr); - MakePathRecur(DirName.c_str(), 0777); + OSUtil::makeDir(DirName.c_str(), 0777); WriteCacheItemBin(FileName + ".bin", Result); WriteCacheItemSrc(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString); @@ -642,13 +613,13 @@ bool ProgramManager::getPIProgramFromDisc(ContextImplPtr ContextImpl, std::string Path{ GetCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString)}; - if (!IsFSEntryPresent(Path.c_str())) + if (!OSUtil::isPathPresent(Path)) return false; int i = 0; std::string FileName{Path + "/" + std::to_string(i)}; - while (IsFSEntryPresent(FileName + ".bin") && - IsFSEntryPresent(FileName + ".src")) { + while (OSUtil::isPathPresent(FileName + ".bin") && + OSUtil::isPathPresent(FileName + ".src")) { auto BinDataItem = ReadCacheItem(FileName + ".bin"); if (BinDataItem.size() && IsCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index 65d8b362ce4ac..d2da633fa434b 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3805,6 +3805,7 @@ _ZN2cl4sycl6detail6OSUtil12getCacheRootB5cxx11Ev _ZN2cl4sycl6detail6OSUtil12getOSMemSizeEv _ZN2cl4sycl6detail6OSUtil16getCurrentDSODirB5cxx11Ev _ZN2cl4sycl6detail6OSUtil17getOSModuleHandleEPKv +_ZN2cl4sycl6detail6OSUtil7makeDirEPKcj _ZN2cl4sycl6device11get_devicesENS0_4info11device_typeE _ZN2cl4sycl6deviceC1EP13_cl_device_id _ZN2cl4sycl6deviceC1ERKNS0_15device_selectorE From f7760066a3430c2e0f0ea53b9b8fca505107b625 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Tue, 23 Mar 2021 15:14:32 +0300 Subject: [PATCH 07/32] Fix failure --- sycl/include/CL/sycl/detail/os_util.hpp | 2 +- sycl/source/detail/os_util.cpp | 10 ++++++---- .../detail/program_manager/program_manager.cpp | 2 +- sycl/test/abi/sycl_symbols_linux.dump | 2 +- sycl/test/on-device/cache/basic.cpp | 14 +++++++------- sycl/test/on-device/cache/spec_consts.cpp | 14 +++++++------- 6 files changed, 23 insertions(+), 21 deletions(-) diff --git a/sycl/include/CL/sycl/detail/os_util.hpp b/sycl/include/CL/sycl/detail/os_util.hpp index e74cafe319443..ad965b974de8c 100644 --- a/sycl/include/CL/sycl/detail/os_util.hpp +++ b/sycl/include/CL/sycl/detail/os_util.hpp @@ -65,7 +65,7 @@ class __SYCL_EXPORT OSUtil { } /// Creates directory recursively - static int makeDir(const char *Dir, mode_t Mode); + static int makeDir(const char *Dir); /// Module handle for the executable module - it is assumed there is always /// single one at most. diff --git a/sycl/source/detail/os_util.cpp b/sycl/source/detail/os_util.cpp index 9d7749204ee9d..a4c5972c01ecb 100644 --- a/sycl/source/detail/os_util.cpp +++ b/sycl/source/detail/os_util.cpp @@ -30,6 +30,7 @@ #elif defined(__SYCL_RT_OS_WINDOWS) #include +#include #include #include @@ -215,7 +216,7 @@ std::string OSUtil::getCurrentDSODir() { std::string OSUtil::getDirName(const char *Path) { std::string Tmp(Path); // Remove trailing directory separators - Tmp.erase(str.find_last_not_of("/\\") + 1, std::string::npos); + Tmp.erase(Tmp.find_last_not_of("/\\") + 1, std::string::npos); int pos = Tmp.find_last_of("/\\"); if (pos != std::string::npos) @@ -288,7 +289,7 @@ std::string OSUtil::getCacheRoot() { return Root; } -int OSUtil::makeDir(const char *Dir, mode_t Mode) { +int OSUtil::makeDir(const char *Dir) { assert((Dir != nullptr) && "Passed null-pointer as directory name."); // Directory is present - do nothing @@ -296,17 +297,18 @@ int OSUtil::makeDir(const char *Dir, mode_t Mode) { return 0; char *CurDir = strdup(Dir); - makeDir(getDirName(CurDir).c_str(), Mode); + makeDir(getDirName(CurDir).c_str()); free(CurDir); #if defined(__SYCL_RT_OS_WINDOWS) return _mkdir(Dir); #else - return mkdir(Dir, Mode); + return mkdir(Dir, 0777); #endif } } // namespace detail } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) + diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 35d38802d0a78..f14d99c533477 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -594,7 +594,7 @@ void ProgramManager::putPIProgramToDisc(const detail::plugin &Plugin, sizeof(char *) * Pointers.size(), Pointers.data(), nullptr); - OSUtil::makeDir(DirName.c_str(), 0777); + OSUtil::makeDir(DirName.c_str()); WriteCacheItemBin(FileName + ".bin", Result); WriteCacheItemSrc(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString); diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index d2da633fa434b..50c49b1532f8f 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3805,7 +3805,7 @@ _ZN2cl4sycl6detail6OSUtil12getCacheRootB5cxx11Ev _ZN2cl4sycl6detail6OSUtil12getOSMemSizeEv _ZN2cl4sycl6detail6OSUtil16getCurrentDSODirB5cxx11Ev _ZN2cl4sycl6detail6OSUtil17getOSModuleHandleEPKv -_ZN2cl4sycl6detail6OSUtil7makeDirEPKcj +_ZN2cl4sycl6detail6OSUtil7makeDirEPKc _ZN2cl4sycl6device11get_devicesENS0_4info11device_typeE _ZN2cl4sycl6deviceC1EP13_cl_device_id _ZN2cl4sycl6deviceC1ERKNS0_15device_selectorE diff --git a/sycl/test/on-device/cache/basic.cpp b/sycl/test/on-device/cache/basic.cpp index 7f6974e5c86c5..487379e012b24 100644 --- a/sycl/test/on-device/cache/basic.cpp +++ b/sycl/test/on-device/cache/basic.cpp @@ -1,13 +1,13 @@ // No JITing for host devices. // REQUIRES: opencl || level_zero || cuda -// RUN: rm -rf %T/cache_dir +// RUN: rm -rf %t/cache_dir // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE // //==----------- basic.cpp --------------------------------------------------==// // diff --git a/sycl/test/on-device/cache/spec_consts.cpp b/sycl/test/on-device/cache/spec_consts.cpp index 0e2793801e3fb..9cd7e37423d22 100644 --- a/sycl/test/on-device/cache/spec_consts.cpp +++ b/sycl/test/on-device/cache/spec_consts.cpp @@ -1,14 +1,14 @@ // No JITing for host devices. // Specialization constant values are not supported on CUDA // REQUIRES: opencl || level_zero -// RUN: rm -rf %T/cache_dir +// RUN: rm -rf %t/cache_dir // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD -// RUN: env SYCL_CACHE_DIR=%T/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE // //==----------- spec_consts.cpp --------------------------------------------==// // From 414780c27ab9d404bc91df8e70c1879f1823f3bd Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Tue, 23 Mar 2021 15:36:01 +0300 Subject: [PATCH 08/32] Fix issues --- sycl/source/detail/os_util.cpp | 1 - .../program_manager/program_manager.cpp | 157 ++++++------------ sycl/test/Unit/lit.cfg.py | 1 + 3 files changed, 52 insertions(+), 107 deletions(-) diff --git a/sycl/source/detail/os_util.cpp b/sycl/source/detail/os_util.cpp index a4c5972c01ecb..11b00e19698f0 100644 --- a/sycl/source/detail/os_util.cpp +++ b/sycl/source/detail/os_util.cpp @@ -311,4 +311,3 @@ int OSUtil::makeDir(const char *Dir) { } // namespace detail } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) - diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index f14d99c533477..891388343746f 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -183,7 +183,6 @@ getOrBuild(KernelProgramCache &KPCache, KeyT &&CacheKey, AcquireFT &&Acquire, // only the building thread will run this try { - RetT *Desired = Build(); #ifndef NDEBUG @@ -355,7 +354,7 @@ std::string getDeviceString(const device &Device) { Device.get_info()}; } -std::string DumpBinData(const unsigned char *Data, size_t Size) { +std::string dumpBinData(const unsigned char *Data, size_t Size) { if (!Size) return "NONE"; std::stringstream ss; @@ -365,21 +364,16 @@ std::string DumpBinData(const unsigned char *Data, size_t Size) { return ss.str(); } -void WriteCacheItemBin(const std::string &FileName, +/* Write built binary to persistent cache + * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] + */ +void writeCacheItemBin(const std::string &FileName, const std::vector> &Data) { std::ofstream FileStream{FileName, std::ios::binary}; - if (DbgProgMgr > 1) { - std::cerr << "####Writing programs built for " << std::dec << Data.size() - << " devices:\n"; - } size_t Size = Data.size(); FileStream.write((char *)&Size, sizeof(Size)); for (size_t i = 0; i < Data.size(); ++i) { - if (DbgProgMgr > 1) { - std::cerr << "####\tWrite " << i << "-th image of size " << std::dec - << Data[i].size() << "\n"; - } Size = Data[i].size(); FileStream.write((char *)&Size, sizeof(Size)); FileStream.write(Data[i].data(), Size); @@ -387,20 +381,38 @@ void WriteCacheItemBin(const std::string &FileName, FileStream.close(); } -void WriteCacheItemSrc(const std::string &FileName, const device &Device, +/* Read built binary to persistent cache + * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] + */ +std::vector> readCacheItem(const std::string &FileName) { + std::vector> Res; + std::ifstream FileStream{FileName, std::ios::binary}; + size_t ImgNum, ImgSize; + FileStream.read((char *)&ImgNum, sizeof(ImgNum)); + Res.resize(ImgNum); + for (size_t i = 0; i < ImgNum; ++i) { + FileStream.read((char *)&ImgSize, sizeof(ImgSize)); + Res[i].resize(ImgSize); + FileStream.read(Res[i].data(), ImgSize); + } + + return Res; +} + +/* Writing cache item key sources to be used for reliable identification + * Format: Four pairs of [size, value] for device, build options, specialization + * constant values, device code SPIR-V image. + */ +void writeCacheItemSrc(const std::string &FileName, const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { std::ofstream FileStream{FileName, std::ios::binary}; std::string ImgString{ - DumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; std::string DeviceString{getDeviceString(Device)}; std::string SpecConstsString{ - DumpBinData(SpecConsts.data(), SpecConsts.size())}; - if (DbgProgMgr > 1) { - std::cerr << "####Writing source for cache item.\n"; - std::cerr << "####'" << DeviceString << "'" << std::endl; - } + dumpBinData(SpecConsts.data(), SpecConsts.size())}; size_t Size = DeviceString.size(); FileStream.write((char *)&Size, sizeof(Size)); @@ -417,42 +429,18 @@ void WriteCacheItemSrc(const std::string &FileName, const device &Device, FileStream.close(); } -std::vector> ReadCacheItem(const std::string &FileName) { - std::vector> Res; - std::ifstream FileStream{FileName, std::ios::binary}; - size_t ImgNum, ImgSize; - FileStream.read((char *)&ImgNum, sizeof(ImgNum)); - if (DbgProgMgr > 1) { - std::cerr << "####Reading programs built for " << std::dec << ImgNum - << " devices:\n"; - } - - Res.resize(ImgNum); - - for (size_t i = 0; i < ImgNum; ++i) { - FileStream.read((char *)&ImgSize, sizeof(ImgSize)); - if (DbgProgMgr > 1) { - std::cerr << "####\tRead " << i << "-th image of size " << std::dec - << ImgSize << "\n"; - } - - Res[i].resize(ImgSize); - FileStream.read(Res[i].data(), ImgSize); - } - - return Res; -} - -bool IsCacheItemSrcEqual(const std::string &FileName, const device &Device, +/* Check that cache item key sources are equal to the current program + */ +bool isCacheItemSrcEqual(const std::string &FileName, const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { std::ifstream FileStream{FileName, std::ios::binary}; std::string ImgString{ - DumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; std::string DeviceString{getDeviceString(Device)}; std::string SpecConstsString{ - DumpBinData(SpecConsts.data(), SpecConsts.size())}; + dumpBinData(SpecConsts.data(), SpecConsts.size())}; size_t Size; std::string res; @@ -460,80 +448,42 @@ bool IsCacheItemSrcEqual(const std::string &FileName, const device &Device, FileStream.read((char *)&Size, sizeof(Size)); res.resize(Size); FileStream.read(&res[0], Size); - if (DeviceString.compare(res)) { - if (DbgProgMgr > 1) { - std::cerr << "####Devices differ:" - << DeviceString.compare(0, Size - 1, res.data()) << "\n"; - std::cerr << "####'" << DeviceString << "'\n"; - std::cerr << "####\t vs\n"; - std::cerr << "####'" << std::string(res.data(), Size) << "'\n"; - std::cerr << "####Cached size " << std::dec << Size << " vs current size " - << DeviceString.size() << std::endl; - for (unsigned int i = 0; i < Size; i++) { - if (res[i] != DeviceString[i]) - std::cerr << "####First diff on " << i << std::endl; - } - } - + if (DeviceString.compare(res)) return false; - } FileStream.read((char *)&Size, sizeof(Size)); res.resize(Size); FileStream.read(&res[0], Size); - if (BuildOptionsString.compare(0, Size, res.data())) { - if (DbgProgMgr > 1) { - std::cerr << "####Build options differ:\n"; - std::cerr << "####'" << BuildOptionsString << "'\n"; - std::cerr << "####\t vs\n"; - std::cerr << "####'" << std::string(res.data(), Size) << "'\n"; - } + if (BuildOptionsString.compare(0, Size, res.data())) return false; - } FileStream.read((char *)&Size, sizeof(Size)); res.resize(Size); FileStream.read(&res[0], Size); - if (SpecConstsString.compare(0, Size, res.data())) { - if (DbgProgMgr > 1) { - std::cerr << "####Specialization constants differ\n"; - std::cerr << "####'" << SpecConstsString << "'\n"; - std::cerr << "####\t vs\n"; - std::cerr << "####'" << std::string(res.data(), Size) << "'\n"; - } + if (SpecConstsString.compare(0, Size, res.data())) return false; - } FileStream.read((char *)&Size, sizeof(Size)); res.resize(Size); FileStream.read(&res[0], Size); - if (ImgString.compare(0, Size, res.data())) { - if (DbgProgMgr > 1) { - std::cerr << "####Images differ\n"; - std::cerr << "####'" << ImgString << "'\n"; - std::cerr << "####\t vs\n"; - std::cerr << "####'" << std::string(res.data(), Size) << "'\n"; - } + if (ImgString.compare(0, Size, res.data())) return false; - } FileStream.close(); - if (DbgProgMgr > 1) - std::cerr << "####Cache item sources are equal\n"; return true; } -std::string GetCacheItemDirName(const device &Device, +std::string getCacheItemDirName(const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { static std::string cache_root{detail::OSUtil::getCacheRoot()}; std::string ImgString{ - DumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; std::string DeviceString{getDeviceString(Device)}; std::string SpecConstsString{ - DumpBinData(SpecConsts.data(), SpecConsts.size())}; + dumpBinData(SpecConsts.data(), SpecConsts.size())}; std::hash StringHasher{}; return {cache_root + "/" + std::to_string(StringHasher(DeviceString)) + "/" + @@ -542,14 +492,9 @@ std::string GetCacheItemDirName(const device &Device, std::to_string(StringHasher(BuildOptionsString))}; } -bool IsPersistentCacheEnabled() { +static bool isPersistentCacheEnabled() { static const char *PersistenCacheDisabled = SYCLConfig::get(); - - if (DbgProgMgr > 0) - std::cerr << "####Persistent cache " - << (PersistenCacheDisabled ? "disabled." : "enabled.") - << std::endl; return !PersistenCacheDisabled; } @@ -559,12 +504,12 @@ void ProgramManager::putPIProgramToDisc(const detail::plugin &Plugin, const SerializedObj &SpecConsts, const std::string &BuildOptionsString, const RT::PiProgram &Program) { - if (!IsPersistentCacheEnabled()) { + if (!isPersistentCacheEnabled()) { return; } std::string DirName = - GetCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); + getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); size_t i = 0; std::string FileName; @@ -595,8 +540,8 @@ void ProgramManager::putPIProgramToDisc(const detail::plugin &Plugin, Pointers.data(), nullptr); OSUtil::makeDir(DirName.c_str()); - WriteCacheItemBin(FileName + ".bin", Result); - WriteCacheItemSrc(FileName + ".src", Device, Img, SpecConsts, + writeCacheItemBin(FileName + ".bin", Result); + writeCacheItemSrc(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString); } @@ -607,11 +552,11 @@ bool ProgramManager::getPIProgramFromDisc(ContextImplPtr ContextImpl, const std::string &BuildOptionsString, RT::PiProgram &NativePrg) { - if (!IsPersistentCacheEnabled()) + if (!isPersistentCacheEnabled()) return false; std::string Path{ - GetCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString)}; + getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString)}; if (!OSUtil::isPathPresent(Path)) return false; @@ -620,9 +565,9 @@ bool ProgramManager::getPIProgramFromDisc(ContextImplPtr ContextImpl, std::string FileName{Path + "/" + std::to_string(i)}; while (OSUtil::isPathPresent(FileName + ".bin") && OSUtil::isPathPresent(FileName + ".src")) { - auto BinDataItem = ReadCacheItem(FileName + ".bin"); + auto BinDataItem = readCacheItem(FileName + ".bin"); if (BinDataItem.size() && - IsCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, + isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString)) { // TODO: Build for multiple devices once supported by program manager NativePrg = createBinaryProgram( diff --git a/sycl/test/Unit/lit.cfg.py b/sycl/test/Unit/lit.cfg.py index 2dbc0d14e9f21..d261014cc20cb 100644 --- a/sycl/test/Unit/lit.cfg.py +++ b/sycl/test/Unit/lit.cfg.py @@ -74,4 +74,5 @@ def find_shlibpath_var(): .format(platform.system())) config.environment['SYCL_DEVICE_FILTER'] = lit_config.params.get('SYCL_PLUGIN', "opencl") +config.environment['SYCL_CACHE_DISABLE_PERSISTENT'] = '1' lit_config.note("Backend: {}".format(config.environment['SYCL_DEVICE_FILTER'])) From d3283f32dc9aba88bd9e54b46ab576b7c7d5d4ad Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Wed, 24 Mar 2021 22:36:10 +0300 Subject: [PATCH 09/32] Move on-disk cache to separate module --- sycl/source/CMakeLists.txt | 1 + sycl/source/detail/persistent_cache.cpp | 246 +++++++++++++++++ sycl/source/detail/persistent_cache.hpp | 77 ++++++ .../program_manager/program_manager.cpp | 257 +----------------- .../program_manager/program_manager.hpp | 11 - 5 files changed, 338 insertions(+), 254 deletions(-) create mode 100644 sycl/source/detail/persistent_cache.cpp create mode 100644 sycl/source/detail/persistent_cache.hpp diff --git a/sycl/source/CMakeLists.txt b/sycl/source/CMakeLists.txt index e7586f38bde2b..7970cb3d55399 100644 --- a/sycl/source/CMakeLists.txt +++ b/sycl/source/CMakeLists.txt @@ -117,6 +117,7 @@ set(SYCL_SOURCES "detail/device_binary_image.cpp" "detail/device_filter.cpp" "detail/device_impl.cpp" + "detail/persistent_cache.cpp" "detail/error_handling/enqueue_kernel.cpp" "detail/event_impl.cpp" "detail/filter_selector_impl.cpp" diff --git a/sycl/source/detail/persistent_cache.cpp b/sycl/source/detail/persistent_cache.cpp new file mode 100644 index 0000000000000..43d244eced68f --- /dev/null +++ b/sycl/source/detail/persistent_cache.cpp @@ -0,0 +1,246 @@ +//==---------- persistent_cache.cpp - On-disk cache for program -*- C++-*---==// +// +// 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 +#include +#include + +__SYCL_INLINE_NAMESPACE(cl) { +namespace sycl { +namespace detail { +void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, + const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString, + const RT::PiProgram &Program) { + if (!isPersistentCacheEnabled()) { + return; + } + + std::string DirName = + getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); + + size_t i = 0; + std::string FileName; + do { + FileName = DirName + "/" + std::to_string(i++); + } while (OSUtil::isPathPresent(FileName + ".bin")); + + unsigned int DeviceNum = 0; + + Plugin.call(Program, PI_PROGRAM_INFO_NUM_DEVICES, + sizeof(DeviceNum), &DeviceNum, + nullptr); + + std::vector BinarySizes(DeviceNum); + Plugin.call( + Program, PI_PROGRAM_INFO_BINARY_SIZES, + sizeof(size_t) * BinarySizes.size(), BinarySizes.data(), nullptr); + + std::vector> Result; + std::vector Pointers; + for (size_t I = 0; I < BinarySizes.size(); ++I) { + Result.emplace_back(BinarySizes[I]); + Pointers.push_back(Result[I].data()); + } + + Plugin.call(Program, PI_PROGRAM_INFO_BINARIES, + sizeof(char *) * Pointers.size(), + Pointers.data(), nullptr); + + OSUtil::makeDir(DirName.c_str()); + writeCacheItemBin(FileName + ".bin", Result); + writeCacheItemSrc(FileName + ".src", Device, Img, SpecConsts, + BuildOptionsString); +} + +std::vector> PersistentCache::getPIProgramFromDisc( + const device &Device, const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, const std::string &BuildOptionsString, + RT::PiProgram &NativePrg) { + + if (!isPersistentCacheEnabled()) + return {}; + + std::string Path{ + getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString)}; + + if (!OSUtil::isPathPresent(Path)) + return {}; + + int i = 0; + std::string FileName{Path + "/" + std::to_string(i)}; + while (OSUtil::isPathPresent(FileName + ".bin") && + OSUtil::isPathPresent(FileName + ".src")) { + if (isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, + BuildOptionsString)) { + return readCacheItem(FileName + ".bin"); + } + FileName = Path + "/" + std::to_string(++i); + } + + return {}; +} + +std::string PersistentCache::getDeviceString(const device &Device) { + return {Device.get_platform().get_info() + "/" + + Device.get_info() + "/" + + Device.get_info() + "/" + + Device.get_info()}; +} + +std::string PersistentCache::dumpBinData(const unsigned char *Data, + size_t Size) { + if (!Size) + return "NONE"; + std::stringstream ss; + for (size_t i = 0; i < Size; i++) { + ss << std::hex << (int)Data[i]; + } + return ss.str(); +} + +/* Write built binary to persistent cache + * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] + */ +void PersistentCache::writeCacheItemBin( + const std::string &FileName, const std::vector> &Data) { + std::ofstream FileStream{FileName, std::ios::binary}; + + size_t Size = Data.size(); + FileStream.write((char *)&Size, sizeof(Size)); + for (size_t i = 0; i < Data.size(); ++i) { + Size = Data[i].size(); + FileStream.write((char *)&Size, sizeof(Size)); + FileStream.write(Data[i].data(), Size); + } + FileStream.close(); +} + +/* Read built binary to persistent cache + * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] + */ +std::vector> +PersistentCache::readCacheItem(const std::string &FileName) { + std::vector> Res; + std::ifstream FileStream{FileName, std::ios::binary}; + size_t ImgNum, ImgSize; + FileStream.read((char *)&ImgNum, sizeof(ImgNum)); + Res.resize(ImgNum); + for (size_t i = 0; i < ImgNum; ++i) { + FileStream.read((char *)&ImgSize, sizeof(ImgSize)); + Res[i].resize(ImgSize); + FileStream.read(Res[i].data(), ImgSize); + } + + return Res; +} + +/* Writing cache item key sources to be used for reliable identification + * Format: Four pairs of [size, value] for device, build options, specialization + * constant values, device code SPIR-V image. + */ +void PersistentCache::writeCacheItemSrc(const std::string &FileName, + const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString) { + std::ofstream FileStream{FileName, std::ios::binary}; + std::string ImgString{ + dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + std::string DeviceString{getDeviceString(Device)}; + std::string SpecConstsString{ + dumpBinData(SpecConsts.data(), SpecConsts.size())}; + + size_t Size = DeviceString.size(); + FileStream.write((char *)&Size, sizeof(Size)); + FileStream.write(DeviceString.data(), Size); + Size = BuildOptionsString.size(); + FileStream.write((char *)&Size, sizeof(Size)); + FileStream.write(BuildOptionsString.data(), Size); + Size = SpecConstsString.size(); + FileStream.write((char *)&Size, sizeof(Size)); + FileStream.write(SpecConstsString.data(), Size); + Size = ImgString.size(); + FileStream.write((char *)&Size, sizeof(Size)); + FileStream.write(ImgString.data(), Size); + FileStream.close(); +} + +/* Check that cache item key sources are equal to the current program + */ +bool PersistentCache::isCacheItemSrcEqual( + const std::string &FileName, const device &Device, + const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, + const std::string &BuildOptionsString) { + std::ifstream FileStream{FileName, std::ios::binary}; + std::string ImgString{ + dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + std::string DeviceString{getDeviceString(Device)}; + std::string SpecConstsString{ + dumpBinData(SpecConsts.data(), SpecConsts.size())}; + + size_t Size; + std::string res; + + FileStream.read((char *)&Size, sizeof(Size)); + res.resize(Size); + FileStream.read(&res[0], Size); + if (DeviceString.compare(res)) + return false; + + FileStream.read((char *)&Size, sizeof(Size)); + res.resize(Size); + FileStream.read(&res[0], Size); + if (BuildOptionsString.compare(0, Size, res.data())) + return false; + + FileStream.read((char *)&Size, sizeof(Size)); + res.resize(Size); + FileStream.read(&res[0], Size); + if (SpecConstsString.compare(0, Size, res.data())) + return false; + + FileStream.read((char *)&Size, sizeof(Size)); + res.resize(Size); + FileStream.read(&res[0], Size); + if (ImgString.compare(0, Size, res.data())) + return false; + + FileStream.close(); + return true; +} + +std::string PersistentCache::getCacheItemDirName( + const device &Device, const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { + static std::string cache_root{detail::OSUtil::getCacheRoot()}; + + std::string ImgString{ + dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + std::string DeviceString{getDeviceString(Device)}; + std::string SpecConstsString{ + dumpBinData(SpecConsts.data(), SpecConsts.size())}; + std::hash StringHasher{}; + + return {cache_root + "/" + std::to_string(StringHasher(DeviceString)) + "/" + + std::to_string(StringHasher(ImgString)) + "/" + + std::to_string(StringHasher(SpecConstsString)) + "/" + + std::to_string(StringHasher(BuildOptionsString))}; +} + +bool PersistentCache::isPersistentCacheEnabled() { + static const char *PersistenCacheDisabled = + SYCLConfig::get(); + return !PersistenCacheDisabled; +} + +} // namespace detail +} // namespace sycl +} // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/detail/persistent_cache.hpp b/sycl/source/detail/persistent_cache.hpp new file mode 100644 index 0000000000000..1a945ae475a01 --- /dev/null +++ b/sycl/source/detail/persistent_cache.hpp @@ -0,0 +1,77 @@ +//==---------- persistent_cache.hpp - On-disk cache for program -*- C++-*---==// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#pragma once + +#include +#include +#include +#include +#include +#include + +__SYCL_INLINE_NAMESPACE(cl) { +namespace sycl { +namespace detail { +class PersistentCache { + /* Form string representing device version */ + static std::string getDeviceString(const device &Device); + /* Form string containing hex representation of the C-string*/ + static std::string dumpBinData(const unsigned char *Data, size_t Size); + /* Write built binary to persistent cache + * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] + */ + static void writeCacheItemBin(const std::string &FileName, + const std::vector> &Data); + /* Read built binary to persistent cache + * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] + */ + static std::vector> + readCacheItem(const std::string &FileName); + /* Writing cache item key sources to be used for reliable identification + * Format: Four pairs of [size, value] for device, build options, + * specialization constant values, device code SPIR-V image. + */ + static void writeCacheItemSrc(const std::string &FileName, + const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString); + /* Check that cache item key sources are equal to the current program + */ + static bool isCacheItemSrcEqual(const std::string &FileName, + const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString); + /* Get directory name for storing current cache item + */ + static std::string getCacheItemDirName(const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString); + /* Check if on-disk cache enabled. + */ + static bool isPersistentCacheEnabled(); + +public: + static std::vector> + getPIProgramFromDisc(const device &Device, const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString, + RT::PiProgram &NativePrg); + static void putPIProgramToDisc(const detail::plugin &Plugin, + const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString, + const RT::PiProgram &Program); +}; +} // namespace detail +} // namespace sycl +} // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 891388343746f..7090f34ba673b 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -21,6 +21,7 @@ #include #include #include +#include #include #include #include @@ -347,240 +348,6 @@ RT::PiProgram ProgramManager::createPIProgram(const RTDeviceBinaryImage &Img, return Res; } -std::string getDeviceString(const device &Device) { - return {Device.get_platform().get_info() + "/" + - Device.get_info() + "/" + - Device.get_info() + "/" + - Device.get_info()}; -} - -std::string dumpBinData(const unsigned char *Data, size_t Size) { - if (!Size) - return "NONE"; - std::stringstream ss; - for (size_t i = 0; i < Size; i++) { - ss << std::hex << (int)Data[i]; - } - return ss.str(); -} - -/* Write built binary to persistent cache - * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] - */ -void writeCacheItemBin(const std::string &FileName, - const std::vector> &Data) { - std::ofstream FileStream{FileName, std::ios::binary}; - - size_t Size = Data.size(); - FileStream.write((char *)&Size, sizeof(Size)); - for (size_t i = 0; i < Data.size(); ++i) { - Size = Data[i].size(); - FileStream.write((char *)&Size, sizeof(Size)); - FileStream.write(Data[i].data(), Size); - } - FileStream.close(); -} - -/* Read built binary to persistent cache - * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] - */ -std::vector> readCacheItem(const std::string &FileName) { - std::vector> Res; - std::ifstream FileStream{FileName, std::ios::binary}; - size_t ImgNum, ImgSize; - FileStream.read((char *)&ImgNum, sizeof(ImgNum)); - Res.resize(ImgNum); - for (size_t i = 0; i < ImgNum; ++i) { - FileStream.read((char *)&ImgSize, sizeof(ImgSize)); - Res[i].resize(ImgSize); - FileStream.read(Res[i].data(), ImgSize); - } - - return Res; -} - -/* Writing cache item key sources to be used for reliable identification - * Format: Four pairs of [size, value] for device, build options, specialization - * constant values, device code SPIR-V image. - */ -void writeCacheItemSrc(const std::string &FileName, const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString) { - std::ofstream FileStream{FileName, std::ios::binary}; - std::string ImgString{ - dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; - std::string DeviceString{getDeviceString(Device)}; - std::string SpecConstsString{ - dumpBinData(SpecConsts.data(), SpecConsts.size())}; - - size_t Size = DeviceString.size(); - FileStream.write((char *)&Size, sizeof(Size)); - FileStream.write(DeviceString.data(), Size); - Size = BuildOptionsString.size(); - FileStream.write((char *)&Size, sizeof(Size)); - FileStream.write(BuildOptionsString.data(), Size); - Size = SpecConstsString.size(); - FileStream.write((char *)&Size, sizeof(Size)); - FileStream.write(SpecConstsString.data(), Size); - Size = ImgString.size(); - FileStream.write((char *)&Size, sizeof(Size)); - FileStream.write(ImgString.data(), Size); - FileStream.close(); -} - -/* Check that cache item key sources are equal to the current program - */ -bool isCacheItemSrcEqual(const std::string &FileName, const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString) { - std::ifstream FileStream{FileName, std::ios::binary}; - std::string ImgString{ - dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; - std::string DeviceString{getDeviceString(Device)}; - std::string SpecConstsString{ - dumpBinData(SpecConsts.data(), SpecConsts.size())}; - - size_t Size; - std::string res; - - FileStream.read((char *)&Size, sizeof(Size)); - res.resize(Size); - FileStream.read(&res[0], Size); - if (DeviceString.compare(res)) - return false; - - FileStream.read((char *)&Size, sizeof(Size)); - res.resize(Size); - FileStream.read(&res[0], Size); - if (BuildOptionsString.compare(0, Size, res.data())) - return false; - - FileStream.read((char *)&Size, sizeof(Size)); - res.resize(Size); - FileStream.read(&res[0], Size); - if (SpecConstsString.compare(0, Size, res.data())) - return false; - - FileStream.read((char *)&Size, sizeof(Size)); - res.resize(Size); - FileStream.read(&res[0], Size); - if (ImgString.compare(0, Size, res.data())) - return false; - - FileStream.close(); - return true; -} - -std::string getCacheItemDirName(const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString) { - static std::string cache_root{detail::OSUtil::getCacheRoot()}; - - std::string ImgString{ - dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; - std::string DeviceString{getDeviceString(Device)}; - std::string SpecConstsString{ - dumpBinData(SpecConsts.data(), SpecConsts.size())}; - std::hash StringHasher{}; - - return {cache_root + "/" + std::to_string(StringHasher(DeviceString)) + "/" + - std::to_string(StringHasher(ImgString)) + "/" + - std::to_string(StringHasher(SpecConstsString)) + "/" + - std::to_string(StringHasher(BuildOptionsString))}; -} - -static bool isPersistentCacheEnabled() { - static const char *PersistenCacheDisabled = - SYCLConfig::get(); - return !PersistenCacheDisabled; -} - -void ProgramManager::putPIProgramToDisc(const detail::plugin &Plugin, - const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString, - const RT::PiProgram &Program) { - if (!isPersistentCacheEnabled()) { - return; - } - - std::string DirName = - getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); - - size_t i = 0; - std::string FileName; - do { - FileName = DirName + "/" + std::to_string(i++); - } while (OSUtil::isPathPresent(FileName + ".bin")); - - unsigned int DeviceNum = 0; - - Plugin.call(Program, PI_PROGRAM_INFO_NUM_DEVICES, - sizeof(DeviceNum), &DeviceNum, - nullptr); - - std::vector BinarySizes(DeviceNum); - Plugin.call( - Program, PI_PROGRAM_INFO_BINARY_SIZES, - sizeof(size_t) * BinarySizes.size(), BinarySizes.data(), nullptr); - - std::vector> Result; - std::vector Pointers; - for (size_t I = 0; I < BinarySizes.size(); ++I) { - Result.emplace_back(BinarySizes[I]); - Pointers.push_back(Result[I].data()); - } - - Plugin.call(Program, PI_PROGRAM_INFO_BINARIES, - sizeof(char *) * Pointers.size(), - Pointers.data(), nullptr); - - OSUtil::makeDir(DirName.c_str()); - writeCacheItemBin(FileName + ".bin", Result); - writeCacheItemSrc(FileName + ".src", Device, Img, SpecConsts, - BuildOptionsString); -} - -bool ProgramManager::getPIProgramFromDisc(ContextImplPtr ContextImpl, - const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString, - RT::PiProgram &NativePrg) { - - if (!isPersistentCacheEnabled()) - return false; - - std::string Path{ - getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString)}; - - if (!OSUtil::isPathPresent(Path)) - return false; - - int i = 0; - std::string FileName{Path + "/" + std::to_string(i)}; - while (OSUtil::isPathPresent(FileName + ".bin") && - OSUtil::isPathPresent(FileName + ".src")) { - auto BinDataItem = readCacheItem(FileName + ".bin"); - if (BinDataItem.size() && - isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, - BuildOptionsString)) { - // TODO: Build for multiple devices once supported by program manager - NativePrg = createBinaryProgram( - ContextImpl, Device, (const unsigned char *)BinDataItem[0].data(), - BinDataItem[0].size()); - return true; - } - FileName = Path + "/" + std::to_string(++i); - } - - return false; -} - RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, const context &Context, const device &Device, @@ -654,10 +421,14 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, ContextImplPtr ContextImpl = getSyclObjImpl(Context); const detail::plugin &Plugin = ContextImpl->getPlugin(); RT::PiProgram NativePrg; - bool LoadedFromDiskCache = - getPIProgramFromDisc(ContextImpl, Device, Img, SpecConsts, - CompileOpts + LinkOpts, NativePrg); - if (!LoadedFromDiskCache) { + auto BinProg = PersistentCache::getPIProgramFromDisc( + Device, Img, SpecConsts, CompileOpts + LinkOpts, NativePrg); + if (BinProg.size()) { + // TODO: Build for multiple devices once supported by program manager + NativePrg = createBinaryProgram(ContextImpl, Device, + (const unsigned char *)BinProg[0].data(), + BinProg[0].size()); + } else { NativePrg = createPIProgram(Img, Context, Device); if (Prg) flushSpecConstants(*Prg, NativePrg, &Img); @@ -673,8 +444,7 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, // If device image is not SPIR-V, DeviceLibReqMask will be 0 which means // no fallback device library will be linked. uint32_t DeviceLibReqMask = 0; - if (!LoadedFromDiskCache && - Img.getFormat() == PI_DEVICE_BINARY_TYPE_SPIRV && + if (!BinProg.size() && Img.getFormat() == PI_DEVICE_BINARY_TYPE_SPIRV && !SYCLConfig::get()) DeviceLibReqMask = getDeviceLibReqMask(Img); @@ -687,9 +457,10 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, std::lock_guard Lock(MNativeProgramsMutex); NativePrograms[BuiltProgram.get()] = &Img; } - if (!LoadedFromDiskCache) - putPIProgramToDisc(Plugin, Device, Img, SpecConsts, - CompileOpts + LinkOpts, BuiltProgram.get()); + if (!BinProg.size()) + PersistentCache::putPIProgramToDisc(Plugin, Device, Img, SpecConsts, + CompileOpts + LinkOpts, + BuiltProgram.get()); return BuiltProgram.release(); }; diff --git a/sycl/source/detail/program_manager/program_manager.hpp b/sycl/source/detail/program_manager/program_manager.hpp index e04d0125fdc03..49be7f29cf9b4 100644 --- a/sycl/source/detail/program_manager/program_manager.hpp +++ b/sycl/source/detail/program_manager/program_manager.hpp @@ -163,17 +163,6 @@ class ProgramManager { bool JITCompilationIsRequired = false); using ProgramPtr = unique_ptr_class, decltype(&::piProgramRelease)>; - bool getPIProgramFromDisc(ContextImplPtr ContextImpl, const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptions, - RT::PiProgram &NativePrg); - - void putPIProgramToDisc(const detail::plugin &Plugin, const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString, - const RT::PiProgram &Program); ProgramPtr build(ProgramPtr Program, const ContextImplPtr Context, const string_class &CompileOptions, From 5f94856896ea9363bd1af82d598d818900221795 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Wed, 24 Mar 2021 22:38:48 +0300 Subject: [PATCH 10/32] revert unnecessary change --- sycl/source/detail/program_manager/program_manager.hpp | 1 - 1 file changed, 1 deletion(-) diff --git a/sycl/source/detail/program_manager/program_manager.hpp b/sycl/source/detail/program_manager/program_manager.hpp index 49be7f29cf9b4..77572330c2fda 100644 --- a/sycl/source/detail/program_manager/program_manager.hpp +++ b/sycl/source/detail/program_manager/program_manager.hpp @@ -163,7 +163,6 @@ class ProgramManager { bool JITCompilationIsRequired = false); using ProgramPtr = unique_ptr_class, decltype(&::piProgramRelease)>; - ProgramPtr build(ProgramPtr Program, const ContextImplPtr Context, const string_class &CompileOptions, const string_class &LinkOptions, const RT::PiDevice &Device, From eece33eaf5d1f6344e1ef078a8bf00a953ae952f Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Thu, 25 Mar 2021 13:25:39 +0300 Subject: [PATCH 11/32] Apply review comments and fix CI issues --- sycl/doc/EnvironmentVariables.md | 2 +- sycl/doc/KernelProgramCache.md | 13 +++---------- sycl/source/detail/os_util.cpp | 20 ++++++++++++++------ sycl/source/detail/persistent_cache.cpp | 14 ++++++++++++++ 4 files changed, 32 insertions(+), 17 deletions(-) diff --git a/sycl/doc/EnvironmentVariables.md b/sycl/doc/EnvironmentVariables.md index 89086205d0938..daf935372f274 100644 --- a/sycl/doc/EnvironmentVariables.md +++ b/sycl/doc/EnvironmentVariables.md @@ -34,7 +34,7 @@ subject to change. Do not rely on these variables in production code. | SYCL_DISABLE_PARALLEL_FOR_RANGE_ROUNDING | Any(\*) | Disables automatic rounding-up of parallel_for invocation ranges. | | SYCL_ENABLE_PCI | Integer | When set to 1, enables obtaining the GPU PCI address when using the Level Zero backend. The default is 0. | | SYCL_HOST_UNIFIED_MEMORY | Integer | Enforce host unified memory support or lack of it for the execution graph builder. If set to 0, it is enforced as not supported by all devices. If set to 1, it is enforced as supported by all devices. | -| SYCL_CACHE_DIR | Path | Path to persistent cache root directory. Default values are `%AppData%\Intel\sycl_program_cache` for Windows and `$HOME/intel/sycl_program_cache` on Linux. | +| SYCL_CACHE_DIR | Path | Path to persistent cache root directory. Default values are `%AppData%\libsycl_cache` for Windows and `$XDG_CACHE_HOME/libsycl_cache` on Linux, if XDG_CACHE_HOME is not set then `$HOME/.cache/libsycl_cache`. | | SYCL_CACHE_DISABLE_PERSISTENT | Any(\*) | Switches persistent cache switch off. Default value is ON. | | SYCL_CACHE_EVICTION_DISABLE | Any(\*) | Switches cache eviction off. Default value is ON. | | SYCL_CACHE_MAX_SIZE | Positive integer | Cache eviction is triggered once total size of cached images exceeds the value in megabytes (default - 8 192 for 8 GB). Set to 0 to disable size-based cache eviction. | diff --git a/sycl/doc/KernelProgramCache.md b/sycl/doc/KernelProgramCache.md index d5313c02acaa3..95e7ea64dd1a0 100644 --- a/sycl/doc/KernelProgramCache.md +++ b/sycl/doc/KernelProgramCache.md @@ -172,16 +172,9 @@ three sources of build options: ## Cache configuration -There is set of configuration parameters which can be set as environment variables or parameters in `sycl.conf` and affect cache behavior: -| Environment variable | Values | Description | -| -------------------- | ------ | ----------- | -| `SYCL_CACHE_DIR` | Path | Path to persistent cache root directory. Default values are `%AppData%\Intel\sycl_program_cache` for Windows and `$HOME/intel/sycl_program_cache` on Linux. | -| `SYCL_CACHE_DISABLE_PERSISTENT` | Any(\*) | Switches persistent cache switch off. Default value is ON. | -| `SYCL_CACHE_EVICTION_DISABLE` | Any(\*) | Switches cache eviction off. Default value is ON. | -| `SYCL_CACHE_MAX_SIZE` | Positive integer | Cache eviction is triggered once total size of cached images exceeds the value in megabytes (default - 8 192 for 8 GB). Set to 0 to disable size-based cache eviction. | -| `SYCL_CACHE_THRESHOLD` | Positive integer | Cache eviction threshold in days (default value is 7 for 1 week). Set to 0 for disabling time-based cache eviction. | -| `SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE` | Positive integer | Minimum size of device code image in kilobytes which is reasonable to cache on disk because disk access operation may take more time than do JIT compilation for it. Default value is 0 to cache all images. | -| `SYCL_CACHE_MAX_DEVICE_IMAGE_SIZE` | Positive integer | Maximum size of device image in megabytes which is cached. Too big kernels may overload disk too fast. Default value is 0 to cache all images. | +There is set of configuration parameters which can be defined as environment +variables or parameters in `sycl.conf` and affect cache behavior. They are +described in [EnvironmentVariables.md](https://github.com/intel/llvm/blob/sycl/sycl/doc/EnvironmentVariables.md). ## Implementation details diff --git a/sycl/source/detail/os_util.cpp b/sycl/source/detail/os_util.cpp index 11b00e19698f0..7fe543328dd51 100644 --- a/sycl/source/detail/os_util.cpp +++ b/sycl/source/detail/os_util.cpp @@ -278,15 +278,23 @@ std::string OSUtil::getCacheRoot() { if (PersistenCacheRoot) return PersistenCacheRoot; + constexpr char SYCLCacheDir[] = "/libsycl_cache"; + + // Use static to calculate directory only once per program run #if defined(__SYCL_RT_OS_LINUX) - static const char *RootDir = std::getenv("HOME"); + static const char *CacheDir = std::getenv("XDG_CACHE_HOME"); + static const char *HomeDir = std::getenv("HOME"); + static std::string Res{ + std::string(CacheDir + ? CacheDir + : (HomeDir ? std::string(HomeDir) + "/.cache" : ".")) + + SYCLCacheDir}; #else - static const char *RootDir = std::getenv("AppData"); + static const char *AppDataDir = std::getenv("AppData"); + static std::string Res{std::string(AppDataDir ? AppDataDir : ".") + + SYCLCacheDir}; #endif - std::string Root{RootDir ? RootDir : "."}; - - Root += "/intel/sycl_cache"; - return Root; + return Res; } int OSUtil::makeDir(const char *Dir) { diff --git a/sycl/source/detail/persistent_cache.cpp b/sycl/source/detail/persistent_cache.cpp index 43d244eced68f..d883f364df9ba 100644 --- a/sycl/source/detail/persistent_cache.cpp +++ b/sycl/source/detail/persistent_cache.cpp @@ -19,6 +19,13 @@ void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, const SerializedObj &SpecConsts, const std::string &BuildOptionsString, const RT::PiProgram &Program) { + // Only SPIRV images are cached + if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV && + (Img.getFormat() == PI_DEVICE_BINARY_TYPE_NONE && + pi::getBinaryImageFormat(Img.getRawData().BinaryStart, Img.getSize()) != + PI_DEVICE_BINARY_TYPE_SPIRV)) + return; + if (!isPersistentCacheEnabled()) { return; } @@ -65,6 +72,13 @@ std::vector> PersistentCache::getPIProgramFromDisc( const SerializedObj &SpecConsts, const std::string &BuildOptionsString, RT::PiProgram &NativePrg) { + // Only SPIRV images are cached + if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV && + (Img.getFormat() == PI_DEVICE_BINARY_TYPE_NONE && + pi::getBinaryImageFormat(Img.getRawData().BinaryStart, Img.getSize()) != + PI_DEVICE_BINARY_TYPE_SPIRV)) + return {}; + if (!isPersistentCacheEnabled()) return {}; From 5b26cd68c6cbbeb8d8a29040f075efa5a93c7066 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Thu, 25 Mar 2021 15:35:14 +0300 Subject: [PATCH 12/32] Apply review comments and remove redundant change --- sycl/source/detail/config.def | 3 ++- sycl/test/Unit/lit.cfg.py | 1 - sycl/test/on-device/cache/basic.cpp | 7 ------- sycl/test/on-device/cache/basic.hpp | 8 -------- sycl/test/on-device/cache/spec_consts.cpp | 7 ------- sycl/test/on-device/cache/spec_consts.hpp | 8 -------- 6 files changed, 2 insertions(+), 32 deletions(-) diff --git a/sycl/source/detail/config.def b/sycl/source/detail/config.def index 9d6dd937e2bff..4ffe9360b4de9 100644 --- a/sycl/source/detail/config.def +++ b/sycl/source/detail/config.def @@ -20,7 +20,8 @@ CONFIG(SYCL_DEVICE_FILTER, 1024, __SYCL_DEVICE_FILTER) CONFIG(SYCL_PROGRAM_LINK_OPTIONS, 64, __SYCL_PROGRAM_LINK_OPTIONS) CONFIG(SYCL_PROGRAM_COMPILE_OPTIONS, 64, __SYCL_PROGRAM_COMPILE_OPTIONS) CONFIG(SYCL_HOST_UNIFIED_MEMORY, 1, __SYCL_HOST_UNIFIED_MEMORY) -CONFIG(SYCL_CACHE_DIR, 1024, __SYCL_CACHE_DIR) +// 260 (Windows limit) - 12 (filename) - 84 (cache directory structure) +CONFIG(SYCL_CACHE_DIR, 172, __SYCL_CACHE_DIR) CONFIG(SYCL_CACHE_DISABLE_PERSISTENT, 1, __SYCL_CACHE_DISABLE_PERSISTENT) CONFIG(SYCL_CACHE_EVICTION_DISABLE, 1, __SYCL_CACHE_EVICTION_DISABLE) CONFIG(SYCL_CACHE_MAX_SIZE, 16, __SYCL_CACHE_MAX_SIZE) diff --git a/sycl/test/Unit/lit.cfg.py b/sycl/test/Unit/lit.cfg.py index d261014cc20cb..2dbc0d14e9f21 100644 --- a/sycl/test/Unit/lit.cfg.py +++ b/sycl/test/Unit/lit.cfg.py @@ -74,5 +74,4 @@ def find_shlibpath_var(): .format(platform.system())) config.environment['SYCL_DEVICE_FILTER'] = lit_config.params.get('SYCL_PLUGIN', "opencl") -config.environment['SYCL_CACHE_DISABLE_PERSISTENT'] = '1' lit_config.note("Backend: {}".format(config.environment['SYCL_DEVICE_FILTER'])) diff --git a/sycl/test/on-device/cache/basic.cpp b/sycl/test/on-device/cache/basic.cpp index 487379e012b24..e94e7f07bcbef 100644 --- a/sycl/test/on-device/cache/basic.cpp +++ b/sycl/test/on-device/cache/basic.cpp @@ -9,13 +9,6 @@ // RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD // RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE // -//==----------- basic.cpp --------------------------------------------------==// -// -// 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 -// -//===----------------------------------------------------------------------===// // The test checks that caching works properly. #include "basic.hpp" diff --git a/sycl/test/on-device/cache/basic.hpp b/sycl/test/on-device/cache/basic.hpp index 50b92c710d2db..9a1d11597957b 100644 --- a/sycl/test/on-device/cache/basic.hpp +++ b/sycl/test/on-device/cache/basic.hpp @@ -1,11 +1,3 @@ -//==-------------- basic.hpp -----------------------------------------------==// -// -// 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 int main(int argc, char **argv) { diff --git a/sycl/test/on-device/cache/spec_consts.cpp b/sycl/test/on-device/cache/spec_consts.cpp index 9cd7e37423d22..289849a815bad 100644 --- a/sycl/test/on-device/cache/spec_consts.cpp +++ b/sycl/test/on-device/cache/spec_consts.cpp @@ -10,13 +10,6 @@ // RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD // RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE // -//==----------- spec_consts.cpp --------------------------------------------==// -// -// 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 -// -//===----------------------------------------------------------------------===// // The test checks that caching works properly for SYCL application containing // specialization constant values. #include "spec_consts.hpp" diff --git a/sycl/test/on-device/cache/spec_consts.hpp b/sycl/test/on-device/cache/spec_consts.hpp index 7f6142fb55f0f..9285d5b500086 100644 --- a/sycl/test/on-device/cache/spec_consts.hpp +++ b/sycl/test/on-device/cache/spec_consts.hpp @@ -1,11 +1,3 @@ -//==-------------- spec_const.hpp ------------------------------------------==// -// -// 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 #include From 09d3ee73f204f919a450e28f9a876d2b7910c40c Mon Sep 17 00:00:00 2001 From: vladimirlaz Date: Thu, 25 Mar 2021 18:39:47 +0300 Subject: [PATCH 13/32] Update sycl/doc/KernelProgramCache.md --- sycl/doc/KernelProgramCache.md | 5 ++--- 1 file changed, 2 insertions(+), 3 deletions(-) diff --git a/sycl/doc/KernelProgramCache.md b/sycl/doc/KernelProgramCache.md index 95e7ea64dd1a0..7a7c7247cdd85 100644 --- a/sycl/doc/KernelProgramCache.md +++ b/sycl/doc/KernelProgramCache.md @@ -172,9 +172,8 @@ three sources of build options: ## Cache configuration -There is set of configuration parameters which can be defined as environment -variables or parameters in `sycl.conf` and affect cache behavior. They are -described in [EnvironmentVariables.md](https://github.com/intel/llvm/blob/sycl/sycl/doc/EnvironmentVariables.md). +The environment variables which affect cache behavior are described in +[EnvironmentVariables.md](https://github.com/intel/llvm/blob/sycl/sycl/doc/EnvironmentVariables.md). ## Implementation details From 8eff1a2742213e13ef92af4088057e9e4bff4656 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Fri, 26 Mar 2021 12:48:19 +0300 Subject: [PATCH 14/32] Fix misprint --- sycl/source/detail/config.def | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/source/detail/config.def b/sycl/source/detail/config.def index 4ffe9360b4de9..91b9eba9e8d07 100644 --- a/sycl/source/detail/config.def +++ b/sycl/source/detail/config.def @@ -21,7 +21,7 @@ CONFIG(SYCL_PROGRAM_LINK_OPTIONS, 64, __SYCL_PROGRAM_LINK_OPTIONS) CONFIG(SYCL_PROGRAM_COMPILE_OPTIONS, 64, __SYCL_PROGRAM_COMPILE_OPTIONS) CONFIG(SYCL_HOST_UNIFIED_MEMORY, 1, __SYCL_HOST_UNIFIED_MEMORY) // 260 (Windows limit) - 12 (filename) - 84 (cache directory structure) -CONFIG(SYCL_CACHE_DIR, 172, __SYCL_CACHE_DIR) +CONFIG(SYCL_CACHE_DIR, 164, __SYCL_CACHE_DIR) CONFIG(SYCL_CACHE_DISABLE_PERSISTENT, 1, __SYCL_CACHE_DISABLE_PERSISTENT) CONFIG(SYCL_CACHE_EVICTION_DISABLE, 1, __SYCL_CACHE_EVICTION_DISABLE) CONFIG(SYCL_CACHE_MAX_SIZE, 16, __SYCL_CACHE_MAX_SIZE) From ed915f5d1312ff40db8d353c39657be4bff23d9e Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Fri, 26 Mar 2021 17:49:23 +0300 Subject: [PATCH 15/32] Apply review comments and remove redundant code --- sycl/source/detail/persistent_cache.cpp | 43 ++++++++----------------- sycl/source/detail/persistent_cache.hpp | 2 -- 2 files changed, 14 insertions(+), 31 deletions(-) diff --git a/sycl/source/detail/persistent_cache.cpp b/sycl/source/detail/persistent_cache.cpp index d883f364df9ba..9a99d44261289 100644 --- a/sycl/source/detail/persistent_cache.cpp +++ b/sycl/source/detail/persistent_cache.cpp @@ -109,17 +109,6 @@ std::string PersistentCache::getDeviceString(const device &Device) { Device.get_info()}; } -std::string PersistentCache::dumpBinData(const unsigned char *Data, - size_t Size) { - if (!Size) - return "NONE"; - std::stringstream ss; - for (size_t i = 0; i < Size; i++) { - ss << std::hex << (int)Data[i]; - } - return ss.str(); -} - /* Write built binary to persistent cache * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] */ @@ -166,11 +155,7 @@ void PersistentCache::writeCacheItemSrc(const std::string &FileName, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { std::ofstream FileStream{FileName, std::ios::binary}; - std::string ImgString{ - dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; std::string DeviceString{getDeviceString(Device)}; - std::string SpecConstsString{ - dumpBinData(SpecConsts.data(), SpecConsts.size())}; size_t Size = DeviceString.size(); FileStream.write((char *)&Size, sizeof(Size)); @@ -178,12 +163,12 @@ void PersistentCache::writeCacheItemSrc(const std::string &FileName, Size = BuildOptionsString.size(); FileStream.write((char *)&Size, sizeof(Size)); FileStream.write(BuildOptionsString.data(), Size); - Size = SpecConstsString.size(); + Size = SpecConsts.size(); FileStream.write((char *)&Size, sizeof(Size)); - FileStream.write(SpecConstsString.data(), Size); - Size = ImgString.size(); + FileStream.write((const char *)SpecConsts.data(), Size); + Size = Img.getSize(); FileStream.write((char *)&Size, sizeof(Size)); - FileStream.write(ImgString.data(), Size); + FileStream.write((const char *)Img.getRawData().BinaryStart, Size); FileStream.close(); } @@ -194,11 +179,11 @@ bool PersistentCache::isCacheItemSrcEqual( const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { std::ifstream FileStream{FileName, std::ios::binary}; - std::string ImgString{ - dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + std::string ImgString{(const char *)Img.getRawData().BinaryStart, + Img.getSize()}; std::string DeviceString{getDeviceString(Device)}; - std::string SpecConstsString{ - dumpBinData(SpecConsts.data(), SpecConsts.size())}; + std::string SpecConstsString{(const char *)SpecConsts.data(), + SpecConsts.size()}; size_t Size; std::string res; @@ -218,13 +203,13 @@ bool PersistentCache::isCacheItemSrcEqual( FileStream.read((char *)&Size, sizeof(Size)); res.resize(Size); FileStream.read(&res[0], Size); - if (SpecConstsString.compare(0, Size, res.data())) + if (SpecConstsString.compare(res)) return false; FileStream.read((char *)&Size, sizeof(Size)); res.resize(Size); FileStream.read(&res[0], Size); - if (ImgString.compare(0, Size, res.data())) + if (ImgString.compare(res)) return false; FileStream.close(); @@ -236,11 +221,11 @@ std::string PersistentCache::getCacheItemDirName( const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { static std::string cache_root{detail::OSUtil::getCacheRoot()}; - std::string ImgString{ - dumpBinData(Img.getRawData().BinaryStart, Img.getSize())}; + std::string ImgString{(const char *)Img.getRawData().BinaryStart, + Img.getSize()}; std::string DeviceString{getDeviceString(Device)}; - std::string SpecConstsString{ - dumpBinData(SpecConsts.data(), SpecConsts.size())}; + std::string SpecConstsString{(const char *)SpecConsts.data(), + SpecConsts.size()}; std::hash StringHasher{}; return {cache_root + "/" + std::to_string(StringHasher(DeviceString)) + "/" + diff --git a/sycl/source/detail/persistent_cache.hpp b/sycl/source/detail/persistent_cache.hpp index 1a945ae475a01..de25009bdb037 100644 --- a/sycl/source/detail/persistent_cache.hpp +++ b/sycl/source/detail/persistent_cache.hpp @@ -21,8 +21,6 @@ namespace detail { class PersistentCache { /* Form string representing device version */ static std::string getDeviceString(const device &Device); - /* Form string containing hex representation of the C-string*/ - static std::string dumpBinData(const unsigned char *Data, size_t Size); /* Write built binary to persistent cache * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] */ From 9cd8c22429877872cf39d1e70124e1d708ceedc4 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Sun, 28 Mar 2021 22:42:02 +0300 Subject: [PATCH 16/32] Apply review comments --- sycl/source/detail/persistent_cache.cpp | 44 ++++++++++++++++--------- sycl/source/detail/persistent_cache.hpp | 6 ++++ 2 files changed, 35 insertions(+), 15 deletions(-) diff --git a/sycl/source/detail/persistent_cache.cpp b/sycl/source/detail/persistent_cache.cpp index 9a99d44261289..668e45c9b228e 100644 --- a/sycl/source/detail/persistent_cache.cpp +++ b/sycl/source/detail/persistent_cache.cpp @@ -13,12 +13,18 @@ __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { namespace detail { +/* Stores build program in persisten cache + */ void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString, const RT::PiProgram &Program) { + + if (!isPersistentCacheEnabled()) + return; + // Only SPIRV images are cached if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV && (Img.getFormat() == PI_DEVICE_BINARY_TYPE_NONE && @@ -26,10 +32,6 @@ void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, PI_DEVICE_BINARY_TYPE_SPIRV)) return; - if (!isPersistentCacheEnabled()) { - return; - } - std::string DirName = getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); @@ -67,11 +69,18 @@ void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, BuildOptionsString); } +/* Program binaries built for one or more devices are read from persistent + * cache and returned in form of vector of programs. Each binary program is + * stored in vector of chars. + */ std::vector> PersistentCache::getPIProgramFromDisc( const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString, RT::PiProgram &NativePrg) { + if (!isPersistentCacheEnabled()) + return {}; + // Only SPIRV images are cached if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV && (Img.getFormat() == PI_DEVICE_BINARY_TYPE_NONE && @@ -79,9 +88,6 @@ std::vector> PersistentCache::getPIProgramFromDisc( PI_DEVICE_BINARY_TYPE_SPIRV)) return {}; - if (!isPersistentCacheEnabled()) - return {}; - std::string Path{ getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString)}; @@ -102,11 +108,13 @@ std::vector> PersistentCache::getPIProgramFromDisc( return {}; } +/* Returns string value which can be used to identify different device + */ std::string PersistentCache::getDeviceString(const device &Device) { - return {Device.get_platform().get_info() + "/" + - Device.get_info() + "/" + - Device.get_info() + "/" + - Device.get_info()}; + return Device.get_platform().get_info() + "/" + + Device.get_info() + "/" + + Device.get_info() + "/" + + Device.get_info(); } /* Write built binary to persistent cache @@ -216,6 +224,9 @@ bool PersistentCache::isCacheItemSrcEqual( return true; } +/* Returns directory name to store specific kernel image for specified + * device, build options and specialization constants values. + */ std::string PersistentCache::getCacheItemDirName( const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { @@ -228,12 +239,15 @@ std::string PersistentCache::getCacheItemDirName( SpecConsts.size()}; std::hash StringHasher{}; - return {cache_root + "/" + std::to_string(StringHasher(DeviceString)) + "/" + - std::to_string(StringHasher(ImgString)) + "/" + - std::to_string(StringHasher(SpecConstsString)) + "/" + - std::to_string(StringHasher(BuildOptionsString))}; + return cache_root + "/" + std::to_string(StringHasher(DeviceString)) + "/" + + std::to_string(StringHasher(ImgString)) + "/" + + std::to_string(StringHasher(SpecConstsString)) + "/" + + std::to_string(StringHasher(BuildOptionsString)); } +/* Returns true if persistent cache enabled. The cache can be disabled by + * setting SYCL_CACHE_EVICTION_DISABLE environmnet variable. + */ bool PersistentCache::isPersistentCacheEnabled() { static const char *PersistenCacheDisabled = SYCLConfig::get(); diff --git a/sycl/source/detail/persistent_cache.hpp b/sycl/source/detail/persistent_cache.hpp index de25009bdb037..9875cea873942 100644 --- a/sycl/source/detail/persistent_cache.hpp +++ b/sycl/source/detail/persistent_cache.hpp @@ -58,11 +58,17 @@ class PersistentCache { static bool isPersistentCacheEnabled(); public: + /* Program binaries built for one or more devices are read from persistent + * cache and returned in form of vector of programs. Each binary program is + * stored in vector of chars. + */ static std::vector> getPIProgramFromDisc(const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString, RT::PiProgram &NativePrg); + /* Stores build program in persisten cache + */ static void putPIProgramToDisc(const detail::plugin &Plugin, const device &Device, const RTDeviceBinaryImage &Img, From 76a433d4d99a4ef06ce6d9640373fd4e3b84ec0f Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Tue, 30 Mar 2021 18:45:27 +0300 Subject: [PATCH 17/32] Fix status --- sycl/source/detail/os_util.cpp | 6 +- sycl/source/detail/persistent_cache.cpp | 15 +- sycl/source/detail/persistent_cache.hpp | 32 ++++ .../kernel-and-program/CMakeLists.txt | 1 + .../PersistenCacheConcurrentAccess.cpp | 143 ++++++++++++++++++ sycl/unittests/thread_safety/ThreadUtils.h | 18 +++ 6 files changed, 204 insertions(+), 11 deletions(-) create mode 100644 sycl/unittests/kernel-and-program/PersistenCacheConcurrentAccess.cpp diff --git a/sycl/source/detail/os_util.cpp b/sycl/source/detail/os_util.cpp index 7fe543328dd51..0f898bc2302dc 100644 --- a/sycl/source/detail/os_util.cpp +++ b/sycl/source/detail/os_util.cpp @@ -309,10 +309,10 @@ int OSUtil::makeDir(const char *Dir) { free(CurDir); -#if defined(__SYCL_RT_OS_WINDOWS) - return _mkdir(Dir); -#else +#if defined(__SYCL_RT_OS_LINUX) return mkdir(Dir, 0777); +#else + return _mkdir(Dir); #endif } diff --git a/sycl/source/detail/persistent_cache.cpp b/sycl/source/detail/persistent_cache.cpp index 668e45c9b228e..286e9d5e01a35 100644 --- a/sycl/source/detail/persistent_cache.cpp +++ b/sycl/source/detail/persistent_cache.cpp @@ -88,15 +88,15 @@ std::vector> PersistentCache::getPIProgramFromDisc( PI_DEVICE_BINARY_TYPE_SPIRV)) return {}; - std::string Path{ - getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString)}; + std::string Path = + getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); if (!OSUtil::isPathPresent(Path)) return {}; int i = 0; std::string FileName{Path + "/" + std::to_string(i)}; - while (OSUtil::isPathPresent(FileName + ".bin") && + while (OSUtil::isPathPresent(FileName + ".bin") || OSUtil::isPathPresent(FileName + ".src")) { if (isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString)) { @@ -104,7 +104,6 @@ std::vector> PersistentCache::getPIProgramFromDisc( } FileName = Path + "/" + std::to_string(++i); } - return {}; } @@ -139,15 +138,15 @@ void PersistentCache::writeCacheItemBin( */ std::vector> PersistentCache::readCacheItem(const std::string &FileName) { - std::vector> Res; std::ifstream FileStream{FileName, std::ios::binary}; size_t ImgNum, ImgSize; FileStream.read((char *)&ImgNum, sizeof(ImgNum)); - Res.resize(ImgNum); + std::vector> Res(ImgNum); for (size_t i = 0; i < ImgNum; ++i) { FileStream.read((char *)&ImgSize, sizeof(ImgSize)); - Res[i].resize(ImgSize); - FileStream.read(Res[i].data(), ImgSize); + std::vector ImgData(ImgSize); + FileStream.read(ImgData.data(), ImgSize); + Res[i] = std::move(ImgData); } return Res; diff --git a/sycl/source/detail/persistent_cache.hpp b/sycl/source/detail/persistent_cache.hpp index 9875cea873942..c4872233ef0ee 100644 --- a/sycl/source/detail/persistent_cache.hpp +++ b/sycl/source/detail/persistent_cache.hpp @@ -12,13 +12,45 @@ #include #include #include +#include #include +#include +#include #include __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { namespace detail { + +/* This is temporary solution until std::filesystem is available when SYCL RT + * is moved to c++17 standard*/ +/*class LockCacheItem { + const std::string FileName; + +public: + LockCacheItem(const std::string &DirName) : FileName(DirName + ".lock") { + int fd; + while ((fd = open(FileName.c_str(), O_CREAT | O_EXCL, S_IWRITE)) == -1) { + std::this_thread::yield(); + } + close(fd); + } + ~LockCacheItem() { std::remove(FileName.c_str()); } +}; +bool lockCacheDir(const std::string &DirName) { + int fd = open((DirName+".lock").c_str(), O_CREAT | O_EXCL, S_IWRITE); + if(fd!=-1){ + close(fd); + return true; + } + return false; +} +void unlockCacheDir(const std::string &DirName) { + std::remove(DirName+".lock"); +}*/ + class PersistentCache { + /* Form string representing device version */ static std::string getDeviceString(const device &Device); /* Write built binary to persistent cache diff --git a/sycl/unittests/kernel-and-program/CMakeLists.txt b/sycl/unittests/kernel-and-program/CMakeLists.txt index f61498b901bd9..35b51a304e771 100644 --- a/sycl/unittests/kernel-and-program/CMakeLists.txt +++ b/sycl/unittests/kernel-and-program/CMakeLists.txt @@ -2,4 +2,5 @@ add_sycl_unittest_with_device(KernelAndProgramTests OBJECT Cache.cpp KernelRelease.cpp KernelInfo.cpp + PersistenCacheConcurrentAccess.cpp ) diff --git a/sycl/unittests/kernel-and-program/PersistenCacheConcurrentAccess.cpp b/sycl/unittests/kernel-and-program/PersistenCacheConcurrentAccess.cpp new file mode 100644 index 0000000000000..0841addc03555 --- /dev/null +++ b/sycl/unittests/kernel-and-program/PersistenCacheConcurrentAccess.cpp @@ -0,0 +1,143 @@ +//==----- PersistenCacheConcurrentAccess.cpp --- Persistent cache tests ----==// +// +// 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 "../thread_safety/ThreadUtils.h" +#include "detail/persistent_cache.hpp" +#include +#include +#include +#include +#include +#include +#include +#include + +namespace { +constexpr auto sycl_read_write = cl::sycl::access::mode::read_write; +using namespace cl::sycl; +constexpr size_t BinNum = 4; +constexpr size_t BinSizes[BinNum] = {1024, 1024 * 1024, 256, 1024 * 64}; +static pi_result redefinedProgramGetInfo(pi_program program, + pi_program_info param_name, + size_t param_value_size, + void *param_value, + size_t *param_value_size_ret) { + if (param_name == PI_PROGRAM_INFO_NUM_DEVICES) { + auto value = reinterpret_cast(param_value); + *value = BinNum; + } + + if (param_name == PI_PROGRAM_INFO_BINARY_SIZES) { + auto value = reinterpret_cast(param_value); + for (int i = 0; i < BinNum; ++i) + value[i] = BinSizes[i]; + } + + if (param_name == PI_PROGRAM_INFO_BINARIES) { + auto value = reinterpret_cast(param_value); + for (int i = 0; i < BinNum; ++i) + for (int j = 0; j < BinSizes[i]; ++j) + value[i][j] = i; + } + + return PI_SUCCESS; +} + +class PersistenCacheConcurrentAccess : public ::testing::Test { +public: +#ifdef _WIN32 + int setenv(const char *name, const char *value, int overwrite) { + int errcode = 0; + if (!overwrite) { + size_t envsize = 0; + errcode = getenv_s(&envsize, NULL, 0, name); + if (errcode || envsize) + return errcode; + } + return _putenv_s(name, value); + } +#endif + + PersistenCacheConcurrentAccess() : Plt{default_selector()} { + const char *envTmp = +#ifdef _WIN32 + std::getenv("TEMP"); +#else + std::getenv("TMP"); +#endif + if (envTmp != nullptr) + cacheRoot += envTmp; + else +#ifdef _WIN32 + cacheRoot += "C:/temp"; +#else + cacheRoot += "/tmp"; +#endif + cacheRoot += "/PersistenCache"; + setenv("SYCL_CACHE_DIR", cacheRoot.c_str(), 0); + std::printf("Use %s as cache root\n", cacheRoot.c_str()); + + if (Plt.is_host() || Plt.get_backend() != backend::opencl) { + std::clog << "This test is only supported on OpenCL devices\n"; + std::clog << "Current platform is " + << Plt.get_info(); + return; + } + + Mock = std::make_unique(Plt); + Dev = Plt.get_devices()[0]; + Mock->redefine( + redefinedProgramGetInfo); + std::filesystem::remove_all(cacheRoot); + } + +protected: + std::string cacheRoot; + detail::OSModuleHandle ModuleHandle = detail::OSUtil::ExeModuleHandle; + platform Plt; + device Dev; + pi_device_binary_struct BinStruct{/*Version*/ 1, /*Kind*/ 4, + /*Format*/ PI_DEVICE_BINARY_TYPE_SPIRV}; + pi_device_binary Bin = &BinStruct; + detail::RTDeviceBinaryImage Img{Bin, ModuleHandle}; + RT::PiProgram NativeProg; + std::unique_ptr Mock; +}; +TEST_F(PersistenCacheConcurrentAccess, ReadWriteCacheItem) { + std::vector> Data = { + std::vector(1024, '1'), std::vector(1024 * 1024, '2'), + std::vector(256, '3'), std::vector(1024 * 64, '4')}; + + constexpr std::size_t threadCount = 300; + + Barrier b(threadCount); + { + auto testLambda = [&](std::size_t threadId) { + b.wait(); + detail::PersistentCache::putPIProgramToDisc( + detail::getSyclObjImpl(Plt)->getPlugin(), Dev, Img, + sycl::vector_class( + {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't'}), + "--build-options", NativeProg); + auto res = detail::PersistentCache::getPIProgramFromDisc( + Dev, Img, + sycl::vector_class( + {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't'}), + "--build-options", NativeProg); + for (int i = 0; i < res.size(); ++i) { + for (int j = 0; j < res[i].size(); ++j) { + assert(res[i][j] == i && + "Corrupted image loaded from persistent cache"); + } + } + }; + + ThreadPool MPool(threadCount, testLambda); + } +} +} // namespace diff --git a/sycl/unittests/thread_safety/ThreadUtils.h b/sycl/unittests/thread_safety/ThreadUtils.h index 18a50bb37a7b9..021f4087ab07f 100644 --- a/sycl/unittests/thread_safety/ThreadUtils.h +++ b/sycl/unittests/thread_safety/ThreadUtils.h @@ -2,6 +2,24 @@ #include #include +class Barrier { +public: + Barrier() = delete; + explicit Barrier(std::size_t count) : threadNum(count) {} + void wait() { + std::unique_lock lock(mutex); + if (--threadNum == 0) { + cv.notify_all(); + } else { + cv.wait(lock, [this] { return threadNum == 0; }); + } + } + +private: + std::mutex mutex; + std::condition_variable cv; + std::size_t threadNum; +}; class ThreadPool { public: From a78d7a505de5a45f79b3149f9c1ca14c77868001 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Wed, 31 Mar 2021 14:22:37 +0300 Subject: [PATCH 18/32] Revert API changes to simplify migration to std::filesystem --- sycl/include/CL/sycl/detail/os_util.hpp | 13 ---- sycl/source/detail/os_util.cpp | 43 ------------- sycl/source/detail/persistent_cache.cpp | 81 +++++++++++++++++++++++-- sycl/source/detail/persistent_cache.hpp | 3 + sycl/test/abi/sycl_symbols_linux.dump | 2 - 5 files changed, 78 insertions(+), 64 deletions(-) diff --git a/sycl/include/CL/sycl/detail/os_util.hpp b/sycl/include/CL/sycl/detail/os_util.hpp index ad965b974de8c..271943484f85b 100644 --- a/sycl/include/CL/sycl/detail/os_util.hpp +++ b/sycl/include/CL/sycl/detail/os_util.hpp @@ -16,7 +16,6 @@ #include #include #include -#include #ifdef _WIN32 #define __SYCL_RT_OS_WINDOWS @@ -58,15 +57,6 @@ class __SYCL_EXPORT OSUtil { /// Returns a directory component of a path. static std::string getDirName(const char *Path); - /// Checks if specified path is present - static inline bool isPathPresent(const std::string &Path) { - struct stat Stat; - return !stat(Path.c_str(), &Stat); - } - - /// Creates directory recursively - static int makeDir(const char *Dir); - /// Module handle for the executable module - it is assumed there is always /// single one at most. static constexpr OSModuleHandle ExeModuleHandle = -1; @@ -90,9 +80,6 @@ class __SYCL_EXPORT OSUtil { /// Deallocates the memory referenced by \p Ptr. static void alignedFree(void *Ptr); - - /// Returns the path to directory storing on-disk SYCL program cache. - static std::string getCacheRoot(); }; } // namespace detail diff --git a/sycl/source/detail/os_util.cpp b/sycl/source/detail/os_util.cpp index 0f898bc2302dc..a95991ae3df01 100644 --- a/sycl/source/detail/os_util.cpp +++ b/sycl/source/detail/os_util.cpp @@ -273,49 +273,6 @@ void OSUtil::alignedFree(void *Ptr) { #endif } -std::string OSUtil::getCacheRoot() { - static const char *PersistenCacheRoot = SYCLConfig::get(); - if (PersistenCacheRoot) - return PersistenCacheRoot; - - constexpr char SYCLCacheDir[] = "/libsycl_cache"; - - // Use static to calculate directory only once per program run -#if defined(__SYCL_RT_OS_LINUX) - static const char *CacheDir = std::getenv("XDG_CACHE_HOME"); - static const char *HomeDir = std::getenv("HOME"); - static std::string Res{ - std::string(CacheDir - ? CacheDir - : (HomeDir ? std::string(HomeDir) + "/.cache" : ".")) + - SYCLCacheDir}; -#else - static const char *AppDataDir = std::getenv("AppData"); - static std::string Res{std::string(AppDataDir ? AppDataDir : ".") + - SYCLCacheDir}; -#endif - return Res; -} - -int OSUtil::makeDir(const char *Dir) { - assert((Dir != nullptr) && "Passed null-pointer as directory name."); - - // Directory is present - do nothing - if (isPathPresent(Dir)) - return 0; - - char *CurDir = strdup(Dir); - makeDir(getDirName(CurDir).c_str()); - - free(CurDir); - -#if defined(__SYCL_RT_OS_LINUX) - return mkdir(Dir, 0777); -#else - return _mkdir(Dir); -#endif -} - } // namespace detail } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/detail/persistent_cache.cpp b/sycl/source/detail/persistent_cache.cpp index 286e9d5e01a35..1011da17d7139 100644 --- a/sycl/source/detail/persistent_cache.cpp +++ b/sycl/source/detail/persistent_cache.cpp @@ -13,6 +13,49 @@ __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { namespace detail { +// These are temporary implementation of file operations until moving to C++17 +// and use of std::filesystem instead + +std::string getDirName(const char *Path) { + std::string Tmp(Path); + // Remove trailing directory separators + Tmp.erase(Tmp.find_last_not_of("/\\") + 1, std::string::npos); + + auto pos = Tmp.find_last_of("/\\"); + if (pos != std::string::npos) + return Tmp.substr(0, pos); + + // If no directory separator is present return initial path like dirname does + return Tmp; +} + +#include +/// Checks if specified path is present +static inline bool isPathPresent(const std::string &Path) { + struct stat Stat; + return !stat(Path.c_str(), &Stat); +} + +int makeDir(const char *Dir) { + assert((Dir != nullptr) && "Passed null-pointer as directory name."); + + // Directory is present - do nothing + if (isPathPresent(Dir)) + return 0; + + char *CurDir = strdup(Dir); + makeDir(getDirName(CurDir).c_str()); + + free(CurDir); + +#if defined(__SYCL_RT_OS_LINUX) + return mkdir(Dir, 0777); +#else + return _mkdir(Dir); +#endif +} + + /* Stores build program in persisten cache */ void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, @@ -39,7 +82,7 @@ void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, std::string FileName; do { FileName = DirName + "/" + std::to_string(i++); - } while (OSUtil::isPathPresent(FileName + ".bin")); + } while (isPathPresent(FileName + ".bin")); unsigned int DeviceNum = 0; @@ -63,7 +106,7 @@ void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, sizeof(char *) * Pointers.size(), Pointers.data(), nullptr); - OSUtil::makeDir(DirName.c_str()); + makeDir(DirName.c_str()); writeCacheItemBin(FileName + ".bin", Result); writeCacheItemSrc(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString); @@ -91,13 +134,13 @@ std::vector> PersistentCache::getPIProgramFromDisc( std::string Path = getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); - if (!OSUtil::isPathPresent(Path)) + if (!isPathPresent(Path)) return {}; int i = 0; std::string FileName{Path + "/" + std::to_string(i)}; - while (OSUtil::isPathPresent(FileName + ".bin") || - OSUtil::isPathPresent(FileName + ".src")) { + while (isPathPresent(FileName + ".bin") || + isPathPresent(FileName + ".src")) { if (isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString)) { return readCacheItem(FileName + ".bin"); @@ -229,7 +272,7 @@ bool PersistentCache::isCacheItemSrcEqual( std::string PersistentCache::getCacheItemDirName( const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { - static std::string cache_root{detail::OSUtil::getCacheRoot()}; + static std::string cache_root{getDeviceCodeCacheRoot()}; std::string ImgString{(const char *)Img.getRawData().BinaryStart, Img.getSize()}; @@ -253,6 +296,32 @@ bool PersistentCache::isPersistentCacheEnabled() { return !PersistenCacheDisabled; } +/* Returns path for device code cache root directory + */ +std::string PersistentCache::getDeviceCodeCacheRoot() { + static const char *RootDir = SYCLConfig::get(); + if (RootDir) + return RootDir; + + constexpr char DeviceCodeCacheDir[] = "/libsycl_cache"; + + // Use static to calculate directory only once per program run +#if defined(__SYCL_RT_OS_LINUX) + static const char *CacheDir = std::getenv("XDG_CACHE_HOME"); + static const char *HomeDir = std::getenv("HOME"); + static std::string Res{ + std::string(CacheDir + ? CacheDir + : (HomeDir ? std::string(HomeDir) + "/.cache" : ".")) + + DeviceCodeCacheDir}; +#else + static const char *AppDataDir = std::getenv("AppData"); + static std::string Res{std::string(AppDataDir ? AppDataDir : ".") + + DeviceCodeCacheDir}; +#endif + return Res; +} + } // namespace detail } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/detail/persistent_cache.hpp b/sycl/source/detail/persistent_cache.hpp index c4872233ef0ee..fc9e7eb35b976 100644 --- a/sycl/source/detail/persistent_cache.hpp +++ b/sycl/source/detail/persistent_cache.hpp @@ -89,6 +89,9 @@ class PersistentCache { */ static bool isPersistentCacheEnabled(); + /* Returns the path to directory storing persistent device code cache.*/ + static std::string getDeviceCodeCacheRoot(); + public: /* Program binaries built for one or more devices are read from persistent * cache and returned in form of vector of programs. Each binary program is diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index 50c49b1532f8f..f9540b8637696 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3801,11 +3801,9 @@ _ZN2cl4sycl6detail2pi9assertionEbPKc _ZN2cl4sycl6detail6OSUtil10getDirNameB5cxx11EPKc _ZN2cl4sycl6detail6OSUtil11alignedFreeEPv _ZN2cl4sycl6detail6OSUtil12alignedAllocEmm -_ZN2cl4sycl6detail6OSUtil12getCacheRootB5cxx11Ev _ZN2cl4sycl6detail6OSUtil12getOSMemSizeEv _ZN2cl4sycl6detail6OSUtil16getCurrentDSODirB5cxx11Ev _ZN2cl4sycl6detail6OSUtil17getOSModuleHandleEPKv -_ZN2cl4sycl6detail6OSUtil7makeDirEPKc _ZN2cl4sycl6device11get_devicesENS0_4info11device_typeE _ZN2cl4sycl6deviceC1EP13_cl_device_id _ZN2cl4sycl6deviceC1ERKNS0_15device_selectorE From a1be142fb2364c38289fa4d2ebefd873ee3f6c5a Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Wed, 31 Mar 2021 23:18:50 +0300 Subject: [PATCH 19/32] Upload for testing --- sycl/source/CMakeLists.txt | 2 +- ...e.cpp => persistent_device_code_cache.cpp} | 111 ++++++++---------- ...e.hpp => persistent_device_code_cache.hpp} | 94 ++++++++------- .../program_manager/program_manager.cpp | 12 +- .../kernel-and-program/CMakeLists.txt | 2 +- ...cess.cpp => PersistentDeviceCodeCache.cpp} | 67 +++++++++-- 6 files changed, 169 insertions(+), 119 deletions(-) rename sycl/source/detail/{persistent_cache.cpp => persistent_device_code_cache.cpp} (73%) rename sycl/source/detail/{persistent_cache.hpp => persistent_device_code_cache.hpp} (58%) rename sycl/unittests/kernel-and-program/{PersistenCacheConcurrentAccess.cpp => PersistentDeviceCodeCache.cpp} (59%) diff --git a/sycl/source/CMakeLists.txt b/sycl/source/CMakeLists.txt index 7970cb3d55399..3ada2e8f47058 100644 --- a/sycl/source/CMakeLists.txt +++ b/sycl/source/CMakeLists.txt @@ -117,7 +117,6 @@ set(SYCL_SOURCES "detail/device_binary_image.cpp" "detail/device_filter.cpp" "detail/device_impl.cpp" - "detail/persistent_cache.cpp" "detail/error_handling/enqueue_kernel.cpp" "detail/event_impl.cpp" "detail/filter_selector_impl.cpp" @@ -136,6 +135,7 @@ set(SYCL_SOURCES "detail/queue_impl.cpp" "detail/online_compiler/online_compiler.cpp" "detail/os_util.cpp" + "detail/persistent_device_code_cache.cpp" "detail/platform_util.cpp" "detail/reduction.cpp" "detail/sampler_impl.cpp" diff --git a/sycl/source/detail/persistent_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp similarity index 73% rename from sycl/source/detail/persistent_cache.cpp rename to sycl/source/detail/persistent_device_code_cache.cpp index 1011da17d7139..70a2db610c569 100644 --- a/sycl/source/detail/persistent_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -7,15 +7,16 @@ //===----------------------------------------------------------------------===// #include -#include +#include +#include #include __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { namespace detail { -// These are temporary implementation of file operations until moving to C++17 -// and use of std::filesystem instead +/* This is temporary solution until std::filesystem is available when SYCL RT + * is moved to c++17 standard*/ std::string getDirName(const char *Path) { std::string Tmp(Path); // Remove trailing directory separators @@ -29,16 +30,8 @@ std::string getDirName(const char *Path) { return Tmp; } -#include -/// Checks if specified path is present -static inline bool isPathPresent(const std::string &Path) { - struct stat Stat; - return !stat(Path.c_str(), &Stat); -} - int makeDir(const char *Dir) { assert((Dir != nullptr) && "Passed null-pointer as directory name."); - // Directory is present - do nothing if (isPathPresent(Dir)) return 0; @@ -55,28 +48,23 @@ int makeDir(const char *Dir) { #endif } - /* Stores build program in persisten cache */ -void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, - const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString, - const RT::PiProgram &Program) { - - if (!isPersistentCacheEnabled()) +void PersistentDeviceCodeCache::putItemToDisc( + const device &Device, const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, const std::string &BuildOptionsString, + const RT::PiProgram &NativePrg) { + + if (!isEnabled()) return; // Only SPIRV images are cached - if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV && - (Img.getFormat() == PI_DEVICE_BINARY_TYPE_NONE && - pi::getBinaryImageFormat(Img.getRawData().BinaryStart, Img.getSize()) != - PI_DEVICE_BINARY_TYPE_SPIRV)) + if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV) return; + auto Plugin = detail::getSyclObjImpl(Device)->getPlugin(); std::string DirName = - getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); + getCacheItemPath(Device, Img, SpecConsts, BuildOptionsString); size_t i = 0; std::string FileName; @@ -86,13 +74,13 @@ void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, unsigned int DeviceNum = 0; - Plugin.call(Program, PI_PROGRAM_INFO_NUM_DEVICES, - sizeof(DeviceNum), &DeviceNum, - nullptr); + Plugin.call( + NativePrg, PI_PROGRAM_INFO_NUM_DEVICES, sizeof(DeviceNum), &DeviceNum, + nullptr); std::vector BinarySizes(DeviceNum); Plugin.call( - Program, PI_PROGRAM_INFO_BINARY_SIZES, + NativePrg, PI_PROGRAM_INFO_BINARY_SIZES, sizeof(size_t) * BinarySizes.size(), BinarySizes.data(), nullptr); std::vector> Result; @@ -102,48 +90,54 @@ void PersistentCache::putPIProgramToDisc(const detail::plugin &Plugin, Pointers.push_back(Result[I].data()); } - Plugin.call(Program, PI_PROGRAM_INFO_BINARIES, + Plugin.call(NativePrg, PI_PROGRAM_INFO_BINARIES, sizeof(char *) * Pointers.size(), Pointers.data(), nullptr); - makeDir(DirName.c_str()); - writeCacheItemBin(FileName + ".bin", Result); - writeCacheItemSrc(FileName + ".src", Device, Img, SpecConsts, + try { + makeDir(DirName.c_str()); + LockCacheItem Lock{DirName}; + writeBinaryDataToFile(FileName + ".bin", Result); + writeSourceItem(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString); + } catch (...) { + // If a problem happens on storing cache item, do nothing + } } /* Program binaries built for one or more devices are read from persistent * cache and returned in form of vector of programs. Each binary program is * stored in vector of chars. */ -std::vector> PersistentCache::getPIProgramFromDisc( +std::vector> PersistentDeviceCodeCache::getItemFromDisc( const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString, RT::PiProgram &NativePrg) { - if (!isPersistentCacheEnabled()) + if (!isEnabled()) return {}; // Only SPIRV images are cached - if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV && - (Img.getFormat() == PI_DEVICE_BINARY_TYPE_NONE && - pi::getBinaryImageFormat(Img.getRawData().BinaryStart, Img.getSize()) != - PI_DEVICE_BINARY_TYPE_SPIRV)) + if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV) return {}; std::string Path = - getCacheItemDirName(Device, Img, SpecConsts, BuildOptionsString); + getCacheItemPath(Device, Img, SpecConsts, BuildOptionsString); if (!isPathPresent(Path)) return {}; int i = 0; + + // If cache directory is locked ignore cache + if (LockCacheItem::isLocked(Path)) + return {}; + std::string FileName{Path + "/" + std::to_string(i)}; - while (isPathPresent(FileName + ".bin") || - isPathPresent(FileName + ".src")) { + while (isPathPresent(FileName + ".bin") || isPathPresent(FileName + ".src")) { if (isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString)) { - return readCacheItem(FileName + ".bin"); + return readBinaryDataFromFile(FileName + ".bin"); } FileName = Path + "/" + std::to_string(++i); } @@ -152,7 +146,7 @@ std::vector> PersistentCache::getPIProgramFromDisc( /* Returns string value which can be used to identify different device */ -std::string PersistentCache::getDeviceString(const device &Device) { +std::string PersistentDeviceCodeCache::getDeviceIDString(const device &Device) { return Device.get_platform().get_info() + "/" + Device.get_info() + "/" + Device.get_info() + "/" + @@ -162,7 +156,7 @@ std::string PersistentCache::getDeviceString(const device &Device) { /* Write built binary to persistent cache * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] */ -void PersistentCache::writeCacheItemBin( +void PersistentDeviceCodeCache::writeBinaryDataToFile( const std::string &FileName, const std::vector> &Data) { std::ofstream FileStream{FileName, std::ios::binary}; @@ -180,7 +174,7 @@ void PersistentCache::writeCacheItemBin( * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] */ std::vector> -PersistentCache::readCacheItem(const std::string &FileName) { +PersistentDeviceCodeCache::readBinaryDataFromFile(const std::string &FileName) { std::ifstream FileStream{FileName, std::ios::binary}; size_t ImgNum, ImgSize; FileStream.read((char *)&ImgNum, sizeof(ImgNum)); @@ -199,13 +193,12 @@ PersistentCache::readCacheItem(const std::string &FileName) { * Format: Four pairs of [size, value] for device, build options, specialization * constant values, device code SPIR-V image. */ -void PersistentCache::writeCacheItemSrc(const std::string &FileName, - const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString) { +void PersistentDeviceCodeCache::writeSourceItem( + const std::string &FileName, const device &Device, + const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, + const std::string &BuildOptionsString) { std::ofstream FileStream{FileName, std::ios::binary}; - std::string DeviceString{getDeviceString(Device)}; + std::string DeviceString{getDeviceIDString(Device)}; size_t Size = DeviceString.size(); FileStream.write((char *)&Size, sizeof(Size)); @@ -224,14 +217,14 @@ void PersistentCache::writeCacheItemSrc(const std::string &FileName, /* Check that cache item key sources are equal to the current program */ -bool PersistentCache::isCacheItemSrcEqual( +bool PersistentDeviceCodeCache::isCacheItemSrcEqual( const std::string &FileName, const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { std::ifstream FileStream{FileName, std::ios::binary}; std::string ImgString{(const char *)Img.getRawData().BinaryStart, Img.getSize()}; - std::string DeviceString{getDeviceString(Device)}; + std::string DeviceString{getDeviceIDString(Device)}; std::string SpecConstsString{(const char *)SpecConsts.data(), SpecConsts.size()}; @@ -269,14 +262,14 @@ bool PersistentCache::isCacheItemSrcEqual( /* Returns directory name to store specific kernel image for specified * device, build options and specialization constants values. */ -std::string PersistentCache::getCacheItemDirName( +std::string PersistentDeviceCodeCache::getCacheItemPath( const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { - static std::string cache_root{getDeviceCodeCacheRoot()}; + static std::string cache_root{getRootDir()}; std::string ImgString{(const char *)Img.getRawData().BinaryStart, Img.getSize()}; - std::string DeviceString{getDeviceString(Device)}; + std::string DeviceString{getDeviceIDString(Device)}; std::string SpecConstsString{(const char *)SpecConsts.data(), SpecConsts.size()}; std::hash StringHasher{}; @@ -290,7 +283,7 @@ std::string PersistentCache::getCacheItemDirName( /* Returns true if persistent cache enabled. The cache can be disabled by * setting SYCL_CACHE_EVICTION_DISABLE environmnet variable. */ -bool PersistentCache::isPersistentCacheEnabled() { +bool PersistentDeviceCodeCache::isEnabled() { static const char *PersistenCacheDisabled = SYCLConfig::get(); return !PersistenCacheDisabled; @@ -298,7 +291,7 @@ bool PersistentCache::isPersistentCacheEnabled() { /* Returns path for device code cache root directory */ -std::string PersistentCache::getDeviceCodeCacheRoot() { +std::string PersistentDeviceCodeCache::getRootDir() { static const char *RootDir = SYCLConfig::get(); if (RootDir) return RootDir; diff --git a/sycl/source/detail/persistent_cache.hpp b/sycl/source/detail/persistent_device_code_cache.hpp similarity index 58% rename from sycl/source/detail/persistent_cache.hpp rename to sycl/source/detail/persistent_device_code_cache.hpp index fc9e7eb35b976..c978b5e9705a1 100644 --- a/sycl/source/detail/persistent_cache.hpp +++ b/sycl/source/detail/persistent_device_code_cache.hpp @@ -24,54 +24,57 @@ namespace detail { /* This is temporary solution until std::filesystem is available when SYCL RT * is moved to c++17 standard*/ -/*class LockCacheItem { +std::string getDirName(const char *Path); + +#include +/// Checks if specified path is present +inline bool isPathPresent(const std::string &Path) { + struct stat Stat; + return !stat(Path.c_str(), &Stat); +} + +int makeDir(const char *Dir); + +class LockCacheItem { const std::string FileName; public: - LockCacheItem(const std::string &DirName) : FileName(DirName + ".lock") { + LockCacheItem(const std::string &DirName) : FileName(DirName + "/.lock") { int fd; while ((fd = open(FileName.c_str(), O_CREAT | O_EXCL, S_IWRITE)) == -1) { std::this_thread::yield(); } close(fd); } + static bool isLocked(const std::string &DirName) { + return isPathPresent(DirName + "/.lock"); + } ~LockCacheItem() { std::remove(FileName.c_str()); } }; -bool lockCacheDir(const std::string &DirName) { - int fd = open((DirName+".lock").c_str(), O_CREAT | O_EXCL, S_IWRITE); - if(fd!=-1){ - close(fd); - return true; - } - return false; -} -void unlockCacheDir(const std::string &DirName) { - std::remove(DirName+".lock"); -}*/ -class PersistentCache { - - /* Form string representing device version */ - static std::string getDeviceString(const device &Device); +class PersistentDeviceCodeCache { +private: /* Write built binary to persistent cache * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] */ - static void writeCacheItemBin(const std::string &FileName, - const std::vector> &Data); + static void writeBinaryDataToFile(const std::string &FileName, + const std::vector> &Data); + /* Read built binary to persistent cache * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] */ static std::vector> - readCacheItem(const std::string &FileName); + readBinaryDataFromFile(const std::string &FileName); + /* Writing cache item key sources to be used for reliable identification * Format: Four pairs of [size, value] for device, build options, * specialization constant values, device code SPIR-V image. */ - static void writeCacheItemSrc(const std::string &FileName, - const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString); + static void writeSourceItem(const std::string &FileName, const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString); + /* Check that cache item key sources are equal to the current program */ static bool isCacheItemSrcEqual(const std::string &FileName, @@ -79,37 +82,42 @@ class PersistentCache { const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString); - /* Get directory name for storing current cache item - */ - static std::string getCacheItemDirName(const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString); + /* Check if on-disk cache enabled. */ - static bool isPersistentCacheEnabled(); + static bool isEnabled(); /* Returns the path to directory storing persistent device code cache.*/ - static std::string getDeviceCodeCacheRoot(); + static std::string getRootDir(); + + /* Form string representing device version */ + static std::string getDeviceIDString(const device &Device); public: + /* Get directory name for storing current cache item + */ + static std::string getCacheItemPath(const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString); + /* Program binaries built for one or more devices are read from persistent * cache and returned in form of vector of programs. Each binary program is * stored in vector of chars. */ static std::vector> - getPIProgramFromDisc(const device &Device, const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString, - RT::PiProgram &NativePrg); + getItemFromDisc(const device &Device, const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString, + RT::PiProgram &NativePrg); + /* Stores build program in persisten cache */ - static void putPIProgramToDisc(const detail::plugin &Plugin, - const device &Device, - const RTDeviceBinaryImage &Img, - const SerializedObj &SpecConsts, - const std::string &BuildOptionsString, - const RT::PiProgram &Program); + static void putItemToDisc(const device &Device, + const RTDeviceBinaryImage &Img, + const SerializedObj &SpecConsts, + const std::string &BuildOptionsString, + const RT::PiProgram &NativePrg); }; } // namespace detail } // namespace sycl diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 7090f34ba673b..5826065d9e411 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -21,7 +21,7 @@ #include #include #include -#include +#include #include #include #include @@ -421,7 +421,8 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, ContextImplPtr ContextImpl = getSyclObjImpl(Context); const detail::plugin &Plugin = ContextImpl->getPlugin(); RT::PiProgram NativePrg; - auto BinProg = PersistentCache::getPIProgramFromDisc( + + auto BinProg = PersistentDeviceCodeCache::getItemFromDisc( Device, Img, SpecConsts, CompileOpts + LinkOpts, NativePrg); if (BinProg.size()) { // TODO: Build for multiple devices once supported by program manager @@ -457,10 +458,11 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, std::lock_guard Lock(MNativeProgramsMutex); NativePrograms[BuiltProgram.get()] = &Img; } + + // Save program to persistent cache if it not there if (!BinProg.size()) - PersistentCache::putPIProgramToDisc(Plugin, Device, Img, SpecConsts, - CompileOpts + LinkOpts, - BuiltProgram.get()); + PersistentDeviceCodeCache::putItemToDisc( + Device, Img, SpecConsts, CompileOpts + LinkOpts, BuiltProgram.get()); return BuiltProgram.release(); }; diff --git a/sycl/unittests/kernel-and-program/CMakeLists.txt b/sycl/unittests/kernel-and-program/CMakeLists.txt index 35b51a304e771..e5033a6568b0f 100644 --- a/sycl/unittests/kernel-and-program/CMakeLists.txt +++ b/sycl/unittests/kernel-and-program/CMakeLists.txt @@ -2,5 +2,5 @@ add_sycl_unittest_with_device(KernelAndProgramTests OBJECT Cache.cpp KernelRelease.cpp KernelInfo.cpp - PersistenCacheConcurrentAccess.cpp + PersistentDeviceCodeCache.cpp ) diff --git a/sycl/unittests/kernel-and-program/PersistenCacheConcurrentAccess.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp similarity index 59% rename from sycl/unittests/kernel-and-program/PersistenCacheConcurrentAccess.cpp rename to sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index 0841addc03555..73605d648d2f6 100644 --- a/sycl/unittests/kernel-and-program/PersistenCacheConcurrentAccess.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -7,7 +7,7 @@ //===----------------------------------------------------------------------===// #include "../thread_safety/ThreadUtils.h" -#include "detail/persistent_cache.hpp" +#include "detail/persistent_device_code_cache.hpp" #include #include #include @@ -48,7 +48,7 @@ static pi_result redefinedProgramGetInfo(pi_program program, return PI_SUCCESS; } -class PersistenCacheConcurrentAccess : public ::testing::Test { +class PersistenDeviceCodeCache : public ::testing::Test { public: #ifdef _WIN32 int setenv(const char *name, const char *value, int overwrite) { @@ -63,7 +63,7 @@ class PersistenCacheConcurrentAccess : public ::testing::Test { } #endif - PersistenCacheConcurrentAccess() : Plt{default_selector()} { + PersistenDeviceCodeCache() : Plt{default_selector()} { const char *envTmp = #ifdef _WIN32 std::getenv("TEMP"); @@ -80,7 +80,6 @@ class PersistenCacheConcurrentAccess : public ::testing::Test { #endif cacheRoot += "/PersistenCache"; setenv("SYCL_CACHE_DIR", cacheRoot.c_str(), 0); - std::printf("Use %s as cache root\n", cacheRoot.c_str()); if (Plt.is_host() || Plt.get_backend() != backend::opencl) { std::clog << "This test is only supported on OpenCL devices\n"; @@ -107,24 +106,24 @@ class PersistenCacheConcurrentAccess : public ::testing::Test { detail::RTDeviceBinaryImage Img{Bin, ModuleHandle}; RT::PiProgram NativeProg; std::unique_ptr Mock; -}; -TEST_F(PersistenCacheConcurrentAccess, ReadWriteCacheItem) { std::vector> Data = { std::vector(1024, '1'), std::vector(1024 * 1024, '2'), std::vector(256, '3'), std::vector(1024 * 64, '4')}; +}; - constexpr std::size_t threadCount = 300; +TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheItem) { + constexpr std::size_t threadCount = 1000; Barrier b(threadCount); { auto testLambda = [&](std::size_t threadId) { b.wait(); - detail::PersistentCache::putPIProgramToDisc( - detail::getSyclObjImpl(Plt)->getPlugin(), Dev, Img, + detail::PersistentDeviceCodeCache::putItemToDisc( + Dev, Img, sycl::vector_class( {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't'}), "--build-options", NativeProg); - auto res = detail::PersistentCache::getPIProgramFromDisc( + auto res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, sycl::vector_class( {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't'}), @@ -140,4 +139,52 @@ TEST_F(PersistenCacheConcurrentAccess, ReadWriteCacheItem) { ThreadPool MPool(threadCount, testLambda); } } +TEST_F(PersistenDeviceCodeCache, AccessDeniedForCacheDir) { + std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( + Dev, Img, {}, "--build-options"); + detail::PersistentDeviceCodeCache::putItemToDisc( + Dev, Img, {}, "--build-options", NativeProg); + assert(std::filesystem::exists(ItemDir + "/0.bin") && "No file created"); + std::filesystem::permissions(ItemDir + "/0.bin", + std::filesystem::perms::owner_all | + std::filesystem::perms::group_all | + std::filesystem::perms::others_all, + std::filesystem::perm_options::remove); + // No access to binary file new cache item to be created + detail::PersistentDeviceCodeCache::putItemToDisc( + Dev, Img, {}, "--build-options", NativeProg); + assert(std::filesystem::exists(ItemDir + "/1.bin") && "No file created"); + + std::filesystem::permissions(ItemDir + "/1.src", + std::filesystem::perms::owner_all | + std::filesystem::perms::group_all | + std::filesystem::perms::others_all, + std::filesystem::perm_options::remove); + auto res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, "--build-options", NativeProg); + std::cout << res.size() << std::endl; + // No image to be read due to lack of permissions + assert(res.size() == 0); + + std::filesystem::permissions(ItemDir + "/0.bin", + std::filesystem::perms::owner_all | + std::filesystem::perms::group_all | + std::filesystem::perms::others_all, + std::filesystem::perm_options::add); + + std::filesystem::permissions(ItemDir + "/1.src", + std::filesystem::perms::owner_all | + std::filesystem::perms::group_all | + std::filesystem::perms::others_all, + std::filesystem::perm_options::add); + + res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, "--build-options", NativeProg); + // Image should be successfully read + for (int i = 0; i < res.size(); ++i) { + for (int j = 0; j < res[i].size(); ++j) { + assert(res[i][j] == i && "Corrupted image loaded from persistent cache"); + } + } +} } // namespace From 4f38f478dcd4edcd94f44efe15023745890cd103 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Thu, 1 Apr 2021 16:21:56 +0300 Subject: [PATCH 20/32] Updates - apply review remarks; - add synchronization for concurent FS access; - IO operations error management; - added more comments. --- sycl/source/detail/os_util.cpp | 6 +- .../detail/persistent_device_code_cache.cpp | 91 +++++++++- .../detail/persistent_device_code_cache.hpp | 43 ++++- .../PersistentDeviceCodeCache.cpp | 156 +++++++++++------- sycl/unittests/thread_safety/ThreadUtils.h | 4 + 5 files changed, 217 insertions(+), 83 deletions(-) diff --git a/sycl/source/detail/os_util.cpp b/sycl/source/detail/os_util.cpp index a95991ae3df01..c95866744fa19 100644 --- a/sycl/source/detail/os_util.cpp +++ b/sycl/source/detail/os_util.cpp @@ -8,7 +8,6 @@ #include #include -#include #include @@ -30,7 +29,6 @@ #elif defined(__SYCL_RT_OS_WINDOWS) #include -#include #include #include @@ -123,7 +121,7 @@ std::string OSUtil::getCurrentDSODir() { // // 4) Extract an absolute path to a filename and get a dirname from it. // - uintptr_t CurrentFunc = (uintptr_t)&getCurrentDSODir; + uintptr_t CurrentFunc = (uintptr_t) &getCurrentDSODir; std::ifstream Stream("/proc/self/maps"); Stream >> std::hex; while (!Stream.eof()) { @@ -168,7 +166,7 @@ std::string OSUtil::getCurrentDSODir() { return ""; } -std::string OSUtil::getDirName(const char *Path) { +std::string OSUtil::getDirName(const char* Path) { std::string Tmp(Path); // dirname(3) needs a writable C string: a null-terminator is written where a // path should split. diff --git a/sycl/source/detail/persistent_device_code_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp index 70a2db610c569..5b40e0b0a10fd 100644 --- a/sycl/source/detail/persistent_device_code_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -10,6 +10,12 @@ #include #include #include +#if defined(__SYCL_RT_OS_LINUX) +#include +#else +#include +#include +#endif __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { @@ -48,6 +54,15 @@ int makeDir(const char *Dir) { #endif } +LockCacheItem::LockCacheItem(const std::string &DirName) + : FileName(DirName + "/.lock") { + int fd; + while ((fd = open(FileName.c_str(), O_CREAT | O_EXCL, S_IWRITE)) == -1) { + std::this_thread::yield(); + } + close(fd); +} + /* Stores build program in persisten cache */ void PersistentDeviceCodeCache::putItemToDisc( @@ -137,7 +152,11 @@ std::vector> PersistentDeviceCodeCache::getItemFromDisc( while (isPathPresent(FileName + ".bin") || isPathPresent(FileName + ".src")) { if (isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString)) { - return readBinaryDataFromFile(FileName + ".bin"); + try { + return readBinaryDataFromFile(FileName + ".bin"); + } catch (...) { + // If read was unsuccessfull try the next item + } } FileName = Path + "/" + std::to_string(++i); } @@ -155,17 +174,27 @@ std::string PersistentDeviceCodeCache::getDeviceIDString(const device &Device) { /* Write built binary to persistent cache * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] + * Return on first unsuccessfull file operation */ void PersistentDeviceCodeCache::writeBinaryDataToFile( const std::string &FileName, const std::vector> &Data) { std::ofstream FileStream{FileName, std::ios::binary}; + if (FileStream.fail()) + return; size_t Size = Data.size(); FileStream.write((char *)&Size, sizeof(Size)); + if (FileStream.fail()) + return; + for (size_t i = 0; i < Data.size(); ++i) { Size = Data[i].size(); FileStream.write((char *)&Size, sizeof(Size)); + if (FileStream.fail()) + return; FileStream.write(Data[i].data(), Size); + if (FileStream.fail()) + return; } FileStream.close(); } @@ -176,13 +205,24 @@ void PersistentDeviceCodeCache::writeBinaryDataToFile( std::vector> PersistentDeviceCodeCache::readBinaryDataFromFile(const std::string &FileName) { std::ifstream FileStream{FileName, std::ios::binary}; + if (FileStream.fail()) + return {}; size_t ImgNum, ImgSize; FileStream.read((char *)&ImgNum, sizeof(ImgNum)); + if (FileStream.fail()) + return {}; + std::vector> Res(ImgNum); for (size_t i = 0; i < ImgNum; ++i) { FileStream.read((char *)&ImgSize, sizeof(ImgSize)); + if (FileStream.fail()) + return {}; + std::vector ImgData(ImgSize); FileStream.read(ImgData.data(), ImgSize); + if (FileStream.fail()) + return {}; + Res[i] = std::move(ImgData); } @@ -198,30 +238,53 @@ void PersistentDeviceCodeCache::writeSourceItem( const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { std::ofstream FileStream{FileName, std::ios::binary}; + if (FileStream.fail()) + return; + std::string DeviceString{getDeviceIDString(Device)}; size_t Size = DeviceString.size(); FileStream.write((char *)&Size, sizeof(Size)); + if (FileStream.fail()) + return; FileStream.write(DeviceString.data(), Size); + if (FileStream.fail()) + return; Size = BuildOptionsString.size(); FileStream.write((char *)&Size, sizeof(Size)); + if (FileStream.fail()) + return; FileStream.write(BuildOptionsString.data(), Size); + if (FileStream.fail()) + return; Size = SpecConsts.size(); FileStream.write((char *)&Size, sizeof(Size)); + if (FileStream.fail()) + return; FileStream.write((const char *)SpecConsts.data(), Size); + if (FileStream.fail()) + return; Size = Img.getSize(); FileStream.write((char *)&Size, sizeof(Size)); + if (FileStream.fail()) + return; FileStream.write((const char *)Img.getRawData().BinaryStart, Size); + if (FileStream.fail()) + return; FileStream.close(); } -/* Check that cache item key sources are equal to the current program +/* Check that cache item key sources are equal to the current program. + * If file read operations fail cache item is treated as not equal. */ bool PersistentDeviceCodeCache::isCacheItemSrcEqual( const std::string &FileName, const device &Device, const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { std::ifstream FileStream{FileName, std::ios::binary}; + if (FileStream.fail()) + return false; + std::string ImgString{(const char *)Img.getRawData().BinaryStart, Img.getSize()}; std::string DeviceString{getDeviceIDString(Device)}; @@ -229,30 +292,40 @@ bool PersistentDeviceCodeCache::isCacheItemSrcEqual( SpecConsts.size()}; size_t Size; - std::string res; - FileStream.read((char *)&Size, sizeof(Size)); - res.resize(Size); + if (FileStream.fail()) + return false; + + std::string res(Size, '\0'); FileStream.read(&res[0], Size); - if (DeviceString.compare(res)) + if (FileStream.fail() || DeviceString.compare(res)) return false; FileStream.read((char *)&Size, sizeof(Size)); + if (FileStream.fail()) + return false; + res.resize(Size); FileStream.read(&res[0], Size); - if (BuildOptionsString.compare(0, Size, res.data())) + if (FileStream.fail() || BuildOptionsString.compare(0, Size, res.data())) return false; FileStream.read((char *)&Size, sizeof(Size)); + if (FileStream.fail()) + return false; + res.resize(Size); FileStream.read(&res[0], Size); - if (SpecConstsString.compare(res)) + if (FileStream.fail() || SpecConstsString.compare(res)) return false; FileStream.read((char *)&Size, sizeof(Size)); + if (FileStream.fail()) + return false; + res.resize(Size); FileStream.read(&res[0], Size); - if (ImgString.compare(res)) + if (FileStream.fail() || ImgString.compare(res)) return false; FileStream.close(); diff --git a/sycl/source/detail/persistent_device_code_cache.hpp b/sycl/source/detail/persistent_device_code_cache.hpp index c978b5e9705a1..bad58236c436b 100644 --- a/sycl/source/detail/persistent_device_code_cache.hpp +++ b/sycl/source/detail/persistent_device_code_cache.hpp @@ -15,7 +15,6 @@ #include #include #include -#include #include __SYCL_INLINE_NAMESPACE(cl) { @@ -39,20 +38,48 @@ class LockCacheItem { const std::string FileName; public: - LockCacheItem(const std::string &DirName) : FileName(DirName + "/.lock") { - int fd; - while ((fd = open(FileName.c_str(), O_CREAT | O_EXCL, S_IWRITE)) == -1) { - std::this_thread::yield(); - } - close(fd); - } + LockCacheItem(const std::string &DirName); static bool isLocked(const std::string &DirName) { return isPathPresent(DirName + "/.lock"); } ~LockCacheItem() { std::remove(FileName.c_str()); } }; +/* End of temporary solution*/ class PersistentDeviceCodeCache { + /* The device code images are stored on file system using structure below: + * / + * / + * / + * / + * / + * .src + * .bin + * .lock + * - root directory storing cache files; + * - hash out of device information used to + * identify target device; + * - hash made out of device image used as + * input for the JIT compilation; - hash for + * specialization constants values; - hash for + * all build options; - sequential number of hash + * collisions. When hashes matches for the specific build but full values + * don't, new cache item is added with incremented value (enumeration started + * from 0). + * Two files per cache item are stored on disk: + * .src - contains full values for build parameters (device information, + * specialization constant values, build options, device image) + * which is used to resolve hash collisions and analysis of cached + * items. + * .bin - contains built device code. + * Also directory lock file is created when cache item is written. Lock item + * .lock - directory lock file. It is created when data is save to + * filesystem. On read operation the absence of file is checked + * but not created to avoid lock. + * All filesystem operations do not treated as SYCL errors and ignored: + * - on cache write operation cache item is not created; + * - on cache read operation it is treated as cache miss. + */ private: /* Write built binary to persistent cache * Format: numImages, 1stImageSize, Image[, NthImageSize, NthImage...] diff --git a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index 73605d648d2f6..2a362787a5979 100644 --- a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -1,4 +1,4 @@ -//==----- PersistenCacheConcurrentAccess.cpp --- Persistent cache tests ----==// +//==----- PersistenDeviceCodeCache.cpp --- Persistent cache tests ----------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. // See https://llvm.org/LICENSE.txt for license information. @@ -17,11 +17,23 @@ #include #include +namespace fs = std::filesystem; + namespace { constexpr auto sycl_read_write = cl::sycl::access::mode::read_write; using namespace cl::sycl; -constexpr size_t BinNum = 4; -constexpr size_t BinSizes[BinNum] = {1024, 1024 * 1024, 256, 1024 * 64}; + +/* Vector of programs which can be used for testing + */ +std::vector> Progs = { + {128}, /*tiny program for 1 target device, 128 B long*/ + {10240}, /*small program for 1 target device, 10 kB long*/ + {1024 * 1024, 1024, 256, 1024 * 64}, /*big program for 4 target + device, ~1 MB long*/ +}; + +static unsigned char DeviceCodeID = 2; + static pi_result redefinedProgramGetInfo(pi_program program, pi_program_info param_name, size_t param_value_size, @@ -29,19 +41,19 @@ static pi_result redefinedProgramGetInfo(pi_program program, size_t *param_value_size_ret) { if (param_name == PI_PROGRAM_INFO_NUM_DEVICES) { auto value = reinterpret_cast(param_value); - *value = BinNum; + *value = Progs[DeviceCodeID].size(); } if (param_name == PI_PROGRAM_INFO_BINARY_SIZES) { auto value = reinterpret_cast(param_value); - for (int i = 0; i < BinNum; ++i) - value[i] = BinSizes[i]; + for (int i = 0; i < Progs[DeviceCodeID].size(); ++i) + value[i] = Progs[DeviceCodeID][i]; } if (param_name == PI_PROGRAM_INFO_BINARIES) { auto value = reinterpret_cast(param_value); - for (int i = 0; i < BinNum; ++i) - for (int j = 0; j < BinSizes[i]; ++j) + for (int i = 0; i < Progs[DeviceCodeID].size(); ++i) + for (int j = 0; j < Progs[DeviceCodeID][i]; ++j) value[i][j] = i; } @@ -92,7 +104,38 @@ class PersistenDeviceCodeCache : public ::testing::Test { Dev = Plt.get_devices()[0]; Mock->redefine( redefinedProgramGetInfo); - std::filesystem::remove_all(cacheRoot); + } + + void ConcurentReadWriteCache(unsigned char ProgramID, size_t ThreadCount) { + DeviceCodeID = ProgramID; + std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( + Dev, Img, {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't', ProgramID}, "--build-options"); + + Barrier b(ThreadCount); + { + auto testLambda = [&](std::size_t threadId) { + b.wait(); + detail::PersistentDeviceCodeCache::putItemToDisc( + Dev, Img, + sycl::vector_class( + {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't', ProgramID}), + "--build-options", NativeProg); + auto res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, + sycl::vector_class( + {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't', ProgramID}), + "--build-options", NativeProg); + for (int i = 0; i < res.size(); ++i) { + for (int j = 0; j < res[i].size(); ++j) { + assert(res[i][j] == i && + "Corrupted image loaded from persistent cache"); + } + } + }; + + ThreadPool MPool(ThreadCount, testLambda); + } + fs::remove_all(ItemDir); } protected: @@ -106,77 +149,65 @@ class PersistenDeviceCodeCache : public ::testing::Test { detail::RTDeviceBinaryImage Img{Bin, ModuleHandle}; RT::PiProgram NativeProg; std::unique_ptr Mock; - std::vector> Data = { - std::vector(1024, '1'), std::vector(1024 * 1024, '2'), - std::vector(256, '3'), std::vector(1024 * 64, '4')}; }; -TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheItem) { - constexpr std::size_t threadCount = 1000; - - Barrier b(threadCount); - { - auto testLambda = [&](std::size_t threadId) { - b.wait(); - detail::PersistentDeviceCodeCache::putItemToDisc( - Dev, Img, - sycl::vector_class( - {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't'}), - "--build-options", NativeProg); - auto res = detail::PersistentDeviceCodeCache::getItemFromDisc( - Dev, Img, - sycl::vector_class( - {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't'}), - "--build-options", NativeProg); - for (int i = 0; i < res.size(); ++i) { - for (int j = 0; j < res[i].size(); ++j) { - assert(res[i][j] == i && - "Corrupted image loaded from persistent cache"); - } - } - }; +/* Do read/write for the same cache item to/from 2000 threads for small device + * code size. Make sure that there is no data corruption or crashes. + */ +TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteSmallItem) { + ConcurentReadWriteCache(0, 2000); +} - ThreadPool MPool(threadCount, testLambda); - } +/* Do read/write for the same cache item to/from 1000 threads for medium device + * code size. Make sure that there is no data corruption or crashes. + */ +TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheMediumItem) { + ConcurentReadWriteCache(1, 1000); } + +/* Do read/write for the same cache item to/from 200 threads from big device + * code size. Make sure that there is no data corruption or crashes. + */ +TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheBigItem) { + ConcurentReadWriteCache(2, 200); +} + +/* Checks cache behavior when filesystem read/write operations fail + */ TEST_F(PersistenDeviceCodeCache, AccessDeniedForCacheDir) { std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( Dev, Img, {}, "--build-options"); detail::PersistentDeviceCodeCache::putItemToDisc( Dev, Img, {}, "--build-options", NativeProg); - assert(std::filesystem::exists(ItemDir + "/0.bin") && "No file created"); - std::filesystem::permissions(ItemDir + "/0.bin", - std::filesystem::perms::owner_all | - std::filesystem::perms::group_all | - std::filesystem::perms::others_all, - std::filesystem::perm_options::remove); + assert(fs::exists(ItemDir + "/0.bin") && "No file created"); + fs::permissions(ItemDir + "/0.bin", + fs::perms::owner_all | fs::perms::group_all | + fs::perms::others_all, + fs::perm_options::remove); // No access to binary file new cache item to be created detail::PersistentDeviceCodeCache::putItemToDisc( Dev, Img, {}, "--build-options", NativeProg); - assert(std::filesystem::exists(ItemDir + "/1.bin") && "No file created"); + assert(fs::exists(ItemDir + "/1.bin") && "No file created"); - std::filesystem::permissions(ItemDir + "/1.src", - std::filesystem::perms::owner_all | - std::filesystem::perms::group_all | - std::filesystem::perms::others_all, - std::filesystem::perm_options::remove); + fs::permissions(ItemDir + "/1.src", + fs::perms::owner_all | fs::perms::group_all | + fs::perms::others_all, + fs::perm_options::remove); auto res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, {}, "--build-options", NativeProg); - std::cout << res.size() << std::endl; - // No image to be read due to lack of permissions + + // No image to be read due to lack of permissions fro source file assert(res.size() == 0); - std::filesystem::permissions(ItemDir + "/0.bin", - std::filesystem::perms::owner_all | - std::filesystem::perms::group_all | - std::filesystem::perms::others_all, - std::filesystem::perm_options::add); + fs::permissions(ItemDir + "/0.bin", + fs::perms::owner_all | fs::perms::group_all | + fs::perms::others_all, + fs::perm_options::add); - std::filesystem::permissions(ItemDir + "/1.src", - std::filesystem::perms::owner_all | - std::filesystem::perms::group_all | - std::filesystem::perms::others_all, - std::filesystem::perm_options::add); + fs::permissions(ItemDir + "/1.src", + fs::perms::owner_all | fs::perms::group_all | + fs::perms::others_all, + fs::perm_options::add); res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, {}, "--build-options", NativeProg); @@ -186,5 +217,6 @@ TEST_F(PersistenDeviceCodeCache, AccessDeniedForCacheDir) { assert(res[i][j] == i && "Corrupted image loaded from persistent cache"); } } + fs::remove_all(ItemDir); } } // namespace diff --git a/sycl/unittests/thread_safety/ThreadUtils.h b/sycl/unittests/thread_safety/ThreadUtils.h index 021f4087ab07f..5aeeb6826d665 100644 --- a/sycl/unittests/thread_safety/ThreadUtils.h +++ b/sycl/unittests/thread_safety/ThreadUtils.h @@ -2,6 +2,10 @@ #include #include + +/* Thread barrier which makes threads wait until defined number of threads reach + * the barrier. + */ class Barrier { public: Barrier() = delete; From 6946c8fcfd4555fff4fb86367c3ae7b040be6b74 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Thu, 1 Apr 2021 17:51:27 +0300 Subject: [PATCH 21/32] Fix CI issues --- .../PersistentDeviceCodeCache.cpp | 57 ++++++++----------- 1 file changed, 23 insertions(+), 34 deletions(-) diff --git a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index 2a362787a5979..54c731ce78306 100644 --- a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -11,14 +11,12 @@ #include #include #include -#include #include #include +#include #include #include -namespace fs = std::filesystem; - namespace { constexpr auto sycl_read_write = cl::sycl::access::mode::read_write; using namespace cl::sycl; @@ -26,10 +24,10 @@ using namespace cl::sycl; /* Vector of programs which can be used for testing */ std::vector> Progs = { - {128}, /*tiny program for 1 target device, 128 B long*/ + {128}, /*tiny program for 1 target device, 128 B long*/ {10240}, /*small program for 1 target device, 10 kB long*/ {1024 * 1024, 1024, 256, 1024 * 64}, /*big program for 4 target - device, ~1 MB long*/ + device, ~1 MB long*/ }; static unsigned char DeviceCodeID = 2; @@ -108,8 +106,9 @@ class PersistenDeviceCodeCache : public ::testing::Test { void ConcurentReadWriteCache(unsigned char ProgramID, size_t ThreadCount) { DeviceCodeID = ProgramID; - std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( - Dev, Img, {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't', ProgramID}, "--build-options"); + std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( + Dev, Img, {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't', ProgramID}, + "--build-options"); Barrier b(ThreadCount); { @@ -135,7 +134,7 @@ class PersistenDeviceCodeCache : public ::testing::Test { ThreadPool MPool(ThreadCount, testLambda); } - fs::remove_all(ItemDir); + llvm::sys::fs::remove_directories(ItemDir); } protected: @@ -151,27 +150,29 @@ class PersistenDeviceCodeCache : public ::testing::Test { std::unique_ptr Mock; }; -/* Do read/write for the same cache item to/from 2000 threads for small device +/* Do read/write for the same cache item to/from 300 threads for small device * code size. Make sure that there is no data corruption or crashes. */ TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteSmallItem) { - ConcurentReadWriteCache(0, 2000); + ConcurentReadWriteCache(0, 300); } -/* Do read/write for the same cache item to/from 1000 threads for medium device +/* Do read/write for the same cache item to/from 100 threads for medium device * code size. Make sure that there is no data corruption or crashes. */ TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheMediumItem) { - ConcurentReadWriteCache(1, 1000); + ConcurentReadWriteCache(1, 100); } -/* Do read/write for the same cache item to/from 200 threads from big device +/* Do read/write for the same cache item to/from 20 threads from big device * code size. Make sure that there is no data corruption or crashes. */ TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheBigItem) { - ConcurentReadWriteCache(2, 200); + ConcurentReadWriteCache(2, 20); } +#ifndef _WIN32 +// llvm::sys::fs::setPermissions doe not make effect on Windows /* Checks cache behavior when filesystem read/write operations fail */ TEST_F(PersistenDeviceCodeCache, AccessDeniedForCacheDir) { @@ -179,35 +180,22 @@ TEST_F(PersistenDeviceCodeCache, AccessDeniedForCacheDir) { Dev, Img, {}, "--build-options"); detail::PersistentDeviceCodeCache::putItemToDisc( Dev, Img, {}, "--build-options", NativeProg); - assert(fs::exists(ItemDir + "/0.bin") && "No file created"); - fs::permissions(ItemDir + "/0.bin", - fs::perms::owner_all | fs::perms::group_all | - fs::perms::others_all, - fs::perm_options::remove); + assert(llvm::sys::fs::exists(ItemDir + "/0.bin") && "No file created"); + llvm::sys::fs::setPermissions(ItemDir + "/0.bin", llvm::sys::fs::no_perms); // No access to binary file new cache item to be created detail::PersistentDeviceCodeCache::putItemToDisc( Dev, Img, {}, "--build-options", NativeProg); - assert(fs::exists(ItemDir + "/1.bin") && "No file created"); + assert(llvm::sys::fs::exists(ItemDir + "/1.bin") && "No file created"); - fs::permissions(ItemDir + "/1.src", - fs::perms::owner_all | fs::perms::group_all | - fs::perms::others_all, - fs::perm_options::remove); + llvm::sys::fs::setPermissions(ItemDir + "/1.bin", llvm::sys::fs::no_perms); auto res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, {}, "--build-options", NativeProg); // No image to be read due to lack of permissions fro source file assert(res.size() == 0); - fs::permissions(ItemDir + "/0.bin", - fs::perms::owner_all | fs::perms::group_all | - fs::perms::others_all, - fs::perm_options::add); - - fs::permissions(ItemDir + "/1.src", - fs::perms::owner_all | fs::perms::group_all | - fs::perms::others_all, - fs::perm_options::add); + llvm::sys::fs::setPermissions(ItemDir + "/0.bin", llvm::sys::fs::all_perms); + llvm::sys::fs::setPermissions(ItemDir + "/1.bin", llvm::sys::fs::all_perms); res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, {}, "--build-options", NativeProg); @@ -217,6 +205,7 @@ TEST_F(PersistenDeviceCodeCache, AccessDeniedForCacheDir) { assert(res[i][j] == i && "Corrupted image loaded from persistent cache"); } } - fs::remove_all(ItemDir); + llvm::sys::fs::remove_directories(ItemDir); } +#endif //_WIN32 } // namespace From d4f395e5e21bf9e7b82c912300f831d99f756626 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Thu, 1 Apr 2021 20:02:37 +0300 Subject: [PATCH 22/32] Fix comments --- sycl/source/detail/persistent_device_code_cache.cpp | 2 +- sycl/source/detail/persistent_device_code_cache.hpp | 2 +- 2 files changed, 2 insertions(+), 2 deletions(-) diff --git a/sycl/source/detail/persistent_device_code_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp index 5b40e0b0a10fd..1a6521dbbfcc2 100644 --- a/sycl/source/detail/persistent_device_code_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -1,4 +1,4 @@ -//==---------- persistent_cache.cpp - On-disk cache for program -*- C++-*---==// +//==---------- persistent_device_code_cache.cpp -----------------*- C++-*---==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. // See https://llvm.org/LICENSE.txt for license information. diff --git a/sycl/source/detail/persistent_device_code_cache.hpp b/sycl/source/detail/persistent_device_code_cache.hpp index bad58236c436b..e63f4e0866832 100644 --- a/sycl/source/detail/persistent_device_code_cache.hpp +++ b/sycl/source/detail/persistent_device_code_cache.hpp @@ -1,4 +1,4 @@ -//==---------- persistent_cache.hpp - On-disk cache for program -*- C++-*---==// +//==---------- persistent_device_code_cache.hpp -----------------*- C++-*---==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. // See https://llvm.org/LICENSE.txt for license information. From 5253e54f1c3de2c0fb9c295286d92e0d0f9df7e3 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Fri, 2 Apr 2021 18:25:01 +0300 Subject: [PATCH 23/32] Apply review comment and fix CUDA failure --- .../detail/persistent_device_code_cache.cpp | 5 ++-- .../detail/persistent_device_code_cache.hpp | 3 ++- .../PersistentDeviceCodeCache.cpp | 24 ++++++------------- sycl/unittests/thread_safety/ThreadUtils.h | 5 ++-- 4 files changed, 14 insertions(+), 23 deletions(-) diff --git a/sycl/source/detail/persistent_device_code_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp index 1a6521dbbfcc2..36c1389d94fb7 100644 --- a/sycl/source/detail/persistent_device_code_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -207,7 +207,7 @@ PersistentDeviceCodeCache::readBinaryDataFromFile(const std::string &FileName) { std::ifstream FileStream{FileName, std::ios::binary}; if (FileStream.fail()) return {}; - size_t ImgNum, ImgSize; + size_t ImgNum = 0, ImgSize = 0; FileStream.read((char *)&ImgNum, sizeof(ImgNum)); if (FileStream.fail()) return {}; @@ -242,7 +242,6 @@ void PersistentDeviceCodeCache::writeSourceItem( return; std::string DeviceString{getDeviceIDString(Device)}; - size_t Size = DeviceString.size(); FileStream.write((char *)&Size, sizeof(Size)); if (FileStream.fail()) @@ -291,7 +290,7 @@ bool PersistentDeviceCodeCache::isCacheItemSrcEqual( std::string SpecConstsString{(const char *)SpecConsts.data(), SpecConsts.size()}; - size_t Size; + size_t Size = 0; FileStream.read((char *)&Size, sizeof(Size)); if (FileStream.fail()) return false; diff --git a/sycl/source/detail/persistent_device_code_cache.hpp b/sycl/source/detail/persistent_device_code_cache.hpp index e63f4e0866832..22c4f85ace01a 100644 --- a/sycl/source/detail/persistent_device_code_cache.hpp +++ b/sycl/source/detail/persistent_device_code_cache.hpp @@ -76,7 +76,8 @@ class PersistentDeviceCodeCache { * .lock - directory lock file. It is created when data is save to * filesystem. On read operation the absence of file is checked * but not created to avoid lock. - * All filesystem operations do not treated as SYCL errors and ignored: + * All filesystem operations are not treated as SYCL errors and ignored. If + * such errors happen warning messages are written to std::err: * - on cache write operation cache item is not created; * - on cache read operation it is treated as cache miss. */ diff --git a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index 54c731ce78306..d77cef82e9239 100644 --- a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -74,22 +74,6 @@ class PersistenDeviceCodeCache : public ::testing::Test { #endif PersistenDeviceCodeCache() : Plt{default_selector()} { - const char *envTmp = -#ifdef _WIN32 - std::getenv("TEMP"); -#else - std::getenv("TMP"); -#endif - if (envTmp != nullptr) - cacheRoot += envTmp; - else -#ifdef _WIN32 - cacheRoot += "C:/temp"; -#else - cacheRoot += "/tmp"; -#endif - cacheRoot += "/PersistenCache"; - setenv("SYCL_CACHE_DIR", cacheRoot.c_str(), 0); if (Plt.is_host() || Plt.get_backend() != backend::opencl) { std::clog << "This test is only supported on OpenCL devices\n"; @@ -105,6 +89,10 @@ class PersistenDeviceCodeCache : public ::testing::Test { } void ConcurentReadWriteCache(unsigned char ProgramID, size_t ThreadCount) { + if (Plt.is_host() || Plt.get_backend() != backend::opencl) { + return; + } + DeviceCodeID = ProgramID; std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( Dev, Img, {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't', ProgramID}, @@ -138,7 +126,6 @@ class PersistenDeviceCodeCache : public ::testing::Test { } protected: - std::string cacheRoot; detail::OSModuleHandle ModuleHandle = detail::OSUtil::ExeModuleHandle; platform Plt; device Dev; @@ -176,6 +163,9 @@ TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheBigItem) { /* Checks cache behavior when filesystem read/write operations fail */ TEST_F(PersistenDeviceCodeCache, AccessDeniedForCacheDir) { + if (Plt.is_host() || Plt.get_backend() != backend::opencl) { + return; + } std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( Dev, Img, {}, "--build-options"); detail::PersistentDeviceCodeCache::putItemToDisc( diff --git a/sycl/unittests/thread_safety/ThreadUtils.h b/sycl/unittests/thread_safety/ThreadUtils.h index 5aeeb6826d665..dc7f75c749493 100644 --- a/sycl/unittests/thread_safety/ThreadUtils.h +++ b/sycl/unittests/thread_safety/ThreadUtils.h @@ -3,8 +3,9 @@ #include #include -/* Thread barrier which makes threads wait until defined number of threads reach - * the barrier. +/* Single use thread barrier which makes threads wait until defined number of + * threads reach it. + * std:barrier should be used instead once compiler is moved to C++20 standard. */ class Barrier { public: From 76fbc2cbcf4b94d19d64a260a108185d0f7718bb Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Mon, 5 Apr 2021 17:43:50 +0300 Subject: [PATCH 24/32] Remove duplicated and recursive functions --- .../detail/persistent_device_code_cache.cpp | 31 +++++++------------ 1 file changed, 12 insertions(+), 19 deletions(-) diff --git a/sycl/source/detail/persistent_device_code_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp index 36c1389d94fb7..a6cd9e09c2f31 100644 --- a/sycl/source/detail/persistent_device_code_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -23,35 +23,28 @@ namespace detail { /* This is temporary solution until std::filesystem is available when SYCL RT * is moved to c++17 standard*/ -std::string getDirName(const char *Path) { - std::string Tmp(Path); - // Remove trailing directory separators - Tmp.erase(Tmp.find_last_not_of("/\\") + 1, std::string::npos); - - auto pos = Tmp.find_last_of("/\\"); - if (pos != std::string::npos) - return Tmp.substr(0, pos); - - // If no directory separator is present return initial path like dirname does - return Tmp; -} +/* Create directory recursively */ int makeDir(const char *Dir) { assert((Dir != nullptr) && "Passed null-pointer as directory name."); - // Directory is present - do nothing if (isPathPresent(Dir)) return 0; - char *CurDir = strdup(Dir); - makeDir(getDirName(CurDir).c_str()); - - free(CurDir); + std::string Path{Dir}, CurPath; + size_t pos = 0; + do { + pos = Path.find_first_of("/\\", ++pos); + CurPath = Path.substr(0, pos); #if defined(__SYCL_RT_OS_LINUX) - return mkdir(Dir, 0777); + auto Res = mkdir(CurPath.c_str(), 0777); #else - return _mkdir(Dir); + auto Res = _mkdir(CurPath.c_str()); #endif + if (Res && errno != EEXIST) + return Res; + } while (pos != std::string::npos); + return 0; } LockCacheItem::LockCacheItem(const std::string &DirName) From 349920c76333ecaa2542fbeefb7e8fe1045af9c2 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Tue, 6 Apr 2021 11:19:16 +0300 Subject: [PATCH 25/32] Save code with obsoleting .lock files --- .../detail/persistent_device_code_cache.cpp | 60 +++++++-- .../detail/persistent_device_code_cache.hpp | 30 ++++- .../PersistentDeviceCodeCache.cpp | 118 ++++++++++++++---- 3 files changed, 171 insertions(+), 37 deletions(-) diff --git a/sycl/source/detail/persistent_device_code_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp index a6cd9e09c2f31..c6945f17ffbed 100644 --- a/sycl/source/detail/persistent_device_code_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -10,6 +10,7 @@ #include #include #include + #if defined(__SYCL_RT_OS_LINUX) #include #else @@ -47,13 +48,50 @@ int makeDir(const char *Dir) { return 0; } -LockCacheItem::LockCacheItem(const std::string &DirName) - : FileName(DirName + "/.lock") { +/// Checks if file age exceeds defined threshold +bool exceedLifeTime(const std::string &Path, time_t sec) { + struct stat Stat; + + if (stat(Path.c_str(), &Stat)) { + time_t CurTime; + time(&CurTime); + return (CurTime - Stat.st_mtime) > (sec * 1000); + } + return false; +} + +const char LockCacheItem::LockSuffix[] = ".lock"; +LockCacheItem::LockCacheItem(const std::string &Path) + : FileName(Path + LockSuffix) { int fd; + if (exceedLifeTime(FileName, 3600)) + std::remove(FileName.c_str()); + + auto Start = std::chrono::high_resolution_clock::now(); + while ((fd = open(FileName.c_str(), O_CREAT | O_EXCL, S_IWRITE)) == -1) { + // if lock file is not created unblock the thread + if (std::chrono::high_resolution_clock::now() - Start > + std::chrono::microseconds(100)) { + return; + } std::this_thread::yield(); } close(fd); + Owned = true; +} + +LockCacheItem::~LockCacheItem() { + if (Owned) { + auto Start = std::chrono::high_resolution_clock::now(); + while (std::remove(FileName.c_str())) { + // if lock file is not cleaned unblock the thread + if (std::chrono::high_resolution_clock::now() - Start > + std::chrono::microseconds(10)) + return; + std::this_thread::yield(); + } + } } /* Stores build program in persisten cache @@ -104,10 +142,12 @@ void PersistentDeviceCodeCache::putItemToDisc( try { makeDir(DirName.c_str()); - LockCacheItem Lock{DirName}; - writeBinaryDataToFile(FileName + ".bin", Result); - writeSourceItem(FileName + ".src", Device, Img, SpecConsts, - BuildOptionsString); + LockCacheItem Lock{FileName}; + if (Lock.isOwned()) { + writeBinaryDataToFile(FileName + ".bin", Result); + writeSourceItem(FileName + ".src", Device, Img, SpecConsts, + BuildOptionsString); + } } catch (...) { // If a problem happens on storing cache item, do nothing } @@ -137,13 +177,11 @@ std::vector> PersistentDeviceCodeCache::getItemFromDisc( int i = 0; - // If cache directory is locked ignore cache - if (LockCacheItem::isLocked(Path)) - return {}; - std::string FileName{Path + "/" + std::to_string(i)}; while (isPathPresent(FileName + ".bin") || isPathPresent(FileName + ".src")) { - if (isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, + + if (!LockCacheItem::isLocked(FileName) && + isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, BuildOptionsString)) { try { return readBinaryDataFromFile(FileName + ".bin"); diff --git a/sycl/source/detail/persistent_device_code_cache.hpp b/sycl/source/detail/persistent_device_code_cache.hpp index 22c4f85ace01a..51ad97503ed9b 100644 --- a/sycl/source/detail/persistent_device_code_cache.hpp +++ b/sycl/source/detail/persistent_device_code_cache.hpp @@ -12,9 +12,11 @@ #include #include #include +#include #include #include #include +#include #include __SYCL_INLINE_NAMESPACE(cl) { @@ -25,24 +27,32 @@ namespace detail { * is moved to c++17 standard*/ std::string getDirName(const char *Path); -#include /// Checks if specified path is present inline bool isPathPresent(const std::string &Path) { struct stat Stat; return !stat(Path.c_str(), &Stat); } +/// Checks if file age exceeds defined threshold +bool exceedLifeTime(const std::string &Path, time_t sec); + +/// Make directory recursibely int makeDir(const char *Dir); class LockCacheItem { +private: const std::string FileName; + bool Owned = false; + static const char LockSuffix[]; public: - LockCacheItem(const std::string &DirName); - static bool isLocked(const std::string &DirName) { - return isPathPresent(DirName + "/.lock"); + LockCacheItem(const std::string &Path); + + bool isOwned() { return Owned; } + static bool isLocked(const std::string &Path) { + return isPathPresent(Path + LockSuffix); } - ~LockCacheItem() { std::remove(FileName.c_str()); } + ~LockCacheItem(); }; /* End of temporary solution*/ @@ -80,6 +90,16 @@ class PersistentDeviceCodeCache { * such errors happen warning messages are written to std::err: * - on cache write operation cache item is not created; * - on cache read operation it is treated as cache miss. + * + * To avoid concurent write operations to the same cache item causing data + * corruption cache item dir is locked using .lock file. It is created on + * write operation and checked on read. + * - Lock is done per cache item. + * - Lock is not blocking. If lock fails cache item read/write is skipped + * and SYCL application flow resumes. There is time threshold for locking + * a cache item: 10 microseconds. + * - If lock file exists for 1 hour it is cleared on next access to resume + * caching for the particular cache item. */ private: /* Write built binary to persistent cache diff --git a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index d77cef82e9239..bb1b5b26860b0 100644 --- a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -14,6 +14,7 @@ #include #include #include +#include #include #include @@ -92,11 +93,13 @@ class PersistenDeviceCodeCache : public ::testing::Test { if (Plt.is_host() || Plt.get_backend() != backend::opencl) { return; } - + std::string BuildOptions{"--concurrent-access=" + + std::to_string(ThreadCount)}; DeviceCodeID = ProgramID; std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( Dev, Img, {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't', ProgramID}, - "--build-options"); + BuildOptions); + llvm::sys::fs::remove_directories(ItemDir); Barrier b(ThreadCount); { @@ -106,15 +109,15 @@ class PersistenDeviceCodeCache : public ::testing::Test { Dev, Img, sycl::vector_class( {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't', ProgramID}), - "--build-options", NativeProg); - auto res = detail::PersistentDeviceCodeCache::getItemFromDisc( + BuildOptions, NativeProg); + auto Res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, sycl::vector_class( {'S', 'p', 'e', 'c', 'C', 'o', 'n', 's', 't', ProgramID}), - "--build-options", NativeProg); - for (int i = 0; i < res.size(); ++i) { - for (int j = 0; j < res[i].size(); ++j) { - assert(res[i][j] == i && + BuildOptions, NativeProg); + for (int i = 0; i < Res.size(); ++i) { + for (int j = 0; j < Res[i].size(); ++j) { + assert(Res[i][j] == i && "Corrupted image loaded from persistent cache"); } } @@ -158,6 +161,77 @@ TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheBigItem) { ConcurentReadWriteCache(2, 20); } +// llvm::sys::fs::setPermissions doe not make effect on Windows +/* Checks cache behavior when filesystem read/write operations fail + */ +TEST_F(PersistenDeviceCodeCache, LockFile) { + std::chrono::time_point OldTime = + std::chrono::system_clock::now() - std::chrono::hours(2); + if (Plt.is_host() || Plt.get_backend() != backend::opencl) { + return; + } + std::string BuildOptions{"--obsolete-lock"}; + std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( + Dev, Img, {}, BuildOptions); + llvm::sys::fs::remove_directories(ItemDir); + detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, + NativeProg); + assert(llvm::sys::fs::exists(ItemDir + "/0.bin") && "No file created"); + std::string LockFile = ItemDir + "/0.lock"; + assert(!llvm::sys::fs::exists(LockFile) && "Cache item locked"); + + int FD = -1; + // Create lock file for cache item + assert(!llvm::sys::fs::openFileForWrite(LockFile, FD, + llvm::sys::fs::CD_CreateNew) && + "Failed to create lock file"); + llvm::sys::Process::SafelyCloseFileDescriptor(FD); + // Cache item is locked - ignore it + auto Res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, BuildOptions, NativeProg); + assert(Res.size() == 0 && "Locked item was read"); + + // Cache item is locked - new cache item to be created + detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, + NativeProg); + assert(llvm::sys::fs::exists(ItemDir + "/1.bin") && "No file created"); + + // Lock second cache item + assert(!llvm::sys::fs::openFileForWrite(ItemDir + "/1.lock", FD, + llvm::sys::fs::CD_CreateNew) && + "Failed to create lock file"); + llvm::sys::Process::SafelyCloseFileDescriptor(FD); + + assert(!llvm::sys::fs::openFileForWrite(LockFile, FD, + llvm::sys::fs::CD_OpenExisting) && + "Failed to open lock file"); + // Make cache Item obsolete (last access time more than theshold) + llvm::sys::fs::setLastAccessAndModificationTime(FD, OldTime, OldTime); + llvm::sys::Process::SafelyCloseFileDescriptor(FD); + + // Lock file is obsolete - clean lock + detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, + NativeProg); + assert(!llvm::sys::fs::exists(ItemDir + "/2.bin") && "File was created"); + + assert(!llvm::sys::fs::openFileForWrite(LockFile, FD, + llvm::sys::fs::CD_OpenExisting) && + "Failed to open lock file"); + // Make cache Item obsolete (last access time more than theshold) + llvm::sys::fs::setLastAccessAndModificationTime(FD, OldTime, OldTime); + llvm::sys::Process::SafelyCloseFileDescriptor(FD); + + Res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, BuildOptions, NativeProg); + // Image should be successfully read + for (int i = 0; i < Res.size(); ++i) { + for (int j = 0; j < Res[i].size(); ++j) { + assert(Res[i][j] == i && "Corrupted image loaded from persistent cache"); + } + } + llvm::sys::fs::remove_directories(ItemDir); +} + #ifndef _WIN32 // llvm::sys::fs::setPermissions doe not make effect on Windows /* Checks cache behavior when filesystem read/write operations fail @@ -166,33 +240,35 @@ TEST_F(PersistenDeviceCodeCache, AccessDeniedForCacheDir) { if (Plt.is_host() || Plt.get_backend() != backend::opencl) { return; } + std::string BuildOptions{"--build-options"}; std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( - Dev, Img, {}, "--build-options"); - detail::PersistentDeviceCodeCache::putItemToDisc( - Dev, Img, {}, "--build-options", NativeProg); + Dev, Img, {}, BuildOptions); + llvm::sys::fs::remove_directories(ItemDir); + detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, + NativeProg); assert(llvm::sys::fs::exists(ItemDir + "/0.bin") && "No file created"); llvm::sys::fs::setPermissions(ItemDir + "/0.bin", llvm::sys::fs::no_perms); // No access to binary file new cache item to be created - detail::PersistentDeviceCodeCache::putItemToDisc( - Dev, Img, {}, "--build-options", NativeProg); + detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, + NativeProg); assert(llvm::sys::fs::exists(ItemDir + "/1.bin") && "No file created"); llvm::sys::fs::setPermissions(ItemDir + "/1.bin", llvm::sys::fs::no_perms); - auto res = detail::PersistentDeviceCodeCache::getItemFromDisc( - Dev, Img, {}, "--build-options", NativeProg); + auto Res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, BuildOptions, NativeProg); // No image to be read due to lack of permissions fro source file - assert(res.size() == 0); + assert(Res.size() == 0); llvm::sys::fs::setPermissions(ItemDir + "/0.bin", llvm::sys::fs::all_perms); llvm::sys::fs::setPermissions(ItemDir + "/1.bin", llvm::sys::fs::all_perms); - res = detail::PersistentDeviceCodeCache::getItemFromDisc( - Dev, Img, {}, "--build-options", NativeProg); + Res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, BuildOptions, NativeProg); // Image should be successfully read - for (int i = 0; i < res.size(); ++i) { - for (int j = 0; j < res[i].size(); ++j) { - assert(res[i][j] == i && "Corrupted image loaded from persistent cache"); + for (int i = 0; i < Res.size(); ++i) { + for (int j = 0; j < Res[i].size(); ++j) { + assert(Res[i][j] == i && "Corrupted image loaded from persistent cache"); } } llvm::sys::fs::remove_directories(ItemDir); From cdda42e4a5f046724071f5d9942a39ebcbfd67da Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Tue, 6 Apr 2021 11:41:35 +0300 Subject: [PATCH 26/32] Implement non-nlocking lock files --- .../detail/persistent_device_code_cache.cpp | 15 ------ .../detail/persistent_device_code_cache.hpp | 49 +++++++------------ .../PersistentDeviceCodeCache.cpp | 32 ++---------- 3 files changed, 24 insertions(+), 72 deletions(-) diff --git a/sycl/source/detail/persistent_device_code_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp index c6945f17ffbed..8bf36e0aa99d9 100644 --- a/sycl/source/detail/persistent_device_code_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -48,25 +48,10 @@ int makeDir(const char *Dir) { return 0; } -/// Checks if file age exceeds defined threshold -bool exceedLifeTime(const std::string &Path, time_t sec) { - struct stat Stat; - - if (stat(Path.c_str(), &Stat)) { - time_t CurTime; - time(&CurTime); - return (CurTime - Stat.st_mtime) > (sec * 1000); - } - return false; -} - const char LockCacheItem::LockSuffix[] = ".lock"; LockCacheItem::LockCacheItem(const std::string &Path) : FileName(Path + LockSuffix) { int fd; - if (exceedLifeTime(FileName, 3600)) - std::remove(FileName.c_str()); - auto Start = std::chrono::high_resolution_clock::now(); while ((fd = open(FileName.c_str(), O_CREAT | O_EXCL, S_IWRITE)) == -1) { diff --git a/sycl/source/detail/persistent_device_code_cache.hpp b/sycl/source/detail/persistent_device_code_cache.hpp index 51ad97503ed9b..3db7cb50438db 100644 --- a/sycl/source/detail/persistent_device_code_cache.hpp +++ b/sycl/source/detail/persistent_device_code_cache.hpp @@ -33,9 +33,6 @@ inline bool isPathPresent(const std::string &Path) { return !stat(Path.c_str(), &Stat); } -/// Checks if file age exceeds defined threshold -bool exceedLifeTime(const std::string &Path, time_t sec); - /// Make directory recursibely int makeDir(const char *Dir); @@ -70,36 +67,28 @@ class PersistentDeviceCodeCache { * - hash out of device information used to * identify target device; * - hash made out of device image used as - * input for the JIT compilation; - hash for - * specialization constants values; - hash for - * all build options; - sequential number of hash - * collisions. When hashes matches for the specific build but full values - * don't, new cache item is added with incremented value (enumeration started - * from 0). + * input for the JIT compilation; + * - hash for specialization constants values; + * - hash for all build options; + * - sequential number of hash collisions. + * When hashes match for the specific build + * but full values don't, new cache item is + * added with incremented value(enumeration + * started from 0). * Two files per cache item are stored on disk: - * .src - contains full values for build parameters (device information, - * specialization constant values, build options, device image) - * which is used to resolve hash collisions and analysis of cached - * items. - * .bin - contains built device code. - * Also directory lock file is created when cache item is written. Lock item - * .lock - directory lock file. It is created when data is save to - * filesystem. On read operation the absence of file is checked - * but not created to avoid lock. - * All filesystem operations are not treated as SYCL errors and ignored. If - * such errors happen warning messages are written to std::err: + * .src - contains full values for build parameters (device information, + * specialization constant values, build options, device image) + * which is used to resolve hash collisions and analysis of + * cached items. + * .bin - contains built device code. + * .lock - cache item lock file. It is created when data is saved to + * filesystem. On read operation the absence of file is checked + * but it is not created to avoid lock. + * All filesystem operation failures are not treated as SYCL errors and + * ignored. If such errors happen warning messages are written to std::cerr + * and: * - on cache write operation cache item is not created; * - on cache read operation it is treated as cache miss. - * - * To avoid concurent write operations to the same cache item causing data - * corruption cache item dir is locked using .lock file. It is created on - * write operation and checked on read. - * - Lock is done per cache item. - * - Lock is not blocking. If lock fails cache item read/write is skipped - * and SYCL application flow resumes. There is time threshold for locking - * a cache item: 10 microseconds. - * - If lock file exists for 1 hour it is cleared on next access to resume - * caching for the particular cache item. */ private: /* Write built binary to persistent cache diff --git a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index bb1b5b26860b0..88d1bca461033 100644 --- a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -14,7 +14,6 @@ #include #include #include -#include #include #include @@ -182,10 +181,8 @@ TEST_F(PersistenDeviceCodeCache, LockFile) { int FD = -1; // Create lock file for cache item - assert(!llvm::sys::fs::openFileForWrite(LockFile, FD, - llvm::sys::fs::CD_CreateNew) && - "Failed to create lock file"); - llvm::sys::Process::SafelyCloseFileDescriptor(FD); + { std::ofstream File{LockFile}; } + // Cache item is locked - ignore it auto Res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, {}, BuildOptions, NativeProg); @@ -197,29 +194,10 @@ TEST_F(PersistenDeviceCodeCache, LockFile) { assert(llvm::sys::fs::exists(ItemDir + "/1.bin") && "No file created"); // Lock second cache item - assert(!llvm::sys::fs::openFileForWrite(ItemDir + "/1.lock", FD, - llvm::sys::fs::CD_CreateNew) && - "Failed to create lock file"); - llvm::sys::Process::SafelyCloseFileDescriptor(FD); - - assert(!llvm::sys::fs::openFileForWrite(LockFile, FD, - llvm::sys::fs::CD_OpenExisting) && - "Failed to open lock file"); - // Make cache Item obsolete (last access time more than theshold) - llvm::sys::fs::setLastAccessAndModificationTime(FD, OldTime, OldTime); - llvm::sys::Process::SafelyCloseFileDescriptor(FD); - - // Lock file is obsolete - clean lock - detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, - NativeProg); - assert(!llvm::sys::fs::exists(ItemDir + "/2.bin") && "File was created"); + { std::ofstream File{ItemDir + "/1.lock"}; } - assert(!llvm::sys::fs::openFileForWrite(LockFile, FD, - llvm::sys::fs::CD_OpenExisting) && - "Failed to open lock file"); - // Make cache Item obsolete (last access time more than theshold) - llvm::sys::fs::setLastAccessAndModificationTime(FD, OldTime, OldTime); - llvm::sys::Process::SafelyCloseFileDescriptor(FD); + // Remove lock file + std::remove(LockFile.c_str()); Res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, {}, BuildOptions, NativeProg); From 2c95c3573a9e6e0d428940c7d70cb8c5c699fcaf Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Tue, 6 Apr 2021 17:30:11 +0300 Subject: [PATCH 27/32] Fix comments --- .../kernel-and-program/PersistentDeviceCodeCache.cpp | 5 ++--- 1 file changed, 2 insertions(+), 3 deletions(-) diff --git a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index 88d1bca461033..06a32ff150ec1 100644 --- a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -160,8 +160,7 @@ TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheBigItem) { ConcurentReadWriteCache(2, 20); } -// llvm::sys::fs::setPermissions doe not make effect on Windows -/* Checks cache behavior when filesystem read/write operations fail +/* Checks that lock file affects cache operations as expected. */ TEST_F(PersistenDeviceCodeCache, LockFile) { std::chrono::time_point OldTime = @@ -211,7 +210,7 @@ TEST_F(PersistenDeviceCodeCache, LockFile) { } #ifndef _WIN32 -// llvm::sys::fs::setPermissions doe not make effect on Windows +// llvm::sys::fs::setPermissions does not make effect on Windows /* Checks cache behavior when filesystem read/write operations fail */ TEST_F(PersistenDeviceCodeCache, AccessDeniedForCacheDir) { From 8dad33661d8c7de7c1f0bf721d28207557bc2742 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Wed, 7 Apr 2021 18:02:50 +0300 Subject: [PATCH 28/32] Apply review remarks --- sycl/doc/EnvironmentVariables.md | 3 +- sycl/source/detail/config.def | 1 + .../detail/persistent_device_code_cache.cpp | 110 ++++++------------ .../detail/persistent_device_code_cache.hpp | 22 +++- .../PersistentDeviceCodeCache.cpp | 95 +++++++++++++-- 5 files changed, 142 insertions(+), 89 deletions(-) diff --git a/sycl/doc/EnvironmentVariables.md b/sycl/doc/EnvironmentVariables.md index e1b2f8041bec9..6eee0f492cdf3 100644 --- a/sycl/doc/EnvironmentVariables.md +++ b/sycl/doc/EnvironmentVariables.md @@ -36,7 +36,8 @@ subject to change. Do not rely on these variables in production code. | SYCL_ENABLE_PCI | Integer | When set to 1, enables obtaining the GPU PCI address when using the Level Zero backend. The default is 0. | | SYCL_HOST_UNIFIED_MEMORY | Integer | Enforce host unified memory support or lack of it for the execution graph builder. If set to 0, it is enforced as not supported by all devices. If set to 1, it is enforced as supported by all devices. | | SYCL_CACHE_DIR | Path | Path to persistent cache root directory. Default values are `%AppData%\libsycl_cache` for Windows and `$XDG_CACHE_HOME/libsycl_cache` on Linux, if XDG_CACHE_HOME is not set then `$HOME/.cache/libsycl_cache`. | -| SYCL_CACHE_DISABLE_PERSISTENT | Any(\*) | Switches persistent cache switch off. Default value is ON. | +| SYCL_CACHE_TRACE | Any(\*) | Enables printing messages to std::cerr when non-blocking failures happen (e.g. unable to access cache item file). Default values if off. | +|| SYCL_CACHE_DISABLE_PERSISTENT | Any(\*) | Switches persistent cache switch off. Default value is ON. | | SYCL_CACHE_EVICTION_DISABLE | Any(\*) | Switches cache eviction off. Default value is ON. | | SYCL_CACHE_MAX_SIZE | Positive integer | Cache eviction is triggered once total size of cached images exceeds the value in megabytes (default - 8 192 for 8 GB). Set to 0 to disable size-based cache eviction. | | SYCL_CACHE_THRESHOLD | Positive integer | Cache eviction threshold in days (default value is 7 for 1 week). Set to 0 for disabling time-based cache eviction. | diff --git a/sycl/source/detail/config.def b/sycl/source/detail/config.def index 91b9eba9e8d07..c28fea915f658 100644 --- a/sycl/source/detail/config.def +++ b/sycl/source/detail/config.def @@ -22,6 +22,7 @@ CONFIG(SYCL_PROGRAM_COMPILE_OPTIONS, 64, __SYCL_PROGRAM_COMPILE_OPTIONS) CONFIG(SYCL_HOST_UNIFIED_MEMORY, 1, __SYCL_HOST_UNIFIED_MEMORY) // 260 (Windows limit) - 12 (filename) - 84 (cache directory structure) CONFIG(SYCL_CACHE_DIR, 164, __SYCL_CACHE_DIR) +CONFIG(SYCL_CACHE_TRACE, 1, __SYCL_CACHE_TRACE) CONFIG(SYCL_CACHE_DISABLE_PERSISTENT, 1, __SYCL_CACHE_DISABLE_PERSISTENT) CONFIG(SYCL_CACHE_EVICTION_DISABLE, 1, __SYCL_CACHE_EVICTION_DISABLE) CONFIG(SYCL_CACHE_MAX_SIZE, 16, __SYCL_CACHE_MAX_SIZE) diff --git a/sycl/source/detail/persistent_device_code_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp index 8bf36e0aa99d9..7ed5103e065ce 100644 --- a/sycl/source/detail/persistent_device_code_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -6,7 +6,7 @@ // //===----------------------------------------------------------------------===// -#include +#include #include #include #include @@ -25,7 +25,7 @@ namespace detail { /* This is temporary solution until std::filesystem is available when SYCL RT * is moved to c++17 standard*/ -/* Create directory recursively */ +/* Create directory recursively and return non zero code on success*/ int makeDir(const char *Dir) { assert((Dir != nullptr) && "Passed null-pointer as directory name."); if (isPathPresent(Dir)) @@ -52,31 +52,20 @@ const char LockCacheItem::LockSuffix[] = ".lock"; LockCacheItem::LockCacheItem(const std::string &Path) : FileName(Path + LockSuffix) { int fd; - auto Start = std::chrono::high_resolution_clock::now(); - while ((fd = open(FileName.c_str(), O_CREAT | O_EXCL, S_IWRITE)) == -1) { - // if lock file is not created unblock the thread - if (std::chrono::high_resolution_clock::now() - Start > - std::chrono::microseconds(100)) { - return; - } - std::this_thread::yield(); + /* If the lock fail is not created */ + if ((fd = open(FileName.c_str(), O_CREAT | O_EXCL, S_IWRITE)) != -1) { + close(fd); + Owned = true; + } else { + PersistentDeviceCodeCache::trace("Failed to aquire lock file: " + FileName); } - close(fd); - Owned = true; } LockCacheItem::~LockCacheItem() { - if (Owned) { - auto Start = std::chrono::high_resolution_clock::now(); - while (std::remove(FileName.c_str())) { - // if lock file is not cleaned unblock the thread - if (std::chrono::high_resolution_clock::now() - Start > - std::chrono::microseconds(10)) - return; - std::this_thread::yield(); - } - } + if (Owned && std::remove(FileName.c_str())) + PersistentDeviceCodeCache::trace("Failed to release lock file: " + + FileName); } /* Stores build program in persisten cache @@ -195,24 +184,18 @@ std::string PersistentDeviceCodeCache::getDeviceIDString(const device &Device) { void PersistentDeviceCodeCache::writeBinaryDataToFile( const std::string &FileName, const std::vector> &Data) { std::ofstream FileStream{FileName, std::ios::binary}; - if (FileStream.fail()) - return; size_t Size = Data.size(); FileStream.write((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return; for (size_t i = 0; i < Data.size(); ++i) { Size = Data[i].size(); FileStream.write((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return; FileStream.write(Data[i].data(), Size); - if (FileStream.fail()) - return; } FileStream.close(); + if (FileStream.fail()) + trace("Failed to write binary file " + FileName); } /* Read built binary to persistent cache @@ -221,26 +204,24 @@ void PersistentDeviceCodeCache::writeBinaryDataToFile( std::vector> PersistentDeviceCodeCache::readBinaryDataFromFile(const std::string &FileName) { std::ifstream FileStream{FileName, std::ios::binary}; - if (FileStream.fail()) - return {}; size_t ImgNum = 0, ImgSize = 0; FileStream.read((char *)&ImgNum, sizeof(ImgNum)); - if (FileStream.fail()) - return {}; std::vector> Res(ImgNum); for (size_t i = 0; i < ImgNum; ++i) { FileStream.read((char *)&ImgSize, sizeof(ImgSize)); - if (FileStream.fail()) - return {}; std::vector ImgData(ImgSize); FileStream.read(ImgData.data(), ImgSize); - if (FileStream.fail()) - return {}; Res[i] = std::move(ImgData); } + FileStream.close(); + + if (FileStream.fail()) { + trace("Failed to read binary file from " + FileName); + return {}; + } return Res; } @@ -254,39 +235,28 @@ void PersistentDeviceCodeCache::writeSourceItem( const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { std::ofstream FileStream{FileName, std::ios::binary}; - if (FileStream.fail()) - return; std::string DeviceString{getDeviceIDString(Device)}; size_t Size = DeviceString.size(); FileStream.write((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return; FileStream.write(DeviceString.data(), Size); - if (FileStream.fail()) - return; + Size = BuildOptionsString.size(); FileStream.write((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return; FileStream.write(BuildOptionsString.data(), Size); - if (FileStream.fail()) - return; + Size = SpecConsts.size(); FileStream.write((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return; FileStream.write((const char *)SpecConsts.data(), Size); - if (FileStream.fail()) - return; + Size = Img.getSize(); FileStream.write((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return; FileStream.write((const char *)Img.getRawData().BinaryStart, Size); - if (FileStream.fail()) - return; FileStream.close(); + + if (FileStream.fail()) { + trace("Failed to write source file to " + FileName); + } } /* Check that cache item key sources are equal to the current program. @@ -297,53 +267,43 @@ bool PersistentDeviceCodeCache::isCacheItemSrcEqual( const RTDeviceBinaryImage &Img, const SerializedObj &SpecConsts, const std::string &BuildOptionsString) { std::ifstream FileStream{FileName, std::ios::binary}; - if (FileStream.fail()) - return false; std::string ImgString{(const char *)Img.getRawData().BinaryStart, Img.getSize()}; - std::string DeviceString{getDeviceIDString(Device)}; std::string SpecConstsString{(const char *)SpecConsts.data(), SpecConsts.size()}; size_t Size = 0; FileStream.read((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return false; - std::string res(Size, '\0'); FileStream.read(&res[0], Size); - if (FileStream.fail() || DeviceString.compare(res)) + if (getDeviceIDString(Device).compare(res)) return false; FileStream.read((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return false; - res.resize(Size); FileStream.read(&res[0], Size); - if (FileStream.fail() || BuildOptionsString.compare(0, Size, res.data())) + if (BuildOptionsString.compare(0, Size, res.data())) return false; FileStream.read((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return false; - res.resize(Size); FileStream.read(&res[0], Size); - if (FileStream.fail() || SpecConstsString.compare(res)) + if (SpecConstsString.compare(res)) return false; FileStream.read((char *)&Size, sizeof(Size)); - if (FileStream.fail()) - return false; - res.resize(Size); FileStream.read(&res[0], Size); - if (FileStream.fail() || ImgString.compare(res)) + if (ImgString.compare(res)) return false; FileStream.close(); + + if (FileStream.fail()) { + trace("Failed to read source file from " + FileName); + } + return true; } diff --git a/sycl/source/detail/persistent_device_code_cache.hpp b/sycl/source/detail/persistent_device_code_cache.hpp index 3db7cb50438db..f178966aec6ec 100644 --- a/sycl/source/detail/persistent_device_code_cache.hpp +++ b/sycl/source/detail/persistent_device_code_cache.hpp @@ -12,7 +12,7 @@ #include #include #include -#include +#include #include #include #include @@ -33,9 +33,20 @@ inline bool isPathPresent(const std::string &Path) { return !stat(Path.c_str(), &Stat); } -/// Make directory recursibely +/// Make directory recursively and returns zero code on success int makeDir(const char *Dir); +/* The class manages inter-process synchronization: + * - Path passed to the constructor is appended with .lock and used as lock + * file. + * - All operations are not blocking and failure ignoring (diagnostic may be + * send to std::cerr when SYCL_CHACE_TRACE environment variable is set). + * - There are two modes of accessing shared resource: + * - write access assumes that lock is aquired (object is created and + * isOwned() method confirms that current executor owns the lock); + * - read access checks that the lock is not aquired for write by others + * with the help of isLocked() method. + */ class LockCacheItem { private: const std::string FileName; @@ -138,6 +149,13 @@ class PersistentDeviceCodeCache { const SerializedObj &SpecConsts, const std::string &BuildOptionsString); + /* Sends message to std:cerr stream when SYCL_CACHE_TRACE environemnt is set*/ + static void trace(const std::string &msg) { + static const char *TraceEnabled = SYCLConfig::get(); + if (TraceEnabled) + std::cerr << msg << std::endl; + } + /* Program binaries built for one or more devices are read from persistent * cache and returned in form of vector of programs. Each binary program is * stored in vector of chars. diff --git a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index 06a32ff150ec1..a603008547b75 100644 --- a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -11,10 +11,10 @@ #include #include #include +#include #include #include #include -#include #include namespace { @@ -88,6 +88,11 @@ class PersistenDeviceCodeCache : public ::testing::Test { redefinedProgramGetInfo); } + /* Helper function for concurent cache item read/write from diffrent number + * of threads with diffrent cache item sizes: + * ProgramID - defines program parameters to be used for testing (see Progs + * vector above. + * ThreadCount - number of parallel executors used for the test*/ void ConcurentReadWriteCache(unsigned char ProgramID, size_t ThreadCount) { if (Plt.is_host() || Plt.get_backend() != backend::opencl) { return; @@ -160,11 +165,77 @@ TEST_F(PersistenDeviceCodeCache, ConcurentReadWriteCacheBigItem) { ConcurentReadWriteCache(2, 20); } -/* Checks that lock file affects cache operations as expected. +/* Checks that no crash happens when cache items are corrupted on cache read. + * The case when source or binary files are corrupted is treated as cache miss. + * - only source file is present; + * - only binary file is present; + * - source file is corrupted; + * - binary file is corrupted. + */ +TEST_F(PersistenDeviceCodeCache, CorruptedCacheFiles) { + if (Plt.is_host() || Plt.get_backend() != backend::opencl) { + return; + } + std::string BuildOptions{"--corrupted-file"}; + std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( + Dev, Img, {}, BuildOptions); + llvm::sys::fs::remove_directories(ItemDir); + + // Only source file is present + detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, + NativeProg); + assert(!llvm::sys::fs::remove(ItemDir + "/0.bin") && + "Failed to remove binary file"); + auto Res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, BuildOptions, NativeProg); + assert(Res.size() == 0 && "Item with missed binary file was read"); + llvm::sys::fs::remove_directories(ItemDir); + + // Only binary file is present + detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, + NativeProg); + assert(!llvm::sys::fs::remove(ItemDir + "/0.src") && + "Failed to remove source file"); + Res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, BuildOptions, NativeProg); + assert(Res.size() == 0 && "Item with missed source file was read"); + llvm::sys::fs::remove_directories(ItemDir); + + // Binary file is corrupted + detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, + NativeProg); + std::ofstream FileStream(ItemDir + "/0.bin", + std::ofstream::out | std::ofstream::trunc); + /* Emulate binary built for 2 devices: first is OK, second is trancated + * from 23 bytes to 4 + */ + FileStream << 2 << 12 << "123456789012" << 23 << "1234"; + FileStream.close(); + assert((!FileStream.fail()) && "Failed to create trancated binary file"); + Res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, BuildOptions, NativeProg); + assert(Res.size() == 0 && "Item with corrupted binary file was read"); + + llvm::sys::fs::remove_directories(ItemDir); + + // Source file is empty + detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, + NativeProg); + { + std::ofstream FileStream(ItemDir + "/0.src", + std::ofstream::out | std::ofstream::trunc); + } + Res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, BuildOptions, NativeProg); + assert(Res.size() == 0 && "Item with corrupted binary file was read"); + llvm::sys::fs::remove_directories(ItemDir); +} + +/* Checks that lock file affects cache operations as expected: + * - new cache item is created if existing one is locked on write operation; + * - cache miss happens on read operation. */ TEST_F(PersistenDeviceCodeCache, LockFile) { - std::chrono::time_point OldTime = - std::chrono::system_clock::now() - std::chrono::hours(2); if (Plt.is_host() || Plt.get_backend() != backend::opencl) { return; } @@ -172,17 +243,18 @@ TEST_F(PersistenDeviceCodeCache, LockFile) { std::string ItemDir = detail::PersistentDeviceCodeCache::getCacheItemPath( Dev, Img, {}, BuildOptions); llvm::sys::fs::remove_directories(ItemDir); + + // Create 1st cahe item detail::PersistentDeviceCodeCache::putItemToDisc(Dev, Img, {}, BuildOptions, NativeProg); assert(llvm::sys::fs::exists(ItemDir + "/0.bin") && "No file created"); std::string LockFile = ItemDir + "/0.lock"; assert(!llvm::sys::fs::exists(LockFile) && "Cache item locked"); - int FD = -1; - // Create lock file for cache item + // Create lock file for the 1st cache item { std::ofstream File{LockFile}; } - // Cache item is locked - ignore it + // Cache item is locked, cache miss happens on read auto Res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, {}, BuildOptions, NativeProg); assert(Res.size() == 0 && "Locked item was read"); @@ -192,15 +264,16 @@ TEST_F(PersistenDeviceCodeCache, LockFile) { NativeProg); assert(llvm::sys::fs::exists(ItemDir + "/1.bin") && "No file created"); - // Lock second cache item + // Second cache item is locked, cache miss happens on read { std::ofstream File{ItemDir + "/1.lock"}; } + Res = detail::PersistentDeviceCodeCache::getItemFromDisc( + Dev, Img, {}, BuildOptions, NativeProg); + assert(Res.size() == 0 && "Locked item was read"); - // Remove lock file + // First cache item was anlocked and successfully read std::remove(LockFile.c_str()); - Res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, {}, BuildOptions, NativeProg); - // Image should be successfully read for (int i = 0; i < Res.size(); ++i) { for (int j = 0; j < Res[i].size(); ++j) { assert(Res[i][j] == i && "Corrupted image loaded from persistent cache"); From 9775e509b85198b82a6c4c61df49683fdf182f9c Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Wed, 7 Apr 2021 22:01:17 +0300 Subject: [PATCH 29/32] Added image sized thresholds --- sycl/doc/EnvironmentVariables.md | 4 +- .../detail/persistent_device_code_cache.cpp | 35 +++-- .../detail/persistent_device_code_cache.hpp | 40 +++++- sycl/test/on-device/cache/cache_env_vars.cpp | 129 ++++++++++++++++++ .../PersistentDeviceCodeCache.cpp | 3 +- 5 files changed, 191 insertions(+), 20 deletions(-) create mode 100644 sycl/test/on-device/cache/cache_env_vars.cpp diff --git a/sycl/doc/EnvironmentVariables.md b/sycl/doc/EnvironmentVariables.md index 6eee0f492cdf3..af357699e82fc 100644 --- a/sycl/doc/EnvironmentVariables.md +++ b/sycl/doc/EnvironmentVariables.md @@ -41,8 +41,8 @@ subject to change. Do not rely on these variables in production code. | SYCL_CACHE_EVICTION_DISABLE | Any(\*) | Switches cache eviction off. Default value is ON. | | SYCL_CACHE_MAX_SIZE | Positive integer | Cache eviction is triggered once total size of cached images exceeds the value in megabytes (default - 8 192 for 8 GB). Set to 0 to disable size-based cache eviction. | | SYCL_CACHE_THRESHOLD | Positive integer | Cache eviction threshold in days (default value is 7 for 1 week). Set to 0 for disabling time-based cache eviction. | -| SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE | Positive integer | Minimum size of device code image in kilobytes which is reasonable to cache on disk because disk access operation may take more time than do JIT compilation for it. Default value is 0 to cache all images. | -| SYCL_CACHE_MAX_DEVICE_IMAGE_SIZE | Positive integer | Maximum size of device image in megabytes which is cached. Too big kernels may overload disk too fast. Default value is 0 to cache all images. | +| SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE | Positive integer | Minimum size of device code image in bytes which is reasonable to cache on disk because disk access operation may take more time than do JIT compilation for it. Default value is 0 to cache all images. | +| SYCL_CACHE_MAX_DEVICE_IMAGE_SIZE | Positive integer | Maximum size of device image in bytes which is cached. Too big kernels may overload disk too fast. Default value is 1 GB. | `(*) Note: Any means this environment variable is effective when set to any non-null value.` diff --git a/sycl/source/detail/persistent_device_code_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp index 7ed5103e065ce..bedffd28ccdf3 100644 --- a/sycl/source/detail/persistent_device_code_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -48,7 +48,9 @@ int makeDir(const char *Dir) { return 0; } +/* Lock file suffix */ const char LockCacheItem::LockSuffix[] = ".lock"; + LockCacheItem::LockCacheItem(const std::string &Path) : FileName(Path + LockSuffix) { int fd; @@ -68,6 +70,27 @@ LockCacheItem::~LockCacheItem() { FileName); } +/* Returns true if specified image should be cached on disk. It checks if + * cache is enabled, image has SPIRV type and matches thresholds. */ +bool PersistentDeviceCodeCache::isImageCached(const RTDeviceBinaryImage &Img) { + // Cache shoould be enabled and image type should be SPIR-V + if (!isEnabled() || Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV) + return false; + + static auto MaxImgSize = getNumParam( + DEFAULT_MAX_DEVICE_IMAGE_SIZE); + static auto MinImgSize = getNumParam( + DEFAULT_MIN_DEVICE_IMAGE_SIZE); + + // Make sure that image size is between caching thresholds if they are set. + // Zero values for threshold is treated as disabled threshold. + if ((MaxImgSize && (Img.getSize() > MaxImgSize)) || + (MinImgSize && (Img.getSize() < MinImgSize))) + return false; + + return true; +} + /* Stores build program in persisten cache */ void PersistentDeviceCodeCache::putItemToDisc( @@ -75,11 +98,7 @@ void PersistentDeviceCodeCache::putItemToDisc( const SerializedObj &SpecConsts, const std::string &BuildOptionsString, const RT::PiProgram &NativePrg) { - if (!isEnabled()) - return; - - // Only SPIRV images are cached - if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV) + if (!isImageCached(Img)) return; auto Plugin = detail::getSyclObjImpl(Device)->getPlugin(); @@ -136,11 +155,7 @@ std::vector> PersistentDeviceCodeCache::getItemFromDisc( const SerializedObj &SpecConsts, const std::string &BuildOptionsString, RT::PiProgram &NativePrg) { - if (!isEnabled()) - return {}; - - // Only SPIRV images are cached - if (Img.getFormat() != PI_DEVICE_BINARY_TYPE_SPIRV) + if (!isImageCached(Img)) return {}; std::string Path = diff --git a/sycl/source/detail/persistent_device_code_cache.hpp b/sycl/source/detail/persistent_device_code_cache.hpp index f178966aec6ec..ee62332383c3c 100644 --- a/sycl/source/detail/persistent_device_code_cache.hpp +++ b/sycl/source/detail/persistent_device_code_cache.hpp @@ -141,6 +141,32 @@ class PersistentDeviceCodeCache { /* Form string representing device version */ static std::string getDeviceIDString(const device &Device); + /* Returns true if specified image should be cached on disk. It checks if + * cache is enabled, image has SPIRV type and matches thresholds. */ + static bool isImageCached(const RTDeviceBinaryImage &Img); + + /* Returns value of specified parameter. Default value is used if failure + * happens during obtaining value. */ + template + static unsigned long getNumParam(unsigned long Default) { + auto Value = SYCLConfig::get(); + try { + if (Value) + return std::stol(Value); + } catch (std::exception const &) { + PersistentDeviceCodeCache::trace("Invalid value provided, use default " + + std::to_string(Default)); + } + return Default; + } + + /* Default value for minimum device code size to be cached on disk in bytes */ + static constexpr unsigned long DEFAULT_MIN_DEVICE_IMAGE_SIZE = 0; + + /* Default value for maximum device code size to be cached on disk in bytes */ + static constexpr unsigned long DEFAULT_MAX_DEVICE_IMAGE_SIZE = + 1024 * 1024 * 1024; + public: /* Get directory name for storing current cache item */ @@ -149,13 +175,6 @@ class PersistentDeviceCodeCache { const SerializedObj &SpecConsts, const std::string &BuildOptionsString); - /* Sends message to std:cerr stream when SYCL_CACHE_TRACE environemnt is set*/ - static void trace(const std::string &msg) { - static const char *TraceEnabled = SYCLConfig::get(); - if (TraceEnabled) - std::cerr << msg << std::endl; - } - /* Program binaries built for one or more devices are read from persistent * cache and returned in form of vector of programs. Each binary program is * stored in vector of chars. @@ -173,6 +192,13 @@ class PersistentDeviceCodeCache { const SerializedObj &SpecConsts, const std::string &BuildOptionsString, const RT::PiProgram &NativePrg); + + /* Sends message to std:cerr stream when SYCL_CACHE_TRACE environemnt is set*/ + static void trace(const std::string &msg) { + static const char *TraceEnabled = SYCLConfig::get(); + if (TraceEnabled) + std::cerr << msg << std::endl; + } }; } // namespace detail } // namespace sycl diff --git a/sycl/test/on-device/cache/cache_env_vars.cpp b/sycl/test/on-device/cache/cache_env_vars.cpp new file mode 100644 index 0000000000000..0d27fa9c9b596 --- /dev/null +++ b/sycl/test/on-device/cache/cache_env_vars.cpp @@ -0,0 +1,129 @@ +// No JITing for host devices. +// REQUIRES: opencl || level_zero || cuda +// RUN: rm -rf %t/cache_dir +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -DTARGET_IMAGE=INC100 +// Build program and add item to cache +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %t.out | FileCheck %s --check-prefixes=CHECK-BUILD +// Ignore caching because image size is less than threshold +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 SYCL_CACHE_MIN_DEVICE_IMAGE_SIZE=100000 %t.out | FileCheck %s --check-prefixes=CHECK-BUILD +// Ignore caching because image size is more than threshold +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 SYCL_CACHE_MAX_DEVICE_IMAGE_SIZE=1000 %t.out | FileCheck %s --check-prefixes=CHECK-BUILD +// Use cache +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %t.out | FileCheck %s --check-prefixes=CHECK-CACHE +// Ignore cache because of environment variable +// RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 SYCL_CACHE_DISABLE_PERSISTENT=1 %t.out | FileCheck %s --check-prefixes=CHECK-BUILD +// +// The test checks environment variables which may disable caching. +// Also it can be used for benchmarking cache: +// Rough data collected on 32 core machine. +// Number of lines 1 10 100 1000 10000 +// Image Size(kB) 2 2 20 165 1700 +// Device code build time in seconds +// CPU OCL JIT 0.12 0.12 0.16 1.1 16 +// CPU OCL Cache 0.01 0.01 0.01 0.02 0.08 + +// CHECK-BUILD: piProgramBuild +// CHECK-BUILD-NOT: piProgramCreateWithBinary + +// CHECK-CACHE-NOT: piProgramBuild +// CHECK-CACHE: piProgramCreateWithBinary + +#define INC1(x) ((x) = (x) + 1); + +#define INC10(x) \ + INC1(x) \ + INC1(x) \ + INC1(x) \ + INC1(x) \ + INC1(x) \ + INC1(x) \ + INC1(x) \ + INC1(x) \ + INC1(x) \ + INC1(x) + +#define INC100(x) \ + INC10(x) \ + INC10(x) \ + INC10(x) \ + INC10(x) \ + INC10(x) \ + INC10(x) \ + INC10(x) \ + INC10(x) \ + INC10(x) \ + INC10(x) + +#define INC1000(x) \ + INC100(x) \ + INC100(x) \ + INC100(x) \ + INC100(x) \ + INC100(x) \ + INC100(x) \ + INC100(x) \ + INC100(x) \ + INC100(x) \ + INC100(x) + +#define INC10000(x) \ + INC1000(x) \ + INC1000(x) \ + INC1000(x) \ + INC1000(x) \ + INC1000(x) \ + INC1000(x) \ + INC1000(x) \ + INC1000(x) \ + INC1000(x) \ + INC1000(x) + +#define INC100000(x) \ + INC10000(x) \ + INC10000(x) \ + INC10000(x) \ + INC10000(x) \ + INC10000(x) \ + INC10000(x) \ + INC10000(x) \ + INC10000(x) \ + INC10000(x) \ + INC10000(x) + +#include +#include +#include +class Inc; +template void check_build_time(cl::sycl::queue &q) { + cl::sycl::program program(q.get_context()); + auto start = std::chrono::steady_clock::now(); + program.build_with_kernel_type(); + auto end = std::chrono::steady_clock::now(); + + std::chrono::duration elapsed_seconds = end - start; + std::cout << "elapsed build time: " << elapsed_seconds.count() << "s\n"; +} +int main(int argc, char **argv) { + auto start = std::chrono::steady_clock::now(); + // Test program and kernel APIs when building a kernel. + { + cl::sycl::queue q; + check_build_time(q); + + int data = 0; + { + cl::sycl::buffer buf(&data, cl::sycl::range<1>(1)); + cl::sycl::range<1> NumOfWorkItems{buf.get_count()}; + + q.submit([&](cl::sycl::handler &cgh) { + auto acc = buf.get_access(cgh); + cgh.parallel_for( + NumOfWorkItems, [=](cl::sycl::id<1> WIid) { TARGET_IMAGE(acc[0]) }); + }); + } + // check_build_time(q); + auto end = std::chrono::steady_clock::now(); + std::chrono::duration elapsed_seconds = end - start; + std::cout << "elapsed kernel time: " << elapsed_seconds.count() << "s\n"; + } +} diff --git a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index a603008547b75..5d3fe7acf0dee 100644 --- a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -5,7 +5,8 @@ // SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception // //===----------------------------------------------------------------------===// - +// This file contains tests covering persistena device code cache functionality. +// Detailed description of the tests cases can be seen per test function. #include "../thread_safety/ThreadUtils.h" #include "detail/persistent_device_code_cache.hpp" #include From b0c6a6dd28937620d100aad7972e7833f7301750 Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Thu, 8 Apr 2021 10:21:19 +0300 Subject: [PATCH 30/32] Aply review comments --- sycl/include/CL/sycl/detail/os_util.hpp | 16 ++++++++ sycl/source/detail/os_util.cpp | 28 ++++++++++++++ .../detail/persistent_device_code_cache.cpp | 37 +++---------------- .../detail/persistent_device_code_cache.hpp | 20 +++------- .../program_manager/program_manager.cpp | 2 +- sycl/test/abi/sycl_symbols_linux.dump | 1 + sycl/test/on-device/cache/basic.cpp | 1 - .../PersistentDeviceCodeCache.cpp | 2 +- 8 files changed, 59 insertions(+), 48 deletions(-) diff --git a/sycl/include/CL/sycl/detail/os_util.hpp b/sycl/include/CL/sycl/detail/os_util.hpp index 271943484f85b..55b41b28b640c 100644 --- a/sycl/include/CL/sycl/detail/os_util.hpp +++ b/sycl/include/CL/sycl/detail/os_util.hpp @@ -16,6 +16,8 @@ #include #include #include +#include +#include #ifdef _WIN32 #define __SYCL_RT_OS_WINDOWS @@ -80,6 +82,20 @@ class __SYCL_EXPORT OSUtil { /// Deallocates the memory referenced by \p Ptr. static void alignedFree(void *Ptr); + + /// Make directory recursively and returns zero code on success + static int makeDir(const char *Dir); + + /// Checks if specified path is present + static inline bool isPathPresent(const std::string &Path) { +#ifdef __SYCL_RT_OS_WINDOWS + struct _stat Stat; + return !_stat(Path.c_str(), &Stat); +#else + struct stat Stat; + return !stat(Path.c_str(), &Stat); +#endif + } }; } // namespace detail diff --git a/sycl/source/detail/os_util.cpp b/sycl/source/detail/os_util.cpp index c95866744fa19..9fa163825aa41 100644 --- a/sycl/source/detail/os_util.cpp +++ b/sycl/source/detail/os_util.cpp @@ -24,11 +24,13 @@ #include // for dirname #include #include // for PATH_MAX +#include #include #elif defined(__SYCL_RT_OS_WINDOWS) #include +#include #include #include @@ -271,6 +273,32 @@ void OSUtil::alignedFree(void *Ptr) { #endif } +/* This is temporary solution until std::filesystem is available when SYCL RT + * is moved to c++17 standard*/ + +/* Create directory recursively and return non zero code on success*/ +int OSUtil::makeDir(const char *Dir) { + assert((Dir != nullptr) && "Passed null-pointer as directory name."); + if (isPathPresent(Dir)) + return 0; + + std::string Path{Dir}, CurPath; + size_t pos = 0; + + do { + pos = Path.find_first_of("/\\", ++pos); + CurPath = Path.substr(0, pos); +#if defined(__SYCL_RT_OS_LINUX) + auto Res = mkdir(CurPath.c_str(), 0777); +#else + auto Res = _mkdir(CurPath.c_str()); +#endif + if (Res && errno != EEXIST) + return Res; + } while (pos != std::string::npos); + return 0; +} + } // namespace detail } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/detail/persistent_device_code_cache.cpp b/sycl/source/detail/persistent_device_code_cache.cpp index bedffd28ccdf3..49b7530c79dcb 100644 --- a/sycl/source/detail/persistent_device_code_cache.cpp +++ b/sycl/source/detail/persistent_device_code_cache.cpp @@ -22,32 +22,6 @@ __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { namespace detail { -/* This is temporary solution until std::filesystem is available when SYCL RT - * is moved to c++17 standard*/ - -/* Create directory recursively and return non zero code on success*/ -int makeDir(const char *Dir) { - assert((Dir != nullptr) && "Passed null-pointer as directory name."); - if (isPathPresent(Dir)) - return 0; - - std::string Path{Dir}, CurPath; - size_t pos = 0; - - do { - pos = Path.find_first_of("/\\", ++pos); - CurPath = Path.substr(0, pos); -#if defined(__SYCL_RT_OS_LINUX) - auto Res = mkdir(CurPath.c_str(), 0777); -#else - auto Res = _mkdir(CurPath.c_str()); -#endif - if (Res && errno != EEXIST) - return Res; - } while (pos != std::string::npos); - return 0; -} - /* Lock file suffix */ const char LockCacheItem::LockSuffix[] = ".lock"; @@ -91,7 +65,7 @@ bool PersistentDeviceCodeCache::isImageCached(const RTDeviceBinaryImage &Img) { return true; } -/* Stores build program in persisten cache +/* Stores built program in persisten cache */ void PersistentDeviceCodeCache::putItemToDisc( const device &Device, const RTDeviceBinaryImage &Img, @@ -109,7 +83,7 @@ void PersistentDeviceCodeCache::putItemToDisc( std::string FileName; do { FileName = DirName + "/" + std::to_string(i++); - } while (isPathPresent(FileName + ".bin")); + } while (OSUtil::isPathPresent(FileName + ".bin")); unsigned int DeviceNum = 0; @@ -134,7 +108,7 @@ void PersistentDeviceCodeCache::putItemToDisc( Pointers.data(), nullptr); try { - makeDir(DirName.c_str()); + OSUtil::makeDir(DirName.c_str()); LockCacheItem Lock{FileName}; if (Lock.isOwned()) { writeBinaryDataToFile(FileName + ".bin", Result); @@ -161,13 +135,14 @@ std::vector> PersistentDeviceCodeCache::getItemFromDisc( std::string Path = getCacheItemPath(Device, Img, SpecConsts, BuildOptionsString); - if (!isPathPresent(Path)) + if (!OSUtil::isPathPresent(Path)) return {}; int i = 0; std::string FileName{Path + "/" + std::to_string(i)}; - while (isPathPresent(FileName + ".bin") || isPathPresent(FileName + ".src")) { + while (OSUtil::isPathPresent(FileName + ".bin") || + OSUtil::isPathPresent(FileName + ".src")) { if (!LockCacheItem::isLocked(FileName) && isCacheItemSrcEqual(FileName + ".src", Device, Img, SpecConsts, diff --git a/sycl/source/detail/persistent_device_code_cache.hpp b/sycl/source/detail/persistent_device_code_cache.hpp index ee62332383c3c..ed7201dab9198 100644 --- a/sycl/source/detail/persistent_device_code_cache.hpp +++ b/sycl/source/detail/persistent_device_code_cache.hpp @@ -9,6 +9,7 @@ #pragma once #include +#include #include #include #include @@ -27,24 +28,15 @@ namespace detail { * is moved to c++17 standard*/ std::string getDirName(const char *Path); -/// Checks if specified path is present -inline bool isPathPresent(const std::string &Path) { - struct stat Stat; - return !stat(Path.c_str(), &Stat); -} - -/// Make directory recursively and returns zero code on success -int makeDir(const char *Dir); - /* The class manages inter-process synchronization: * - Path passed to the constructor is appended with .lock and used as lock * file. * - All operations are not blocking and failure ignoring (diagnostic may be - * send to std::cerr when SYCL_CHACE_TRACE environment variable is set). + * sent to std::cerr when SYCL_CACHE_TRACE environment variable is set). * - There are two modes of accessing shared resource: - * - write access assumes that lock is aquired (object is created and + * - write access assumes that lock is acquired (object is created and * isOwned() method confirms that current executor owns the lock); - * - read access checks that the lock is not aquired for write by others + * - read access checks that the lock is not acquired for write by others * with the help of isLocked() method. */ class LockCacheItem { @@ -58,7 +50,7 @@ class LockCacheItem { bool isOwned() { return Owned; } static bool isLocked(const std::string &Path) { - return isPathPresent(Path + LockSuffix); + return OSUtil::isPathPresent(Path + LockSuffix); } ~LockCacheItem(); }; @@ -197,7 +189,7 @@ class PersistentDeviceCodeCache { static void trace(const std::string &msg) { static const char *TraceEnabled = SYCLConfig::get(); if (TraceEnabled) - std::cerr << msg << std::endl; + std::cerr << "*** Code caching: " << msg << std::endl; } }; } // namespace detail diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 29112823a79ef..0e548b349fc8c 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -459,7 +459,7 @@ RT::PiProgram ProgramManager::getBuiltPIProgram(OSModuleHandle M, NativePrograms[BuiltProgram.get()] = &Img; } - // Save program to persistent cache if it not there + // Save program to persistent cache if it is not there if (!BinProg.size()) PersistentDeviceCodeCache::putItemToDisc( Device, Img, SpecConsts, CompileOpts + LinkOpts, BuiltProgram.get()); diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index 4601549761311..0c252ed96a7e9 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3807,6 +3807,7 @@ _ZN2cl4sycl6detail6OSUtil12alignedAllocEmm _ZN2cl4sycl6detail6OSUtil12getOSMemSizeEv _ZN2cl4sycl6detail6OSUtil16getCurrentDSODirB5cxx11Ev _ZN2cl4sycl6detail6OSUtil17getOSModuleHandleEPKv +_ZN2cl4sycl6detail6OSUtil7makeDirEPKc _ZN2cl4sycl6detail9join_implERKSt6vectorISt10shared_ptrINS1_18kernel_bundle_implEESaIS5_EE _ZN2cl4sycl6detail9link_implERKSt6vectorINS0_13kernel_bundleILNS0_12bundle_stateE1EEESaIS5_EERKS2_INS0_6deviceESaISA_EERKNS0_13property_listE _ZN2cl4sycl6device11get_devicesENS0_4info11device_typeE diff --git a/sycl/test/on-device/cache/basic.cpp b/sycl/test/on-device/cache/basic.cpp index e94e7f07bcbef..552aa920f1f23 100644 --- a/sycl/test/on-device/cache/basic.cpp +++ b/sycl/test/on-device/cache/basic.cpp @@ -13,7 +13,6 @@ #include "basic.hpp" // CHECK-BUILD: piProgramBuild -// CHECK-BUILD: piProgramCreateWithBinary // CHECK-CACHE-NOT: piProgramBuild // CHECK-CACHE: piProgramCreateWithBinary diff --git a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp index 5d3fe7acf0dee..2bbc4e90125e6 100644 --- a/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp +++ b/sycl/unittests/kernel-and-program/PersistentDeviceCodeCache.cpp @@ -271,7 +271,7 @@ TEST_F(PersistenDeviceCodeCache, LockFile) { Dev, Img, {}, BuildOptions, NativeProg); assert(Res.size() == 0 && "Locked item was read"); - // First cache item was anlocked and successfully read + // First cache item was unlocked and successfully read std::remove(LockFile.c_str()); Res = detail::PersistentDeviceCodeCache::getItemFromDisc( Dev, Img, {}, BuildOptions, NativeProg); From 740408b22e7679179a9e800cb8f2fe93439fe13a Mon Sep 17 00:00:00 2001 From: Vladimir Lazarev Date: Thu, 8 Apr 2021 13:22:05 +0300 Subject: [PATCH 31/32] Disable test on CUDA because non-SPIRV target are not covered by persistent cache --- sycl/test/on-device/cache/basic.cpp | 2 +- sycl/test/on-device/cache/cache_env_vars.cpp | 2 +- 2 files changed, 2 insertions(+), 2 deletions(-) diff --git a/sycl/test/on-device/cache/basic.cpp b/sycl/test/on-device/cache/basic.cpp index 552aa920f1f23..d033cdce97a26 100644 --- a/sycl/test/on-device/cache/basic.cpp +++ b/sycl/test/on-device/cache/basic.cpp @@ -1,5 +1,5 @@ // No JITing for host devices. -// REQUIRES: opencl || level_zero || cuda +// REQUIRES: opencl || level_zero // RUN: rm -rf %t/cache_dir // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: env SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD diff --git a/sycl/test/on-device/cache/cache_env_vars.cpp b/sycl/test/on-device/cache/cache_env_vars.cpp index 0d27fa9c9b596..39cb403362847 100644 --- a/sycl/test/on-device/cache/cache_env_vars.cpp +++ b/sycl/test/on-device/cache/cache_env_vars.cpp @@ -1,5 +1,5 @@ // No JITing for host devices. -// REQUIRES: opencl || level_zero || cuda +// REQUIRES: opencl || level_zero // RUN: rm -rf %t/cache_dir // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -DTARGET_IMAGE=INC100 // Build program and add item to cache From 163fb07576cf546d3bb00736716da5d502071fff Mon Sep 17 00:00:00 2001 From: vladimirlaz Date: Thu, 8 Apr 2021 13:44:52 +0300 Subject: [PATCH 32/32] Apply suggestions from code review Co-authored-by: sergei <57672082+s-kanaev@users.noreply.github.com> --- sycl/doc/EnvironmentVariables.md | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/doc/EnvironmentVariables.md b/sycl/doc/EnvironmentVariables.md index af357699e82fc..dc93cf8de8fb1 100644 --- a/sycl/doc/EnvironmentVariables.md +++ b/sycl/doc/EnvironmentVariables.md @@ -37,7 +37,7 @@ subject to change. Do not rely on these variables in production code. | SYCL_HOST_UNIFIED_MEMORY | Integer | Enforce host unified memory support or lack of it for the execution graph builder. If set to 0, it is enforced as not supported by all devices. If set to 1, it is enforced as supported by all devices. | | SYCL_CACHE_DIR | Path | Path to persistent cache root directory. Default values are `%AppData%\libsycl_cache` for Windows and `$XDG_CACHE_HOME/libsycl_cache` on Linux, if XDG_CACHE_HOME is not set then `$HOME/.cache/libsycl_cache`. | | SYCL_CACHE_TRACE | Any(\*) | Enables printing messages to std::cerr when non-blocking failures happen (e.g. unable to access cache item file). Default values if off. | -|| SYCL_CACHE_DISABLE_PERSISTENT | Any(\*) | Switches persistent cache switch off. Default value is ON. | +| SYCL_CACHE_DISABLE_PERSISTENT | Any(\*) | Switches persistent cache switch off. Default value is ON. | | SYCL_CACHE_EVICTION_DISABLE | Any(\*) | Switches cache eviction off. Default value is ON. | | SYCL_CACHE_MAX_SIZE | Positive integer | Cache eviction is triggered once total size of cached images exceeds the value in megabytes (default - 8 192 for 8 GB). Set to 0 to disable size-based cache eviction. | | SYCL_CACHE_THRESHOLD | Positive integer | Cache eviction threshold in days (default value is 7 for 1 week). Set to 0 for disabling time-based cache eviction. |