Skip to content

Commit 828dd3c

Browse files
author
iclsrc
committed
Merge from 'sycl' to 'sycl-web'
2 parents a7bc1a7 + a395886 commit 828dd3c

File tree

99 files changed

+2629
-1484
lines changed

Some content is hidden

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

99 files changed

+2629
-1484
lines changed

‎.github/workflows/sycl_containers.yaml

Lines changed: 115 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -135,3 +135,118 @@ jobs:
135135
fpgaemu_tag=${{fromJson(steps.deps.outputs.deps).linux_staging.fpgaemu.github_tag}}
136136
cpu_tag=${{fromJson(steps.deps.outputs.deps).linux_staging.oclcpu.github_tag}}
137137
138+
base_image_ubuntu2204:
139+
if: github.repository == 'intel/llvm'
140+
name: Base Ubuntu 22.04 Docker image
141+
runs-on: ubuntu-latest
142+
steps:
143+
- name: Checkout
144+
uses: actions/checkout@v3
145+
with:
146+
fetch-depth: 2
147+
- name: Build and Push Container
148+
uses: ./devops/actions/build_container
149+
with:
150+
push: ${{ github.event_name != 'pull_request' }}
151+
file: ubuntu2204_base
152+
username: ${{ github.repository_owner }}
153+
password: ${{ secrets.GITHUB_TOKEN }}
154+
tags: |
155+
ghcr.io/${{ github.repository }}/ubuntu2204_base:${{ github.sha }}
156+
ghcr.io/${{ github.repository }}/ubuntu2204_base:latest
157+
build_image_ubuntu2204:
158+
if: github.repository == 'intel/llvm'
159+
name: Build Ubuntu Docker image
160+
runs-on: ubuntu-latest
161+
steps:
162+
- name: Checkout
163+
uses: actions/checkout@v3
164+
with:
165+
fetch-depth: 2
166+
- name: Build and Push Container
167+
uses: ./devops/actions/build_container
168+
with:
169+
push: ${{ github.event_name != 'pull_request' }}
170+
file: ubuntu2204_build
171+
username: ${{ github.repository_owner }}
172+
password: ${{ secrets.GITHUB_TOKEN }}
173+
tags: |
174+
ghcr.io/${{ github.repository }}/ubuntu2204_build:${{ github.sha }}
175+
ghcr.io/${{ github.repository }}/ubuntu2204_build:latest
176+
177+
# This job produces a Docker container with the latest versions of Intel
178+
# drivers, that can be found on GitHub.
179+
drivers_image_ubuntu2204:
180+
if: github.repository == 'intel/llvm'
181+
name: Intel Drivers Ubuntu 22.04 Docker image
182+
runs-on: ubuntu-latest
183+
needs: base_image_ubuntu2204
184+
steps:
185+
- name: Checkout
186+
uses: actions/checkout@v3
187+
with:
188+
fetch-depth: 2
189+
- name: Get dependencies configuration
190+
id: deps
191+
run: |
192+
DEPS=`cat devops/dependencies.json`
193+
DEPS="${DEPS//'%'/'%25'}"
194+
DEPS="${DEPS//$'\n'/'%0A'}"
195+
DEPS="${DEPS//$'\r'/'%0D'}"
196+
echo $DEPS
197+
echo "::set-output name=deps::$DEPS"
198+
- name: Build and Push Container
199+
uses: ./devops/actions/build_container
200+
with:
201+
push: ${{ github.event_name != 'pull_request' }}
202+
file: ubuntu2204_intel_drivers
203+
username: ${{ github.repository_owner }}
204+
password: ${{ secrets.GITHUB_TOKEN }}
205+
tags: |
206+
ghcr.io/${{ github.repository }}/ubuntu2204_intel_drivers:latest-${{ github.sha }}
207+
ghcr.io/${{ github.repository }}/ubuntu2204_intel_drivers:latest
208+
build-args: |
209+
compute_runtime_tag=${{fromJson(steps.deps.outputs.deps).linux.compute_runtime.github_tag}}
210+
igc_tag=${{fromJson(steps.deps.outputs.deps).linux.igc.github_tag}}
211+
tbb_tag=${{fromJson(steps.deps.outputs.deps).linux.tbb.github_tag}}
212+
fpgaemu_tag=${{fromJson(steps.deps.outputs.deps).linux.fpgaemu.github_tag}}
213+
cpu_tag=${{fromJson(steps.deps.outputs.deps).linux.oclcpu.github_tag}}
214+
215+
# This job produces a Docker container with the latest versions of Intel
216+
# drivers, that can be found on GitHub.
217+
drivers_image_ubuntu2204_unstable:
218+
if: github.repository == 'intel/llvm'
219+
name: Intel Drivers (unstable) Ubuntu 22.04 Docker image
220+
runs-on: ubuntu-latest
221+
needs: base_image_ubuntu2204
222+
steps:
223+
- name: Checkout
224+
uses: actions/checkout@v3
225+
with:
226+
fetch-depth: 2
227+
- name: Get dependencies configuration
228+
id: deps
229+
run: |
230+
DEPS=`cat devops/dependencies.json`
231+
DEPS="${DEPS//'%'/'%25'}"
232+
DEPS="${DEPS//$'\n'/'%0A'}"
233+
DEPS="${DEPS//$'\r'/'%0D'}"
234+
echo $DEPS
235+
echo "::set-output name=deps::$DEPS"
236+
- name: Build and Push Container
237+
uses: ./devops/actions/build_container
238+
with:
239+
push: ${{ github.event_name != 'pull_request' }}
240+
file: ubuntu2204_intel_drivers
241+
username: ${{ github.repository_owner }}
242+
password: ${{ secrets.GITHUB_TOKEN }}
243+
tags: |
244+
ghcr.io/${{ github.repository }}/ubuntu2204_intel_drivers:unstable-${{ github.sha }}
245+
ghcr.io/${{ github.repository }}/ubuntu2204_intel_drivers:unstable
246+
build-args: |
247+
compute_runtime_tag=${{fromJson(steps.deps.outputs.deps).linux_staging.compute_runtime.github_tag}}
248+
igc_tag=${{fromJson(steps.deps.outputs.deps).linux_staging.igc.github_tag}}
249+
tbb_tag=${{fromJson(steps.deps.outputs.deps).linux_staging.tbb.github_tag}}
250+
fpgaemu_tag=${{fromJson(steps.deps.outputs.deps).linux_staging.fpgaemu.github_tag}}
251+
cpu_tag=${{fromJson(steps.deps.outputs.deps).linux_staging.oclcpu.github_tag}}
252+

