Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

[SYCL] Move bfloat support from experimental to supported. #6524

Merged
merged 99 commits into from
Nov 28, 2022
Merged
Show file tree
Hide file tree
Changes from 86 commits
Commits
Show all changes
99 commits
Select commit Hold shift + click to select a range
6014cef
[SYCL] Move bfloat support from experimental to supported.
rdeodhar Aug 3, 2022
bdd88e5
Corrections to tests.
rdeodhar Aug 3, 2022
73ed541
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Aug 24, 2022
0fe1884
Moved another file out of experimental space.
rdeodhar Aug 24, 2022
feb9d5f
Responses to review comments.
rdeodhar Aug 25, 2022
129f53f
Removed unneeded sycl::half conversion and updated doc.
rdeodhar Aug 26, 2022
2115f09
Added conversion from sycl::half to bfloat16.
rdeodhar Aug 29, 2022
3c2eb80
Cleanup of documentation.
rdeodhar Aug 31, 2022
74aa175
Hooked up bfloat16 aspect within OpenCL plugin.
rdeodhar Sep 2, 2022
bd05711
Support for bfloat16 aspect, and native or fallback support.
rdeodhar Sep 8, 2022
f8e894c
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 8, 2022
2ad68f6
Formatting changes.
rdeodhar Sep 8, 2022
4b78c03
Formatting changes.
rdeodhar Sep 8, 2022
0fce16d
Update to documentation.
rdeodhar Sep 8, 2022
4bcb383
Deprecate bfloat16 aspect.
rdeodhar Sep 8, 2022
35308f8
Fixes for ESIMD.
rdeodhar Sep 9, 2022
fa045e2
Reinstated to_float and from_float, used by NVidia, updated doc.
rdeodhar Sep 9, 2022
3322d6a
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 12, 2022
b12fd94
Update to doc.
rdeodhar Sep 12, 2022
87b0f09
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 14, 2022
f217eb4
Corrections to headers.
rdeodhar Sep 14, 2022
a908b11
Formatting change.
rdeodhar Sep 14, 2022
aab4c78
bfloat16 class supports all sm_xx devices.
Sep 15, 2022
a2568ba
Merge pull request #1 from JackAKirk/bfloat16-cuda-allarch
rdeodhar Sep 15, 2022
4d7a22b
Changes to keep bfloat math functions experimental for now.
rdeodhar Sep 16, 2022
38e5ad4
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 16, 2022
b9accad
Cleanup of bfloat16_math extension.
rdeodhar Sep 16, 2022
ca7880a
Document updates and minor changes.
rdeodhar Sep 19, 2022
dc3b2b5
Fixes for long lines in doc, a different way to check for NaN.
rdeodhar Sep 19, 2022
c955d36
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 20, 2022
1aa6ad3
Broke long lines into multiple lines.
rdeodhar Sep 20, 2022
ff04ce1
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 21, 2022
802f502
Changed library order on Windows.
rdeodhar Sep 21, 2022
8d7f46a
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 22, 2022
190f2a3
Fix for AOT compilation and correction to new headers.
rdeodhar Sep 22, 2022
84c50f3
Noted AOT limitation in doc.
rdeodhar Sep 23, 2022
df058ba
Adjustment for AOT compilation.
rdeodhar Sep 24, 2022
fed4d1d
Fixes for AOT builds.
rdeodhar Sep 26, 2022
28259d0
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 26, 2022
c11115b
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 26, 2022
6b05a2a
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 27, 2022
a82d73a
Fixes for AOT multiple devices.
rdeodhar Sep 27, 2022
3fc8885
Updated documentation.
rdeodhar Sep 27, 2022
1ec6838
Added back missing Status section in documentation.
rdeodhar Sep 27, 2022
105094b
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 27, 2022
432e775
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Sep 29, 2022
c135643
Added tests, corrected aspect check.
rdeodhar Oct 1, 2022
4eca414
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 1, 2022
8876ac8
Added missing newlines.
rdeodhar Oct 3, 2022
f0f2727
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 3, 2022
17673bf
Corrections to tests and macros, added host code emulation.
rdeodhar Oct 4, 2022
1094b8c
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 4, 2022
8d40228
Small corrections.
rdeodhar Oct 4, 2022
c5a85cf
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 4, 2022
cf8f6e0
Fixes for AOT.
rdeodhar Oct 4, 2022
5e50646
Formatting change.
rdeodhar Oct 4, 2022
45d3e70
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 5, 2022
a7be718
Renamed the bfloat aspects.
rdeodhar Oct 5, 2022
cac1c18
Fixes for generic JIT compilation.
rdeodhar Oct 6, 2022
208c09a
Changes for AOT sycl-targets switch.
rdeodhar Oct 6, 2022
46f406d
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 6, 2022
6830857
Corrected aspects queries.
rdeodhar Oct 6, 2022
46e5278
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 6, 2022
10fc9a3
Change in the way fallback/native libs are selected.
rdeodhar Oct 8, 2022
6195545
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 8, 2022
437e34a
Changed type of string.
rdeodhar Oct 10, 2022
09dc4c5
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 12, 2022
386353e
Replaced bfloat16 aspect with bfloat16_math_functions aspect.
rdeodhar Oct 12, 2022
0f93586
Improved devices check in clang driver.
rdeodhar Oct 13, 2022
48f3cac
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 13, 2022
d33cb10
Enhanced test for improved bfloat16 target detection.
rdeodhar Oct 13, 2022
28992c2
Updated bfloat16 driver test for windows.
rdeodhar Oct 13, 2022
ec28c8b
Use STL for parsing devices.
rdeodhar Oct 13, 2022
b958fc7
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 24, 2022
ec70b20
Allow spir64 target to be JIT even when combined with AOT targets.
rdeodhar Oct 24, 2022
1b86012
Updated documentation.
rdeodhar Oct 24, 2022
3e1e681
Modifications for mixed JIT and AOT compilations, added tests.
rdeodhar Oct 25, 2022
8c633d3
Corrections to comments.
rdeodhar Oct 25, 2022
1a59e03
Update to documentation.
rdeodhar Oct 25, 2022
b2fd6cc
Updated doc.
rdeodhar Oct 25, 2022
fab2e54
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 26, 2022
35b8910
Adjustments to tests.
rdeodhar Oct 27, 2022
a05c872
Test cleanup.
rdeodhar Oct 27, 2022
ac5f603
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Oct 27, 2022
6d45ed1
Adjustments to more tests.
rdeodhar Oct 27, 2022
077d0fe
Change to tests to ensure AOT components are available.
rdeodhar Oct 28, 2022
2ff6a9d
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Nov 7, 2022
d7c80ee
Adjustment to test for new bfloat16 header.
rdeodhar Nov 7, 2022
20d13df
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Nov 8, 2022
cd1d0a2
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Nov 15, 2022
4bf60b9
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Nov 18, 2022
45c32f7
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Nov 21, 2022
5de1bf7
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Nov 22, 2022
6ec2bb9
Changes for indirect accesses.
rdeodhar Nov 22, 2022
49e9cd1
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Nov 22, 2022
2065060
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Nov 23, 2022
e24e57b
Fixed conflicts.
rdeodhar Nov 23, 2022
41098ab
Merge branch 'sycl' of https://github.com/intel/llvm into bfloat16
rdeodhar Nov 25, 2022
37b05f0
Correction to library list.
rdeodhar Nov 25, 2022
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
2 changes: 1 addition & 1 deletion clang/lib/Basic/Targets/NVPTX.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -181,7 +181,7 @@ void NVPTXTargetInfo::getTargetDefines(const LangOptions &Opts,
MacroBuilder &Builder) const {
Builder.defineMacro("__PTX__");
Builder.defineMacro("__NVPTX__");
if (Opts.CUDAIsDevice || Opts.OpenMPIsDevice) {
if (Opts.CUDAIsDevice || Opts.OpenMPIsDevice || Opts.SYCLIsDevice) {
// Set __CUDA_ARCH__ for the GPU specified.
std::string CUDAArchCode = [this] {
switch (GPU) {
Expand Down
90 changes: 89 additions & 1 deletion clang/lib/Driver/Driver.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -101,6 +101,8 @@
#include <cstdlib> // ::getenv
#include <map>
#include <memory>
#include <regex>
#include <sstream>
#include <utility>
#if LLVM_ON_UNIX
#include <unistd.h> // getpid
Expand Down Expand Up @@ -5121,6 +5123,76 @@ class OffloadingActionBuilder final {
}
}

// Return whether to use native bfloat16 library.
bool selectBfloatLibs(const ToolChain *TC, bool &useNative) {
const OptTable &Opts = C.getDriver().getOpts();
const char *TargetOpt = nullptr;
const char *DeviceOpt = nullptr;
bool needLibs = false;
for (auto *A : Args) {
llvm::Triple *TargetBE = nullptr;

auto GetTripleIt = [&, this](llvm::StringRef Triple) {
llvm::Triple TargetTriple{Triple};
auto TripleIt = llvm::find_if(SYCLTripleList, [&](auto &SYCLTriple) {
return SYCLTriple == TargetTriple;
});
return TripleIt != SYCLTripleList.end() ? &*TripleIt : nullptr;
};

if (A->getOption().matches(options::OPT_fsycl_targets_EQ)) {
// spir64 target is actually JIT compilation, so we defer selection of
// bfloat16 libraries to runtime. For AOT we need libraries.
needLibs = TC->getTriple().getSubArch() != llvm::Triple::NoSubArch;
TargetBE = GetTripleIt(A->getValue(0));
if (TargetBE)
TargetOpt = A->getValue(0);
else
continue;
} else if (A->getOption().matches(options::OPT_Xsycl_backend_EQ)) {
// Passing device args: -Xsycl-target-backend=<triple> <opt>
TargetBE = GetTripleIt(A->getValue(0));
if (TargetBE)
DeviceOpt = A->getValue(1);
else
continue;
} else if (A->getOption().matches(options::OPT_Xsycl_backend)) {
// Passing device args: -Xsycl-target-backend <opt>
TargetBE = &SYCLTripleList.front();
DeviceOpt = A->getValue(0);
} else if (A->getOption().matches(options::OPT_Xs_separate)) {
// Passing device args: -Xs <opt>
DeviceOpt = A->getValue(0);
} else {
mdtoguchi marked this conversation as resolved.
Show resolved Hide resolved
continue;
};
}
useNative = false;
if (needLibs)
if (TC->getTriple().getSubArch() == llvm::Triple::SPIRSubArch_gen &&
TargetOpt && DeviceOpt) {

auto checkBF = [=](std::string &Dev) {
static const std::regex BFFs("pvc.*|ats.*");
return std::regex_match(Dev, BFFs);
};

needLibs = true;
std::string Params{DeviceOpt};
size_t DevicesPos = Params.find("-device ");
useNative = false;
if (DevicesPos != std::string::npos) {
useNative = true;
std::istringstream Devices(Params.substr(DevicesPos + 8));
for (std::string S; std::getline(Devices, S, ',');) {
useNative &= checkBF(S);
}
}
}

return needLibs;
}

bool addSYCLDeviceLibs(const ToolChain *TC, ActionList &DeviceLinkObjects,
bool isSpirvAOT, bool isMSVCEnv) {
struct DeviceLibOptInfo {
Expand All @@ -5134,7 +5206,8 @@ class OffloadingActionBuilder final {
// of "internal" libraries cannot be affected via -fno-sycl-device-lib.
llvm::StringMap<bool> devicelib_link_info = {
{"libc", true}, {"libm-fp32", true}, {"libm-fp64", true},
{"libimf-fp32", true}, {"libimf-fp64", true}, {"internal", true}};
{"libimf-fp32", true}, {"libimf-fp64", true}, {"libm-bfloat16", true},
{"internal", true}};
if (Arg *A = Args.getLastArg(options::OPT_fsycl_device_lib_EQ,
options::OPT_fno_sycl_device_lib_EQ)) {
if (A->getValues().size() == 0)
Expand Down Expand Up @@ -5193,6 +5266,10 @@ class OffloadingActionBuilder final {
{"libsycl-fallback-cmath-fp64", "libm-fp64"},
{"libsycl-fallback-imf", "libimf-fp32"},
{"libsycl-fallback-imf-fp64", "libimf-fp64"}};
const SYCLDeviceLibsList sycl_device_bfloat16_fallback_lib = {
{"libsycl-fallback-bfloat16", "libm-bfloat16"}};
const SYCLDeviceLibsList sycl_device_bfloat16_native_lib = {
{"libsycl-native-bfloat16", "libm-bfloat16"}};
// ITT annotation libraries are linked in separately whenever the device
// code instrumentation is enabled.
const SYCLDeviceLibsList sycl_device_annotation_libs = {
Expand Down Expand Up @@ -5242,6 +5319,17 @@ class OffloadingActionBuilder final {
addInputs(sycl_device_wrapper_libs);
if (isSpirvAOT || TC->getTriple().isNVPTX())
addInputs(sycl_device_fallback_libs);

bool nativeBfloatLibs;
bool needBfloatLibs = selectBfloatLibs(TC, nativeBfloatLibs);
if (needBfloatLibs) {
// Add native or fallback bfloat16 library.
if (nativeBfloatLibs)
addInputs(sycl_device_bfloat16_native_lib);
else
addInputs(sycl_device_bfloat16_fallback_lib);
}
mdtoguchi marked this conversation as resolved.
Show resolved Hide resolved

if (Args.hasFlag(options::OPT_fsycl_instrument_device_code,
options::OPT_fno_sycl_instrument_device_code, true))
addInputs(sycl_device_annotation_libs);
Expand Down
Loading