Skip to content

[SYCL] Disable inlining in 'ID queries fit in int' test #3751

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.

Sign up for GitHub

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

Merged
merged 10 commits into from
May 27, 2021
Merged

[SYCL] Disable inlining in 'ID queries fit in int' test #3751

Changes from all commits
Commits
Show all changes
10 commits
Select commit Hold shift + click to select a range
e1ab75d
[SYCL] Disable inlining in 'ID queries fit in int' test
May 13, 2021
652dc2d
Fix typo
May 17, 2021
46723d3
Leave -O0
May 18, 2021
e541be3
Replace -fno-inline-functions with -disable-llvm-passes
May 18, 2021
b4eccf6
Remove redundant argument
May 19, 2021
8bdf821
Merge remote-tracking branch 'public/sycl' into fix-test
May 26, 2021
c44a25c
Make code the device one
May 26, 2021
6a81ac2
Fix style issue
May 26, 2021
af1f0ef
Eliminate use of runtime queue and macro to disable parallel for rang…
May 26, 2021
1b5ca99
Removed main
May 26, 2021
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
20 changes: 7 additions & 13 deletions sycl/test/check_device_code/id_queries_fit_int.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,25 +1,22 @@
// RUN: %clangxx -fsycl -fsycl-id-queries-fit-in-int -O0 -c -S -emit-llvm -o %t.ll %s
// RUN: %clangxx -fsycl-device-only -fsycl-id-queries-fit-in-int -fno-sycl-early-optimizations -S -o %t.ll %s
// RUN: FileCheck %s --input-file %t.ll

#include <CL/sycl.hpp>

using namespace sycl;

// CHECK: define dso_local i32 @main() {{.*}} {
int main() {
item<1, true> TestItem = detail::Builder::createItem<1, true>({3}, {2}, {1});
// CHECK: define {{.*}}dso_local spir_func void @{{.*}}testItem{{.*}}(%"class.{{.*}}item"*{{.*}}%TestItem)
SYCL_EXTERNAL void testItem(item<1> TestItem) {
// CHECK: call void @llvm.assume(i1 {{.*}})
int Id = TestItem.get_id(0);
// CHECK: call void @llvm.assume(i1 {{.*}})
int Range = TestItem.get_range(0);
// CHECK: call void @llvm.assume(i1 {{.*}})
int LinearId = TestItem.get_linear_id();
}

cl::sycl::nd_item<1> TestNDItem =
detail::Builder::createNDItem<1>(detail::Builder::createItem<1, false>({4}, {2}),
detail::Builder::createItem<1, false>({2}, {0}),
detail::Builder::createGroup<1>({4}, {2}, {1}));

// CHECK: define {{.*}}dso_local spir_func void @{{.*}}testNDItem{{.*}}(%"class.{{.*}}nd_item"*{{.*}}%TestNDItem)
SYCL_EXTERNAL void testNDItem(nd_item<1> TestNDItem) {
// CHECK: call void @llvm.assume(i1 {{.*}})
int GlobalId = TestNDItem.get_global_id(0);
// CHECK: call void @llvm.assume(i1 {{.*}})
Expand All @@ -39,13 +36,10 @@ int main() {
// CHECK: call void @llvm.assume(i1 {{.*}})
int LocalRange = TestNDItem.get_local_range(0);

// CHECK: call void @llvm.assume(i1 {{.*}})
int GlobalIdConverted = TestNDItem.get_global_id();
// CHECK: call void @llvm.assume(i1 {{.*}})
int LocalIdConverted = TestNDItem.get_local_id();
// CHECK: call void @llvm.assume(i1 {{.*}})
int OffsetConferted = TestNDItem.get_offset();
// CHECK: call void @llvm.assume(i1 {{.*}})

return 0;
}
// CHECK: }