‎.github/workflows/sycl_precommit.yml

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -9,6 +9,7 @@ on:
99
- '.github/ISSUE_TEMPLATE/**'
1010
- '.github/CODEOWNERS'
1111
- '.github/workflows/sycl_update_gpu_driver.yml'
12+
- '.github/workflows/sycl_containers.yaml'
1213
- '.github/workflows/sycl_windows_build_and_test.yml'
1314
- '.github/workflows/sycl_macos_build_and_test.yml'
1415
- 'devops/containers/**'

‎clang/include/clang/Basic/DiagnosticFrontendKinds.td

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -283,8 +283,8 @@ def err_avx_calling_convention : Error<warn_avx_calling_convention.Text>;
283283

284284
def warn_sycl_device_has_aspect_mismatch
285285
: Warning<"function '%0' uses aspect '%1' not listed in its "
286-
"'sycl::device_has' attribute">, BackendInfo,
287-
InGroup<SyclAspectMismatch>;
286+
"%select{'device_has' property|'sycl::device_has' attribute}2">,
287+
BackendInfo, InGroup<SyclAspectMismatch>;
288288
def note_sycl_aspect_propagated_from_call
289289
: Note<"propagated from call to function '%0'">, BackendInfo;
290290

‎clang/include/clang/Basic/LangOptions.def

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -293,6 +293,8 @@ LANGOPT(
293293
LANGOPT(SYCLDisableRangeRounding, 1, 0, "Disable parallel for range rounding")
294294
LANGOPT(SYCLEnableIntHeaderDiags, 1, 0, "Enable diagnostics that require the "
295295
"SYCL integration header")
296+
LANGOPT(SYCLAllowVirtualFunctions, 1, 0,
297+
"Allow virtual functions calls in code for SYCL device")
296298

297299
LANGOPT(HIPUseNewLaunchAPI, 1, 0, "Use new kernel launching API for HIP")
298300

‎clang/include/clang/Driver/Options.td

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -6627,6 +6627,9 @@ def fsycl_use_main_file_name : Flag<["-"], "fsycl-use-main-file-name">,
66276627
HelpText<"Tells compiler that -main-file-name contains an absolute path and "
66286628
"file specified there should be used for checksum calculation.">,
66296629
MarshallingInfoFlag<CodeGenOpts<"SYCLUseMainFileName">>;
6630+
def fsycl_allow_virtual_functions : Flag<["-"], "fsycl-allow-virtual-functions">,
6631+
HelpText<"Allow virtual functions calls in code for SYCL device">,
6632+
MarshallingInfoFlag<LangOpts<"SYCLAllowVirtualFunctions">>;
66306633

66316634
} // let Flags = [CC1Option, NoDriverOption]
66326635

