Skip to content

Commit 29a1f16

Browse files
baderbader
authored
[SYCL] Remove sycldevice target triple environment component (#3929)
Today DPC++ compiler requires users to set "sycldevice" target triple component to separate SYCL compilation mode for SPIR target from other compilation mode (e.g. OpenCL). Triple environment component is not intended to be used in such context and this patch removes `sycldevice`. This patch also preserves backward compatibility with binaries built using old versions of the compiler, which enforced `sycldevice` environment component of the target triple. It's provided by `clang-offload-bundler` tool, which now implicitly looks for the additional SYCL offload kind bundle with a "legacy" triple. A driver warning was added to let users know that `sycldevice` environment component is ignored now.
1 parent 6a4fa3a commit 29a1f16

File tree

240 files changed

+817
-808
lines changed

Some content is hidden

Large Commits have some content hidden by default. Use the searchbox below for content that may be hidden.

240 files changed

+817
-808
lines changed

‎clang/lib/Basic/Targets.cpp

Lines changed: 26 additions & 32 deletions
Original file line numberDiff line numberDiff line change
@@ -595,45 +595,39 @@ TargetInfo *AllocateTarget(const llvm::Triple &Triple,
595595
}
596596

597597
case llvm::Triple::spir: {
598-
if (Triple.getEnvironment() == llvm::Triple::SYCLDevice) {
599-
llvm::Triple HT(Opts.HostTriple);
600-
switch (HT.getOS()) {
601-
case llvm::Triple::Win32:
602-
switch (HT.getEnvironment()) {
603-
default: // Assume MSVC for unknown environments
604-
case llvm::Triple::MSVC:
605-
assert(HT.getArch() == llvm::Triple::x86 &&
606-
"Unsupported host architecture");
607-
return new MicrosoftX86_32SPIRTargetInfo(Triple, Opts);
608-
}
609-
case llvm::Triple::Linux:
610-
return new LinuxTargetInfo<SPIR32SYCLDeviceTargetInfo>(Triple, Opts);
611-
default:
612-
return new SPIR32SYCLDeviceTargetInfo(Triple, Opts);
598+
llvm::Triple HT(Opts.HostTriple);
599+
switch (HT.getOS()) {
600+
case llvm::Triple::Win32:
601+
switch (HT.getEnvironment()) {
602+
default: // Assume MSVC for unknown environments
603+
case llvm::Triple::MSVC:
604+
assert(HT.getArch() == llvm::Triple::x86 &&
605+
"Unsupported host architecture");
606+
return new MicrosoftX86_32SPIRTargetInfo(Triple, Opts);
613607
}
608+
case llvm::Triple::Linux:
609+
return new LinuxTargetInfo<SPIR32TargetInfo>(Triple, Opts);
610+
default:
611+
return new SPIR32TargetInfo(Triple, Opts);
614612
}
615-
return new SPIR32TargetInfo(Triple, Opts);
616613
}
617614

618615
case llvm::Triple::spir64: {
619-
if (Triple.getEnvironment() == llvm::Triple::SYCLDevice) {
620-
llvm::Triple HT(Opts.HostTriple);
621-
switch (HT.getOS()) {
622-
case llvm::Triple::Win32:
623-
switch (HT.getEnvironment()) {
624-
default: // Assume MSVC for unknown environments
625-
case llvm::Triple::MSVC:
626-
assert(HT.getArch() == llvm::Triple::x86_64 &&
627-
"Unsupported host architecture");
628-
return new MicrosoftX86_64_SPIR64TargetInfo(Triple, Opts);
629-
}
630-
case llvm::Triple::Linux:
631-
return new LinuxTargetInfo<SPIR64SYCLDeviceTargetInfo>(Triple, Opts);
632-
default:
633-
return new SPIR64SYCLDeviceTargetInfo(Triple, Opts);
616+
llvm::Triple HT(Opts.HostTriple);
617+
switch (HT.getOS()) {
618+
case llvm::Triple::Win32:
619+
switch (HT.getEnvironment()) {
620+
default: // Assume MSVC for unknown environments
621+
case llvm::Triple::MSVC:
622+
assert(HT.getArch() == llvm::Triple::x86_64 &&
623+
"Unsupported host architecture");
624+
return new MicrosoftX86_64_SPIR64TargetInfo(Triple, Opts);
634625
}
626+
case llvm::Triple::Linux:
627+
return new LinuxTargetInfo<SPIR64TargetInfo>(Triple, Opts);
628+
default:
629+
return new SPIR64TargetInfo(Triple, Opts);
635630
}
636-
return new SPIR64TargetInfo(Triple, Opts);
637631
}
638632

639633
case llvm::Triple::wasm32:

‎clang/lib/Basic/Targets/SPIR.h

Lines changed: 8 additions & 34 deletions
Original file line numberDiff line numberDiff line change
@@ -192,40 +192,13 @@ class LLVM_LIBRARY_VISIBILITY SPIR64TargetInfo : public SPIRTargetInfo {
192192
MacroBuilder &Builder) const override;
193193
};
194194

195-
class LLVM_LIBRARY_VISIBILITY SPIR32SYCLDeviceTargetInfo
196-
: public SPIR32TargetInfo {
197-
public:
198-
SPIR32SYCLDeviceTargetInfo(const llvm::Triple &Triple,
199-
const TargetOptions &Opts)
200-
: SPIR32TargetInfo(Triple, Opts) {
201-
// This is workaround for exception_ptr class.
202-
// Exceptions is not allowed in sycl device code but we should be able
203-
// to parse host code. So we allow compilation of exception_ptr but
204-
// if exceptions are used in device code we should emit a diagnostic.
205-
MaxAtomicInlineWidth = 32;
206-
}
207-
};
208-
209-
class LLVM_LIBRARY_VISIBILITY SPIR64SYCLDeviceTargetInfo
210-
: public SPIR64TargetInfo {
211-
public:
212-
SPIR64SYCLDeviceTargetInfo(const llvm::Triple &Triple,
213-
const TargetOptions &Opts)
214-
: SPIR64TargetInfo(Triple, Opts) {
215-
// This is workaround for exception_ptr class.
216-
// Exceptions is not allowed in sycl device code but we should be able
217-
// to parse host code. So we allow compilation of exception_ptr but
218-
// if exceptions are used in device code we should emit a diagnostic.
219-
MaxAtomicInlineWidth = 64;
220-
}
221-
};
222-
223195
// x86-32 SPIR Windows target
224196
class LLVM_LIBRARY_VISIBILITY WindowsX86_32SPIRTargetInfo
225-
: public WindowsTargetInfo<SPIR32SYCLDeviceTargetInfo> {
197+
: public WindowsTargetInfo<SPIR32TargetInfo> {
226198
public:
227-
WindowsX86_32SPIRTargetInfo(const llvm::Triple &Triple, const TargetOptions &Opts)
228-
: WindowsTargetInfo<SPIR32SYCLDeviceTargetInfo>(Triple, Opts) {
199+
WindowsX86_32SPIRTargetInfo(const llvm::Triple &Triple,
200+
const TargetOptions &Opts)
201+
: WindowsTargetInfo<SPIR32TargetInfo>(Triple, Opts) {
229202
DoubleAlign = LongLongAlign = 64;
230203
WCharType = UnsignedShort;
231204
}
@@ -265,10 +238,11 @@ class LLVM_LIBRARY_VISIBILITY MicrosoftX86_32SPIRTargetInfo
265238

266239
// x86-64 SPIR64 Windows target
267240
class LLVM_LIBRARY_VISIBILITY WindowsX86_64_SPIR64TargetInfo
268-
: public WindowsTargetInfo<SPIR64SYCLDeviceTargetInfo> {
241+
: public WindowsTargetInfo<SPIR64TargetInfo> {
269242
public:
270-
WindowsX86_64_SPIR64TargetInfo(const llvm::Triple &Triple, const TargetOptions &Opts)
271-
: WindowsTargetInfo<SPIR64SYCLDeviceTargetInfo>(Triple, Opts) {
243+
WindowsX86_64_SPIR64TargetInfo(const llvm::Triple &Triple,
244+
const TargetOptions &Opts)
245+
: WindowsTargetInfo<SPIR64TargetInfo>(Triple, Opts) {
272246
LongWidth = LongAlign = 32;
273247
DoubleAlign = LongLongAlign = 64;
274248
IntMaxType = SignedLongLong;

‎clang/lib/Driver/Driver.cpp

Lines changed: 13 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -923,6 +923,18 @@ void Driver::CreateOffloadingDeviceToolChains(Compilation &C,
923923
continue;
924924
}
925925

926+
// Warn about deprecated `sycldevice` environment component.
927+
if (TT.getEnvironmentName() == "sycldevice") {
928+
Diag(clang::diag::warn_drv_deprecated_arg)
929+
<< TT.str() << TT.getArchName();
930+
// Drop environment component.
931+
std::string EffectiveTriple =
932+
Twine(TT.getArchName() + "-" + TT.getVendorName() + "-" +
933+
TT.getOSName())
934+
.str();
935+
TT.setTriple(EffectiveTriple);
936+
}
937+
926938
// Store the current triple so that we can check for duplicates in
927939
// the following iterations.
928940
FoundNormalizedTriples[NormalizedName] = Val;
@@ -985,7 +997,7 @@ void Driver::CreateOffloadingDeviceToolChains(Compilation &C,
985997
else
986998
SYCLTargetArch = "spir64";
987999
else if (HasValidSYCLRuntime)
988-
// Triple for -fintelfpga is spir64_fpga-unknown-unknown-sycldevice.
1000+
// Triple for -fintelfpga is spir64_fpga.
9891001
SYCLTargetArch = SYCLfpga ? "spir64_fpga" : "spir64";
9901002
if (!SYCLTargetArch.empty()) {
9911003
UniqueSYCLTriplesVec.push_back(MakeSYCLDeviceTriple(SYCLTargetArch));
@@ -1877,7 +1889,6 @@ llvm::Triple Driver::MakeSYCLDeviceTriple(StringRef TargetArch) const {
18771889
TT.setArchName(TargetArch);
18781890
TT.setVendor(llvm::Triple::UnknownVendor);
18791891
TT.setOS(llvm::Triple::UnknownOS);
1880-
TT.setEnvironment(llvm::Triple::SYCLDevice);
18811892
return TT;
18821893
}
18831894
return llvm::Triple(TargetArch);
@@ -2759,7 +2770,6 @@ bool hasFPGABinary(Compilation &C, std::string Object, types::ID Type) {
27592770
TT.setArchName(types::getTypeName(Type));
27602771
TT.setVendorName("intel");
27612772
TT.setOS(llvm::Triple::UnknownOS);
2762-
TT.setEnvironment(llvm::Triple::SYCLDevice);
27632773

27642774
// Checking uses -check-section option with the input file, no output
27652775
// file and the target triple being looked for.
@@ -7470,7 +7480,6 @@ const ToolChain &Driver::getToolChain(const ArgList &Args,
74707480
break;
74717481
case llvm::Triple::MSVC:
74727482
case llvm::Triple::UnknownEnvironment:
7473-
case llvm::Triple::SYCLDevice:
74747483
if (Args.getLastArgValue(options::OPT_fuse_ld_EQ)
74757484
.startswith_insensitive("bfd"))
74767485
TC = std::make_unique<toolchains::CrossWindowsToolChain>(

‎clang/lib/Driver/ToolChains/Clang.cpp

Lines changed: 8 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -7627,9 +7627,9 @@ void Clang::AddClangCLArgs(const ArgList &Args, types::ID InputType,
76277627
bool *EmitCodeView) const {
76287628
unsigned RTOptionID = options::OPT__SLASH_MT;
76297629
bool isNVPTX = getToolChain().getTriple().isNVPTX();
7630-
bool isSYCLDevice =
7631-
getToolChain().getTriple().getEnvironment() == llvm::Triple::SYCLDevice;
7632-
bool isSYCL = Args.hasArg(options::OPT_fsycl) || isSYCLDevice;
7630+
bool isSPIR = getToolChain().getTriple().isSPIR();
7631+
// FIXME: isSYCL should not be enabled by "isSPIR"
7632+
bool isSYCL = Args.hasArg(options::OPT_fsycl) || isSPIR;
76337633
// For SYCL Windows, /MD is the default.
76347634
if (isSYCL)
76357635
RTOptionID = options::OPT__SLASH_MD;
@@ -7641,7 +7641,7 @@ void Clang::AddClangCLArgs(const ArgList &Args, types::ID InputType,
76417641

76427642
if (Arg *A = Args.getLastArg(options::OPT__SLASH_M_Group)) {
76437643
RTOptionID = A->getOption().getID();
7644-
if (isSYCL && !isSYCLDevice &&
7644+
if (isSYCL && !isSPIR &&
76457645
(RTOptionID == options::OPT__SLASH_MT ||
76467646
RTOptionID == options::OPT__SLASH_MTd))
76477647
// Use of /MT or /MTd is not supported for SYCL.
@@ -7653,9 +7653,9 @@ void Clang::AddClangCLArgs(const ArgList &Args, types::ID InputType,
76537653
auto addPreDefines = [&](unsigned Defines) {
76547654
if (Defines & addDEBUG)
76557655
CmdArgs.push_back("-D_DEBUG");
7656-
if (Defines & addMT && !isSYCLDevice)
7656+
if (Defines & addMT && !isSPIR)
76577657
CmdArgs.push_back("-D_MT");
7658-
if (Defines & addDLL && !isSYCLDevice)
7658+
if (Defines & addDLL && !isSPIR)
76597659
CmdArgs.push_back("-D_DLL");
76607660
};
76617661
StringRef FlagForCRT;
@@ -8346,7 +8346,6 @@ void OffloadBundler::ConstructJobMultipleOutputs(
83468346
TT.setArchName(types::getTypeName(InputType));
83478347
TT.setVendorName("intel");
83488348
TT.setOS(getToolChain().getTriple().getOS());
8349-
TT.setEnvironment(llvm::Triple::SYCLDevice);
83508349
Triples += "sycl-";
83518350
Triples += TT.normalize();
83528351
continue;
@@ -8462,7 +8461,7 @@ void OffloadWrapper::ConstructJob(Compilation &C, const JobAction &JA,
84628461
llvm::Triple TT = getToolChain().getTriple();
84638462
SmallString<128> TargetTripleOpt = TT.getArchName();
84648463
// When wrapping an FPGA device binary, we need to be sure to apply the
8465-
// appropriate triple that corresponds (fpga_aoc[xr]-intel-<os>-sycldevice)
8464+
// appropriate triple that corresponds (fpga_aoc[xr]-intel-<os>)
84668465
// to the target triple setting.
84678466
if (TT.getSubArch() == llvm::Triple::SPIRSubArch_fpga &&
84688467
TCArgs.hasArg(options::OPT_fsycl_link_EQ)) {
@@ -8474,7 +8473,6 @@ void OffloadWrapper::ConstructJob(Compilation &C, const JobAction &JA,
84748473
FPGAArch += "_emu";
84758474
TT.setArchName(FPGAArch);
84768475
TT.setVendorName("intel");
8477-
TT.setEnvironment(llvm::Triple::SYCLDevice);
84788476
TargetTripleOpt = TT.str();
84798477
// When wrapping an FPGA aocx binary to archive, do not emit registration
84808478
// functions
@@ -8724,7 +8722,7 @@ void SPIRVTranslator::ConstructJob(Compilation &C, const JobAction &JA,
87248722

87258723
TranslatorArgs.push_back("-o");
87268724
TranslatorArgs.push_back(Output.getFilename());
8727-
if (getToolChain().getTriple().isSYCLDeviceEnvironment()) {
8725+
if (JA.isDeviceOffloading(Action::OFK_SYCL)) {
87288726
TranslatorArgs.push_back("-spirv-max-version=1.3");
87298727
// TODO: align debug info for FPGA H/W when its SPIR-V consumer is ready
87308728
if (C.getDriver().isFPGAEmulationMode())

‎clang/lib/Driver/ToolChains/Linux.cpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -177,8 +177,8 @@ static StringRef getOSLibDir(const llvm::Triple &Triple, const ArgList &Args) {
177177

178178
Linux::Linux(const Driver &D, const llvm::Triple &Triple, const ArgList &Args)
179179
: Generic_ELF(D, Triple, Args) {
180-
// for SYCL device targets, we rely on the host GCC for proper compilation
181-
if (Triple.getEnvironment() == llvm::Triple::SYCLDevice) {
180+
// For SPIR device target, we rely on the host GCC for proper compilation.
181+
if (Triple.isSPIR()) {
182182
GCCInstallation.init(llvm::Triple(llvm::sys::getProcessTriple()), Args);
183183
} else {
184184
GCCInstallation.init(Triple, Args);

‎clang/test/CodeGenSYCL/accessor-readonly-invalid-lib.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -triple spir64-unknown-unknown-sycldevice -fsycl-is-device -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -triple spir64-unknown-unknown -fsycl-is-device -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
22
//
33
// Test which verifies that readonly attribute is generated for unexpected access mode value.
44

‎clang/test/CodeGenSYCL/accessor-readonly.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -triple spir64-unknown-unknown-sycldevice -fsycl-is-device -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -triple spir64-unknown-unknown -fsycl-is-device -emit-llvm %s -o - | FileCheck %s
22

33
// Test to check that readonly attribute is applied to accessors with access mode read.
44

‎clang/test/CodeGenSYCL/accessor_inheritance.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
22
#include "Inputs/sycl.hpp"
33

44
struct Base {

‎clang/test/CodeGenSYCL/accessor_no_alias_property.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
22

33
// This test validates the behaviour of noalias parameter attribute.
44

‎clang/test/CodeGenSYCL/address-space-cond-op.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,5 +1,5 @@
11
// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py
2-
// RUN: %clang_cc1 -x c++ -triple spir64-unknown-linux-sycldevice -disable-llvm-passes -fsycl-is-device -emit-llvm %s -o - | FileCheck %s
2+
// RUN: %clang_cc1 -x c++ -triple spir64-unknown-linux -disable-llvm-passes -fsycl-is-device -emit-llvm %s -o - | FileCheck %s
33

44
struct S {
55
unsigned short x;

‎clang/test/CodeGenSYCL/address-space-initializer.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice \
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown \
22
// RUN: -emit-llvm -disable-llvm-passes -o - %s | FileCheck %s
33

44
// This test checks that data for big constant initializer lists is placed

‎clang/test/CodeGenSYCL/address-space-new.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
22

33
struct SpaceWaster {
44
int i, j;

‎clang/test/CodeGenSYCL/address-space-of-returns.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -triple spir64-unknown-linux-sycldevice -fsycl-is-device -disable-llvm-passes -emit-llvm -x c++ %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -triple spir64-unknown-linux -fsycl-is-device -disable-llvm-passes -emit-llvm -x c++ %s -o - | FileCheck %s
22

33
struct A {
44
int B[42];

‎clang/test/CodeGenSYCL/address-space-parameter-conversions.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
22
void bar(int & Data) {}
33
// CHECK-DAG: define {{.*}}spir_func void @[[RAW_REF:[a-zA-Z0-9_]+]](i32 addrspace(4)* align 4 dereferenceable(4) %
44
void bar2(int & Data) {}

‎clang/test/CodeGenSYCL/anonymous_integration_footer.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -std=c++17 -triple spir64-unknown-unknown-sycldevice -fsycl-int-footer=%t.h %s -emit-llvm -o %t.ll
1+
// RUN: %clang_cc1 -fsycl-is-device -std=c++17 -triple spir64-unknown-unknown -fsycl-int-footer=%t.h %s -emit-llvm -o %t.ll
22
// RUN: FileCheck -input-file=%t.h %s
33
// A test that validates the more complex cases of the specialization-constant
44
// integration footer details, basically any situation we can come up with that

‎clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -std=c++17 -triple spir64-unknown-unknown-sycldevice -fsycl-int-footer=%t.h %s -emit-llvm -o %t.ll
1+
// RUN: %clang_cc1 -fsycl-is-device -std=c++17 -triple spir64-unknown-unknown -fsycl-int-footer=%t.h %s -emit-llvm -o %t.ll
22
// RUN: FileCheck -input-file=%t.h %s
33
// A test that validates the more complex cases of the specialization-constant
44
// integration footer details, basically any situation we can come up with that

‎clang/test/CodeGenSYCL/basic-kernel-wrapper.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
22

33
// This test checks that compiler generates correct kernel wrapper for basic
44
// case.

‎clang/test/CodeGenSYCL/buffer_location.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
22

33
// CHECK: define {{.*}}spir_kernel void @_ZTSZ4mainE15kernel_function{{.*}} !kernel_arg_buffer_location ![[MDBL:[0-9]+]]
44
// CHECK: ![[MDBL]] = !{i32 3, i32 -1, i32 -1, i32 -1, i32 -1, i32 -1, i32 2, i32 -1, i32 -1, i32 -1, i32 2, i32 -1, i32 -1, i32 -1, i32 -1}

‎clang/test/CodeGenSYCL/check-direct-attribute-propagation.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -sycl-std=2020 -emit-llvm -o - %s | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -triple spir64-unknown-unknown -disable-llvm-passes -sycl-std=2020 -emit-llvm -o - %s | FileCheck %s
22

33
// Tests for IR of [[intel::scheduler_target_fmax_mhz()]], [[intel::num_simd_work_items()]],
44
// [[intel::no_global_work_offset()]], [[intel::max_global_work_dim()]], [[sycl::reqd_sub_group_size()]],

‎clang/test/CodeGenSYCL/const-wg-init.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
22

33
#include "Inputs/sycl.hpp"
44

‎clang/test/CodeGenSYCL/device-functions.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
22

33
template <typename T>
44
T bar(T arg);

‎clang/test/CodeGenSYCL/device-indirectly-callable-attr.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
22

33
void helper() {}
44

‎clang/test/CodeGenSYCL/device-variables.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s
22

33
enum class test_type { value1, value2, value3 };
44

‎clang/test/CodeGenSYCL/disable_loop_pipelining.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -sycl-std=2020 -emit-llvm -o - %s | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -triple spir64-unknown-unknown -disable-llvm-passes -sycl-std=2020 -emit-llvm -o - %s | FileCheck %s
22

33
#include "sycl.hpp"
44

‎clang/test/CodeGenSYCL/emit-all-decls-crash.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -femit-all-decls -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -disable-llvm-passes -emit-llvm %s -femit-all-decls -o - | FileCheck %s
22

33
// This should not crash and we should not emit this declaration, even though
44
// we have 'emit-all-decls'.

‎clang/test/CodeGenSYCL/emit-kernel-in-virtual-func.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
22

33
template <typename name, typename Func>
44
__attribute__((sycl_kernel)) void kernel_single_task(const Func &kernelFunc) {

‎clang/test/CodeGenSYCL/emit-kernel-in-virtual-func2.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -emit-llvm %s -o - | FileCheck %s
1+
// RUN: %clang_cc1 -fsycl-is-device -triple spir64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
22

33
template <typename name, typename Func>
44
__attribute__((sycl_kernel)) void kernel_single_task(const Func &kernelFunc) {

0 commit comments

Comments
 (0)