Skip to content

Commit 5b3d87c

Browse files
alexbatashevAlexander Batashev
alexbatashev
authored and
Alexander Batashev
committed
[SYCL] Remove usages of program class API (intel#474)
1 parent cfce15b commit 5b3d87c

35 files changed

+269
-89
lines changed

‎SYCL/Basic/get_backend.cpp

Lines changed: 0 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -43,11 +43,6 @@ int main() {
4343
return_fail();
4444
}
4545

46-
program prog(c);
47-
if (prog.get_backend() != plt.get_backend()) {
48-
return_fail();
49-
}
50-
5146
default_selector sel;
5247
queue q(c, sel);
5348
if (q.get_backend() != plt.get_backend()) {

‎SYCL/Basic/kernel_info.cpp

Lines changed: 5 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -23,13 +23,13 @@ int main() {
2323
queue q;
2424

2525
buffer<int, 1> buf(range<1>(1));
26-
program prg(q.get_context());
27-
28-
prg.build_with_kernel_type<class SingleTask>();
29-
assert(prg.has_kernel<class SingleTask>());
30-
kernel krn = prg.get_kernel<class SingleTask>();
26+
auto KernelID = sycl::get_kernel_id<class SingleTask>();
27+
auto KB =
28+
get_kernel_bundle<bundle_state::executable>(q.get_context(), {KernelID});
29+
kernel krn = KB.get_kernel(KernelID);
3130

3231
q.submit([&](handler &cgh) {
32+
cgh.use_kernel_bundle(KB);
3333
auto acc = buf.get_access<access::mode::read_write>(cgh);
3434
cgh.single_task<class SingleTask>(krn, [=]() { acc[0] = acc[0] + 1; });
3535
});
@@ -40,8 +40,6 @@ int main() {
4040
assert(krnArgCount > 0);
4141
const context krnCtx = krn.get_info<info::kernel::context>();
4242
assert(krnCtx == q.get_context());
43-
const program krnPrg = krn.get_info<info::kernel::program>();
44-
assert(krnPrg == prg);
4543
const cl_uint krnRefCount = krn.get_info<info::kernel::reference_count>();
4644
assert(krnRefCount > 0);
4745
const std::string krnAttr = krn.get_info<info::kernel::attributes>();

‎SYCL/SpecConstants/1.2.1/composite-in-functor.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-in-functor.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
// UNSUPPORTED: cuda || rocm
22
//
3-
// RUN: %clangxx -fsycl %s -o %t.out
3+
// RUN: %clangxx -fsycl %s -D__SYCL_INTERNAL_API -o %t.out
44
// RUN: %HOST_RUN_PLACEHOLDER %t.out %HOST_CHECK_PLACEHOLDER
55
// RUN: %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER
66
// RUN: %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER

‎SYCL/SpecConstants/1.2.1/composite-type.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-type.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
// UNSUPPORTED: cuda || rocm
22
//
3-
// RUN: %clangxx -fsycl %s -o %t.out
3+
// RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out
44
// RUN: %HOST_RUN_PLACEHOLDER %t.out
55
// RUN: %CPU_RUN_PLACEHOLDER %t.out
66
// RUN: %GPU_RUN_PLACEHOLDER %t.out

‎SYCL/SpecConstants/1.2.1/multiple-usages-of-composite.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/multiple-usages-of-composite.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
// UNSUPPORTED: cuda || rocm
22
//
3-
// RUN: %clangxx -fsycl %s -o %t.out -v
3+
// RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out -v
44
// RUN: %HOST_RUN_PLACEHOLDER %t.out %HOST_CHECK_PLACEHOLDER
55
// RUN: %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER
66
// RUN: %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER

‎SYCL/SpecConstants/1.2.1/spec_const_hw.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_hw.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
// UNSUPPORTED: cuda || rocm
22
//
3-
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
3+
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -D__SYCL_INTERNAL_API %s -o %t.out
44
// RUN: %HOST_RUN_PLACEHOLDER %t.out
55
// RUN: %CPU_RUN_PLACEHOLDER %t.out
66
// RUN: %GPU_RUN_PLACEHOLDER %t.out

‎SYCL/SpecConstants/1.2.1/spec_const_hw_accelerator.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_hw_accelerator.cpp

File renamed without changes.

‎SYCL/SpecConstants/1.2.1/spec_const_neg.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_neg.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clangxx -fsycl %s -o %t.out
1+
// RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out
22
// RUN: %HOST_RUN_PLACEHOLDER %t.out
33
// RUN: %CPU_RUN_PLACEHOLDER %t.out
44
// RUN: %GPU_RUN_PLACEHOLDER %t.out

‎SYCL/SpecConstants/1.2.1/spec_const_redefine.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_redefine.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -2,7 +2,7 @@
22
//
33
// FIXME Disable fallback assert so that it doesn't interferes with number of
44
// program builds at run-time
5-
// RUN: %clangxx -DSYCL_DISABLE_FALLBACK_ASSERT -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
5+
// RUN: %clangxx -DSYCL_DISABLE_FALLBACK_ASSERT -D__SYCL_INTERNAL_API -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
66
// RUN: %HOST_RUN_PLACEHOLDER %t.out
77
// RUN: env SYCL_PI_TRACE=2 %CPU_RUN_PLACEHOLDER %t.out 2>&1 %CPU_CHECK_PLACEHOLDER
88
// RUN: env SYCL_PI_TRACE=2 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER

‎SYCL/SpecConstants/1.2.1/spec_const_redefine_accelerator.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_redefine_accelerator.cpp

File renamed without changes.

‎SYCL/SpecConstants/1.2.1/specialization_constants.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clangxx -fsycl %s -o %t.out
1+
// RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out
22
// RUN: %HOST_RUN_PLACEHOLDER %t.out
33
// RUN: %CPU_RUN_PLACEHOLDER %t.out
44
// RUN: %GPU_RUN_PLACEHOLDER %t.out

‎SYCL/SpecConstants/1.2.1/specialization_constants_negative.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_negative.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clangxx -fsycl %s -o %t.out
1+
// RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out
22
// RUN: %HOST_RUN_PLACEHOLDER %t.out
33
// RUN: %CPU_RUN_PLACEHOLDER %t.out
44
// RUN: %GPU_RUN_PLACEHOLDER %t.out

‎SYCL/SpecConstants/1.2.1/specialization_constants_override.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_override.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clangxx -fsycl %s -o %t.out
1+
// RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out
22
// RUN: %HOST_RUN_PLACEHOLDER %t.out
33
// RUN: %CPU_RUN_PLACEHOLDER %t.out
44
// RUN: %GPU_RUN_PLACEHOLDER %t.out

‎SYCL/SpecConstants/1.2.1/unpacked-composite-type.cpp renamed to ‎SYCL/DeprecatedFeatures/SpecConsts1.2.1/unpacked-composite-type.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
// UNSUPPORTED: cuda || rocm
22
//
3-
// RUN: %clangxx -fsycl %s -o %t.out
3+
// RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out
44
// RUN: %HOST_RUN_PLACEHOLDER %t.out %HOST_CHECK_PLACEHOLDER
55
// RUN: %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER
66
// RUN: %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER

‎SYCL/KernelAndProgram/basic-program.cpp renamed to ‎SYCL/DeprecatedFeatures/basic-program.cpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1,5 +1,5 @@
1-
// XFAIL: cuda
2-
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
1+
// XFAIL: cuda || hip
2+
// RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
33
// RUN: %CPU_RUN_PLACEHOLDER %t.out
44
// RUN: %GPU_RUN_PLACEHOLDER %t.out
55
// RUN: %ACC_RUN_PLACEHOLDER %t.out

‎SYCL/KernelAndProgram/basic.cpp renamed to ‎SYCL/DeprecatedFeatures/basic.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,7 +1,7 @@
11
// No JITing for host devices.
22
// REQUIRES: opencl || level_zero
33
// RUN: rm -rf %t/cache_dir
4-
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
4+
// RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
55
// RUN: env SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD
66
// RUN: env SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE
77
// RUN: env SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD

‎SYCL/KernelAndProgram/basic.hpp renamed to ‎SYCL/DeprecatedFeatures/basic.hpp

File renamed without changes.

‎SYCL/KernelAndProgram/get-options.cpp renamed to ‎SYCL/DeprecatedFeatures/get-options.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clangxx -fsycl %s -o %t.out
1+
// RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl %s -o %t.out
22
// RUN: %HOST_RUN_PLACEHOLDER %t.out
33
// RUN: %CPU_RUN_PLACEHOLDER %t.out
44
// RUN: %GPU_RUN_PLACEHOLDER %t.out

‎SYCL/DeprecatedFeatures/get_backend.cpp

Lines changed: 71 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,71 @@
1+
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -D__SYCL_INTERNAL_API %s -o %t.out
2+
// RUN: env SYCL_DEVICE_FILTER=%sycl_be %t.out
3+
//
4+
//==----------------- get_backend.cpp ------------------------==//
5+
// This is a test of get_backend().
6+
// Do not set SYCL_DEVICE_FILTER. We do not want the preferred
7+
// backend.
8+
//==----------------------------------------------------------==//
9+
10+
#include <CL/sycl.hpp>
11+
#include <CL/sycl/backend_types.hpp>
12+
#include <iostream>
13+
14+
using namespace cl::sycl;
15+
16+
bool check(backend be) {
17+
switch (be) {
18+
case backend::opencl:
19+
case backend::level_zero:
20+
case backend::cuda:
21+
case backend::host:
22+
return true;
23+
default:
24+
return false;
25+
}
26+
return false;
27+
}
28+
29+
inline void return_fail() {
30+
std::cout << "Failed" << std::endl;
31+
exit(1);
32+
}
33+
34+
int main() {
35+
for (const auto &plt : platform::get_platforms()) {
36+
if (!plt.is_host()) {
37+
if (check(plt.get_backend()) == false) {
38+
return_fail();
39+
}
40+
41+
context c(plt);
42+
if (c.get_backend() != plt.get_backend()) {
43+
return_fail();
44+
}
45+
46+
program prog(c);
47+
if (prog.get_backend() != plt.get_backend()) {
48+
return_fail();
49+
}
50+
51+
default_selector sel;
52+
queue q(c, sel);
53+
if (q.get_backend() != plt.get_backend()) {
54+
return_fail();
55+
}
56+
57+
auto device = q.get_device();
58+
if (device.get_backend() != plt.get_backend()) {
59+
return_fail();
60+
}
61+
62+
unsigned char *HostAlloc = (unsigned char *)malloc_host(1, c);
63+
auto e = q.memset(HostAlloc, 42, 1);
64+
if (e.get_backend() != plt.get_backend()) {
65+
return_fail();
66+
}
67+
}
68+
}
69+
std::cout << "Passed" << std::endl;
70+
return 0;
71+
}

‎SYCL/KernelAndProgram/kernel-and-program.cpp renamed to ‎SYCL/DeprecatedFeatures/kernel-and-program.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
1+
// RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
22
// RUN: %HOST_RUN_PLACEHOLDER %t.out
33
// RUN: %CPU_RUN_PLACEHOLDER %t.out
44
// RUN: %GPU_RUN_PLACEHOLDER %t.out

‎SYCL/DeprecatedFeatures/kernel_info.cpp

Lines changed: 65 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,65 @@
1+
// RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
2+
// RUN: %CPU_RUN_PLACEHOLDER %t.out
3+
// RUN: %GPU_RUN_PLACEHOLDER %t.out
4+
// RUN: %ACC_RUN_PLACEHOLDER %t.out
5+
//
6+
// Fail is flaky for level_zero, enable when fixed.
7+
// UNSUPPORTED: level_zero
8+
9+
//==--- kernel_info.cpp - SYCL kernel info test ----------------------------==//
10+
//
11+
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
12+
// See https://llvm.org/LICENSE.txt for license information.
13+
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
14+
//
15+
//===----------------------------------------------------------------------===//
16+
17+
#include <CL/sycl.hpp>
18+
#include <cassert>
19+
20+
using namespace cl::sycl;
21+
22+
int main() {
23+
queue q;
24+
25+
buffer<int, 1> buf(range<1>(1));
26+
program prg(q.get_context());
27+
28+
prg.build_with_kernel_type<class SingleTask>();
29+
assert(prg.has_kernel<class SingleTask>());
30+
kernel krn = prg.get_kernel<class SingleTask>();
31+
32+
q.submit([&](handler &cgh) {
33+
auto acc = buf.get_access<access::mode::read_write>(cgh);
34+
cgh.single_task<class SingleTask>(krn, [=]() { acc[0] = acc[0] + 1; });
35+
});
36+
37+
const std::string krnName = krn.get_info<info::kernel::function_name>();
38+
assert(!krnName.empty());
39+
const cl_uint krnArgCount = krn.get_info<info::kernel::num_args>();
40+
assert(krnArgCount > 0);
41+
const context krnCtx = krn.get_info<info::kernel::context>();
42+
assert(krnCtx == q.get_context());
43+
const program krnPrg = krn.get_info<info::kernel::program>();
44+
assert(krnPrg == prg);
45+
const cl_uint krnRefCount = krn.get_info<info::kernel::reference_count>();
46+
assert(krnRefCount > 0);
47+
const std::string krnAttr = krn.get_info<info::kernel::attributes>();
48+
assert(krnAttr.empty());
49+
50+
device dev = q.get_device();
51+
const size_t wgSize =
52+
krn.get_work_group_info<info::kernel_work_group::work_group_size>(dev);
53+
assert(wgSize > 0);
54+
const size_t wgSizeNew =
55+
krn.get_info<info::kernel_device_specific::work_group_size>(dev);
56+
assert(wgSizeNew > 0);
57+
assert(wgSize == wgSizeNew);
58+
const size_t prefWGSizeMult = krn.get_work_group_info<
59+
info::kernel_work_group::preferred_work_group_size_multiple>(dev);
60+
assert(prefWGSizeMult > 0);
61+
const size_t prefWGSizeMultNew = krn.get_info<
62+
info::kernel_device_specific::preferred_work_group_size_multiple>(dev);
63+
assert(prefWGSizeMultNew > 0);
64+
assert(prefWGSizeMult == prefWGSizeMultNew);
65+
}

‎SYCL/Basic/parallel_for_range.cpp renamed to ‎SYCL/DeprecatedFeatures/parallel_for_range.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -2,7 +2,7 @@
22
// UNSUPPORTED: windows
33
// Level0 testing times out on Windows.
44

5-
// RUN: %clangxx -fsycl -fno-sycl-id-queries-fit-in-int -fsycl-targets=%sycl_triple %s -o %t.out
5+
// RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl -fno-sycl-id-queries-fit-in-int -fsycl-targets=%sycl_triple %s -o %t.out
66
// RUN: %CPU_RUN_PLACEHOLDER %t.out
77
// RUN: %GPU_RUN_PLACEHOLDER %t.out
88
// RUN: %ACC_RUN_PLACEHOLDER %t.out

‎SYCL/Config/program_link.cpp renamed to ‎SYCL/DeprecatedFeatures/program_link.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
1+
// RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
22
// RUN: %CPU_RUN_PLACEHOLDER SYCL_PI_TRACE=-1 SYCL_PROGRAM_LINK_OPTIONS="-cl-fast-relaxed-math" %t.out %CPU_CHECK_PLACEHOLDER --check-prefix=CHECK-IS-RELAXED-MATH
33
// RUN: %GPU_RUN_PLACEHOLDER SYCL_PI_TRACE=-1 SYCL_PROGRAM_LINK_OPTIONS="-cl-fast-relaxed-math" %t.out %GPU_CHECK_PLACEHOLDER --check-prefix=CHECK-IS-RELAXED-MATH
44
// RUN: %ACC_RUN_PLACEHOLDER SYCL_PI_TRACE=-1 SYCL_PROGRAM_LINK_OPTIONS="-cl-fast-relaxed-math" %t.out %ACC_CHECK_PLACEHOLDER --check-prefix=CHECK-IS-RELAXED-MATH

‎SYCL/KernelAndProgram/spec_consts.cpp renamed to ‎SYCL/DeprecatedFeatures/spec_consts.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -3,7 +3,7 @@
33
// REQUIRES: opencl || level_zero
44
// RUN: rm -rf %t/cache_dir
55
// FIXME Temporary disable fallback assert here until fixed
6-
// RUN: %clangxx -fsycl -DSYCL_DISABLE_FALLBACK_ASSERT=1 -fsycl-targets=%sycl_triple %s -o %t.out
6+
// RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl -DSYCL_DISABLE_FALLBACK_ASSERT=1 -fsycl-targets=%sycl_triple %s -o %t.out
77
// RUN: env SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD
88
// RUN: env SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-CACHE
99
// RUN: env SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_DIR=%t/cache_dir SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK-BUILD

‎SYCL/KernelAndProgram/spec_consts.hpp renamed to ‎SYCL/DeprecatedFeatures/spec_consts.hpp

File renamed without changes.

‎SYCL/DeviceCodeSplit/Inputs/split-per-source-second-file.cpp

Lines changed: 9 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -4,15 +4,19 @@ void runKernelsFromFile2() {
44
cl::sycl::queue Q;
55
int Data = 0;
66
{
7-
cl::sycl::program Prg(Q.get_context());
87
cl::sycl::buffer<int, 1> Buf(&Data, cl::sycl::range<1>(1));
9-
Prg.build_with_kernel_type<File2Kern1>();
10-
cl::sycl::kernel Krn = Prg.get_kernel<File2Kern1>();
8+
auto KernelID1 = sycl::get_kernel_id<File2Kern1>();
9+
auto KernelID2 = sycl::get_kernel_id<File1Kern1>();
10+
auto KernelID3 = sycl::get_kernel_id<File1Kern2>();
11+
auto KB = sycl::get_kernel_bundle<sycl::bundle_state::executable>(
12+
Q.get_context(), {KernelID1});
13+
auto Krn = KB.get_kernel(KernelID1);
1114

12-
assert(!Prg.has_kernel<File1Kern1>());
13-
assert(!Prg.has_kernel<File1Kern2>());
15+
assert(!KB.has_kernel(KernelID2));
16+
assert(!KB.has_kernel(KernelID3));
1417

1518
Q.submit([&](cl::sycl::handler &Cgh) {
19+
Cgh.use_kernel_bundle(KB);
1620
auto Acc = Buf.get_access<cl::sycl::access::mode::read_write>(Cgh);
1721
Cgh.single_task<File2Kern1>(Krn, [=]() { Acc[0] = 3; });
1822
});

0 commit comments

Comments
 (0)