‎clang/lib/CodeGen/CGBuiltin.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -21467,7 +21467,7 @@ RValue CodeGenFunction::EmitIntelFPGAMemBuiltin(const CallExpr *E) {
2146721467

2146821468
llvm::Value *Ann = EmitAnnotationCall(F, PtrVal, AnnotStr, SourceLocation());
2146921469

21470-
cast<CallBase>(Ann)->addFnAttr(llvm::Attribute::ReadNone);
21470+
cast<CallBase>(Ann)->setDoesNotAccessMemory();
2147121471

2147221472
return RValue::get(Ann);
2147321473
}

‎clang/lib/CodeGen/CodeGenAction.cpp

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -867,7 +867,8 @@ void BackendConsumer::AspectMismatchDiagHandler(
867867
assert(LocCookie.isValid() &&
868868
"Invalid location for caller in aspect mismatch diagnostic");
869869
Diags.Report(LocCookie, diag::warn_sycl_device_has_aspect_mismatch)
870-
<< llvm::demangle(D.getFunctionName().str()) << D.getAspect();
870+
<< llvm::demangle(D.getFunctionName().str()) << D.getAspect()
871+
<< D.isFromDeviceHasAttribute();
871872
for (const std::pair<StringRef, unsigned> &CalleeInfo : D.getCallChain()) {
872873
LocCookie = SourceLocation::getFromRawEncoding(CalleeInfo.second);
873874
assert(LocCookie.isValid() &&

‎clang/lib/CodeGen/ItaniumCXXABI.cpp

Lines changed: 31 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -3573,8 +3573,33 @@ void ItaniumRTTIBuilder::BuildVTablePointer(const Type *Ty) {
35733573
// Check if the alias exists. If it doesn't, then get or create the global.
35743574
if (CGM.getItaniumVTableContext().isRelativeLayout())
35753575
VTable = CGM.getModule().getNamedAlias(VTableName);
3576-
if (!VTable)
3577-
VTable = CGM.getModule().getOrInsertGlobal(VTableName, CGM.Int8PtrTy);
3576+
3577+
// To generate valid device code global pointers should have global address
3578+
// space in SYCL.
3579+
bool GenTyInfoGVWithGlobalAS =
3580+
CGM.getLangOpts().SYCLIsDevice &&
3581+
CGM.getLangOpts().SYCLAllowVirtualFunctions &&
3582+
(VTableName == ClassTypeInfo || VTableName == SIClassTypeInfo);
3583+
auto VTableTy =
3584+
GenTyInfoGVWithGlobalAS
3585+
? CGM.Int8Ty->getPointerTo(
3586+
CGM.getContext().getTargetAddressSpace(LangAS::sycl_global))
3587+
: CGM.Int8PtrTy;
3588+
if (!VTable) {
3589+
if (GenTyInfoGVWithGlobalAS) {
3590+
VTable = CGM.getModule().getOrInsertGlobal(VTableName, VTableTy, [&] {
3591+
return new llvm::GlobalVariable(
3592+
CGM.getModule(), VTableTy, /*isConstant=*/false,
3593+
llvm::GlobalVariable::ExternalLinkage, /*Initializer=*/nullptr,
3594+
VTableName, /*InsertBefore=*/nullptr,
3595+
llvm::GlobalValue::ThreadLocalMode::NotThreadLocal,
3596+
llvm::Optional<unsigned>(
3597+
CGM.getContext().getTargetAddressSpace(LangAS::sycl_global)));
3598+
});
3599+
} else {
3600+
VTable = CGM.getModule().getOrInsertGlobal(VTableName, VTableTy);
3601+
}
3602+
}
35783603

35793604
CGM.setDSOLocal(cast<llvm::GlobalValue>(VTable->stripPointerCasts()));
35803605

@@ -3586,15 +3611,15 @@ void ItaniumRTTIBuilder::BuildVTablePointer(const Type *Ty) {
35863611
// The vtable address point is 8 bytes after its start:
35873612
// 4 for the offset to top + 4 for the relative offset to rtti.
35883613
llvm::Constant *Eight = llvm::ConstantInt::get(CGM.Int32Ty, 8);
3589-
VTable = llvm::ConstantExpr::getBitCast(VTable, CGM.Int8PtrTy);
3614+
VTable = llvm::ConstantExpr::getBitCast(VTable, VTableTy);
35903615
VTable =
35913616
llvm::ConstantExpr::getInBoundsGetElementPtr(CGM.Int8Ty, VTable, Eight);
35923617
} else {
35933618
llvm::Constant *Two = llvm::ConstantInt::get(PtrDiffTy, 2);
3594-
VTable = llvm::ConstantExpr::getInBoundsGetElementPtr(CGM.Int8PtrTy, VTable,
3595-
Two);
3619+
VTable =
3620+
llvm::ConstantExpr::getInBoundsGetElementPtr(VTableTy, VTable, Two);
35963621
}
3597-
VTable = llvm::ConstantExpr::getBitCast(VTable, CGM.Int8PtrTy);
3622+
VTable = llvm::ConstantExpr::getBitCast(VTable, VTableTy);
35983623

35993624
Fields.push_back(VTable);
36003625
}

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

Lines changed: 5 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -349,7 +349,11 @@ void tools::gnutools::Linker::constructLLVMARCommand(
349349
Compilation &C, const JobAction &JA, const InputInfo &Output,
350350
const InputInfoList &Input, const ArgList &Args) const {
351351
ArgStringList CmdArgs;
352-
CmdArgs.push_back("cr");
352+
// Use 'cqL' to create the archive. This allows for any fat archives that
353+
// are passed on the command line to be added via contents instead of the
354+
// full archive. Any usage of the generated archive will then have full
355+
// access to resolve any dependencies.
356+
CmdArgs.push_back("cqL");
353357
const char *OutputFilename = Output.getFilename();
354358
if (llvm::sys::fs::exists(OutputFilename)) {
355359
C.getDriver().Diag(clang::diag::warn_drv_existing_archive_append)

‎clang/lib/Headers/avxintrin.h

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -45,9 +45,11 @@ typedef _Float16 __v16hf __attribute__((__vector_size__(32), __aligned__(32)));
4545
typedef _Float16 __m256h __attribute__((__vector_size__(32), __aligned__(32)));
4646
typedef _Float16 __m256h_u __attribute__((__vector_size__(32), __aligned__(1)));
4747

48+
#ifndef __SYCL_DEVICE_ONLY__
4849
typedef __bf16 __v16bf __attribute__((__vector_size__(32), __aligned__(32)));
4950
typedef __bf16 __m256bh __attribute__((__vector_size__(32), __aligned__(32)));
5051
#endif
52+
#endif
5153

5254
/* Define the default attributes for the functions in this file. */
5355
#define __DEFAULT_FN_ATTRS __attribute__((__always_inline__, __nodebug__, __target__("avx"), __min_vector_width__(256)))

0 commit comments

Comments
 (0)