Skip to content

Commit 2260094

Browse files
krisrakpraveenkk123
krisrak
authored and
praveenkk123
committed
removed ext::oneapi namespace in Jupyter DPC++ Reduction code samples (oneapi-src#779)
* removed ext::oneapi namespace in Jupyter DPC++ Reduction code samples * updated Makefile with Reduction cpp file name changes
1 parent 9953d9f commit 2260094

14 files changed

Lines changed: 34 additions & 51 deletions

‎DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/Reductions.ipynb‎

Lines changed: 20 additions & 31 deletions
Original file line numberDiff line numberDiff line change
@@ -433,10 +433,10 @@
433433
"\n",
434434
"It is common for parallel kernels to produce a single output resulting from some combination of all inputs (e.g. the sum). Writing efficient reductions is a complex task, depending on both device and runtime characteristics. Providing an abstraction for reductions in SYCL would greatly improve programmer productivity.\n",
435435
"\n",
436-
"`sycl::ext::oneapi::reduction` object in parallel_for encapsulates the reduction variable, an optional operator identity and the reduction operator as shown below:\n",
436+
"`sycl::reduction` object in parallel_for encapsulates the reduction variable, an optional operator identity and the reduction operator as shown below:\n",
437437
"\n",
438438
"```cpp\n",
439-
" q.parallel_for(nd_range<1>{N, B}, sycl::ext::oneapi::reduction(sum, 0, sycl::plus<>()), [=](nd_item<1> it, auto& temp) {\n",
439+
" q.parallel_for(nd_range<1>{N, B}, sycl::reduction(sum, sycl::plus<>()), [=](nd_item<1> it, auto& temp) {\n",
440440
" int i = it.get_global_id(0);\n",
441441
" temp.combine(data[i]);\n",
442442
" });\n",
@@ -454,7 +454,7 @@
454454
"cell_type": "markdown",
455455
"metadata": {},
456456
"source": [
457-
"The code below uses __sycl::ext::oneapi::reduction__ object in _parallel_for_ to compute the reduction with just one kernel using Unified Shared Memory(USM) for memory management.\n",
457+
"The code below uses __sycl::reduction__ object in _parallel_for_ to compute the reduction with just one kernel using Unified Shared Memory(USM) for memory management.\n",
458458
"\n",
459459
"The DPC++ code below demonstrates reduction in parallel_for with USM: Inspect code, there are no modifications necessary:\n",
460460
"\n",
@@ -469,7 +469,7 @@
469469
"metadata": {},
470470
"outputs": [],
471471
"source": [
472-
"%%writefile lab/sum_oneapi_reduction_usm.cpp\n",
472+
"%%writefile lab/sum_reduction_usm.cpp\n",
473473
"//==============================================================\n",
474474
"// Copyright © 2020 Intel Corporation\n",
475475
"//\n",
@@ -496,7 +496,7 @@
496496
" *sum = 0;\n",
497497
"\n",
498498
" //# nd-range kernel parallel_for with reduction parameter\n",
499-
" q.parallel_for(nd_range<1>{N, B}, ext::oneapi::reduction(sum, 0, plus<>()), [=](nd_item<1> it, auto& temp) {\n",
499+
" q.parallel_for(nd_range<1>{N, B}, reduction(sum, plus<>()), [=](nd_item<1> it, auto& temp) {\n",
500500
" auto i = it.get_global_id(0);\n",
501501
" temp.combine(data[i]);\n",
502502
" }).wait();\n",
@@ -523,7 +523,7 @@
523523
"metadata": {},
524524
"outputs": [],
525525
"source": [
526-
"! chmod 755 q; chmod 755 run_sum_oneapi_reduction_usm.sh; if [ -x \"$(command -v qsub)\" ]; then ./q run_sum_oneapi_reduction_usm.sh; else ./run_sum_oneapi_reduction_usm.sh; fi"
526+
"! chmod 755 q; chmod 755 run_sum_reduction_usm.sh; if [ -x \"$(command -v qsub)\" ]; then ./q run_sum_reduction_usm.sh; else ./run_sum_reduction_usm.sh; fi"
527527
]
528528
},
529529
{
@@ -545,7 +545,7 @@
545545
"cell_type": "markdown",
546546
"metadata": {},
547547
"source": [
548-
"The code below uses __sycl::ext::oneapi::reduction__ object in _parallel_for_ to compute the reduction with just one kernel using SYCL buffers and accessors for memory management.\n",
548+
"The code below uses __sycl::reduction__ object in _parallel_for_ to compute the reduction with just one kernel using SYCL buffers and accessors for memory management.\n",
549549
"\n",
550550
"The DPC++ code below demonstrates reduction in parallel_for with Buffers: Inspect code, there are no modifications necessary:\n",
551551
"\n",
@@ -560,7 +560,7 @@
560560
"metadata": {},
561561
"outputs": [],
562562
"source": [
563-
"%%writefile lab/sum_oneapi_reduction_buffers.cpp\n",
563+
"%%writefile lab/sum_reduction_buffers.cpp\n",
564564
"//==============================================================\n",
565565
"// Copyright © 2020 Intel Corporation\n",
566566
"//\n",
@@ -586,12 +586,11 @@
586586
" buffer buf_sum(&sum, range(1));\n",
587587
"\n",
588588
" q.submit([&](handler& h) {\n",
589-
" //# create accessors for buffers\n",
589+
" //# create accessors for buffer\n",
590590
" accessor acc_data(buf_data, h, read_only);\n",
591-
" accessor acc_sum(buf_sum, h);\n",
592591
"\n",
593592
" //# nd-range kernel parallel_for with reduction parameter\n",
594-
" h.parallel_for(nd_range<1>{N, B}, ext::oneapi::reduction(acc_sum, 0, plus<>()), [=](nd_item<1> it, auto& temp) {\n",
593+
" h.parallel_for(nd_range<1>{N, B}, reduction(buf_sum, h, plus<>()), [=](nd_item<1> it, auto& temp) {\n",
595594
" auto i = it.get_global_id(0);\n",
596595
" temp.combine(acc_data[i]);\n",
597596
" });\n",
@@ -617,7 +616,7 @@
617616
"metadata": {},
618617
"outputs": [],
619618
"source": [
620-
"! chmod 755 q; chmod 755 run_sum_oneapi_reduction_buffers.sh; if [ -x \"$(command -v qsub)\" ]; then ./q run_sum_oneapi_reduction_buffers.sh; else ./run_sum_oneapi_reduction_buffers.sh; fi"
619+
"! chmod 755 q; chmod 755 run_sum_reduction_buffers.sh; if [ -x \"$(command -v qsub)\" ]; then ./q run_sum_reduction_buffers.sh; else ./run_sum_reduction_buffers.sh; fi"
621620
]
622621
},
623622
{
@@ -639,7 +638,7 @@
639638
"cell_type": "markdown",
640639
"metadata": {},
641640
"source": [
642-
"The code below uses multiple __sycl::ext::oneapi::reduction__ objects in _parallel_for_ to compute the reductions with just one kernel using SYCL buffers and accessors for memory management.\n",
641+
"The code below uses multiple __sycl::reduction__ objects in _parallel_for_ to compute the reductions with just one kernel using SYCL buffers and accessors for memory management.\n",
643642
"\n",
644643
"Multiple reductions are also supported with just one kernel, the code snippet below shows how to definne a kernel using parallel_for with multiple reduction objects:\n",
645644
"\n",
@@ -693,14 +692,11 @@
693692
" q.submit([&](handler& h) {\n",
694693
" //# create accessors for data and results\n",
695694
" accessor acc_data(buf_data, h, read_only);\n",
696-
" accessor acc_sum(buf_sum, h);\n",
697-
" accessor acc_min(buf_min, h);\n",
698-
" accessor acc_max(buf_max, h);\n",
699695
" \n",
700696
" //# define reduction objects for sum, min, max reduction\n",
701-
" auto reduction_sum = ext::oneapi::reduction(acc_sum, 0, plus<>());\n",
702-
" auto reduction_min = ext::oneapi::reduction(acc_min, 0, minimum<>());\n",
703-
" auto reduction_max = ext::oneapi::reduction(acc_max, 0, maximum<>());\n",
697+
" auto reduction_sum = reduction(buf_sum, h, plus<>());\n",
698+
" auto reduction_min = reduction(buf_min, h, minimum<>());\n",
699+
" auto reduction_max = reduction(buf_max, h, maximum<>());\n",
704700
" \n",
705701
" //# parallel_for with multiple reduction objects\n",
706702
" h.parallel_for(nd_range<1>{N, B}, reduction_sum, reduction_min, reduction_max, [=](nd_item<1> it, auto& temp_sum, auto& temp_min, auto& temp_max) {\n",
@@ -760,7 +756,7 @@
760756
"cell_type": "markdown",
761757
"metadata": {},
762758
"source": [
763-
"The code below uses __sycl::ext::oneapi::reduction__ object in _parallel_for_ to compute the reduction object that uses a custom operator to find minumum value and index.\n",
759+
"The code below uses __sycl::reduction__ object in _parallel_for_ to compute the reduction object that uses a custom operator to find minumum value and index.\n",
764760
"\n",
765761
"The DPC++ code below demonstrates reduction in parallel_for with custom user defined operator to perform reduction: Inspect code, there are no modifications necessary:\n",
766762
"\n",
@@ -775,7 +771,7 @@
775771
"metadata": {},
776772
"outputs": [],
777773
"source": [
778-
"%%writefile lab/oneapi_reduction_custom_operator.cpp\n",
774+
"%%writefile lab/reduction_custom_operator.cpp\n",
779775
"//==============================================================\n",
780776
"// Copyright © 2020 Intel Corporation\n",
781777
"//\n",
@@ -815,7 +811,7 @@
815811
" //# custom operator for reduction to find minumum and index\n",
816812
" pair<int, int> operator_identity = {std::numeric_limits<int>::max(), std::numeric_limits<int>::min()};\n",
817813
" *result = operator_identity;\n",
818-
" auto reduction_object = ext::oneapi::reduction(result, operator_identity, minimum<pair<int, int>>());\n",
814+
" auto reduction_object = reduction(result, operator_identity, minimum<pair<int, int>>());\n",
819815
"\n",
820816
" //# parallel_for with user defined reduction object\n",
821817
" q.parallel_for(nd_range<1>{N, B}, reduction_object, [=](nd_item<1> item, auto& temp) {\n",
@@ -845,7 +841,7 @@
845841
"metadata": {},
846842
"outputs": [],
847843
"source": [
848-
"! chmod 755 q; chmod 755 run_oneapi_reduction_custom_operator.sh; if [ -x \"$(command -v qsub)\" ]; then ./q run_oneapi_reduction_custom_operator.sh; else ./run_oneapi_reduction_custom_operator.sh; fi"
844+
"! chmod 755 q; chmod 755 run_reduction_custom_operator.sh; if [ -x \"$(command -v qsub)\" ]; then ./q run_reduction_custom_operator.sh; else ./run_reduction_custom_operator.sh; fi"
849845
]
850846
},
851847
{
@@ -867,15 +863,8 @@
867863
"cell_type": "markdown",
868864
"metadata": {},
869865
"source": [
870-
"`sycl::reduce_over_group` function for sub_group/work_group and `sycl::ext::oneapi::reduction` in parallel_for helps to optimize and simplify reduction computation in DPC++"
866+
"`sycl::reduce_over_group` function for sub_group/work_group and `sycl::reduction` in parallel_for helps to optimize and simplify reduction computation in DPC++"
871867
]
872-
},
873-
{
874-
"cell_type": "code",
875-
"execution_count": null,
876-
"metadata": {},
877-
"outputs": [],
878-
"source": []
879868
}
880869
],
881870
"metadata": {

‎DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/assets/hwmapping.png‎

-49.1 KB
Loading

‎DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/lab/multiple_reductions_buffers.cpp‎

Lines changed: 3 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -28,14 +28,11 @@ int main() {
2828
q.submit([&](handler& h) {
2929
//# create accessors for data and results
3030
accessor acc_data(buf_data, h, read_only);
31-
accessor acc_sum(buf_sum, h);
32-
accessor acc_min(buf_min, h);
33-
accessor acc_max(buf_max, h);
3431

3532
//# define reduction objects for sum, min, max reduction
36-
auto reduction_sum = ext::oneapi::reduction(acc_sum, 0, plus<>());
37-
auto reduction_min = ext::oneapi::reduction(acc_min, 0, minimum<>());
38-
auto reduction_max = ext::oneapi::reduction(acc_max, 0, maximum<>());
33+
auto reduction_sum = reduction(buf_sum, h, plus<>());
34+
auto reduction_min = reduction(buf_min, h, minimum<>());
35+
auto reduction_max = reduction(buf_max, h, maximum<>());
3936

4037
//# parallel_for with multiple reduction objects
4138
h.parallel_for(nd_range<1>{N, B}, reduction_sum, reduction_min, reduction_max, [=](nd_item<1> it, auto& temp_sum, auto& temp_min, auto& temp_max) {

DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/lab/oneapi_reduction_custom_operator.cpp renamed to DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/lab/reduction_custom_operator.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -37,7 +37,7 @@ int main() {
3737
//# custom operator for reduction to find minumum and index
3838
pair<int, int> operator_identity = {std::numeric_limits<int>::max(), std::numeric_limits<int>::min()};
3939
*result = operator_identity;
40-
auto reduction_object = ext::oneapi::reduction(result, operator_identity, minimum<pair<int, int>>());
40+
auto reduction_object = reduction(result, operator_identity, minimum<pair<int, int>>());
4141

4242
//# parallel_for with user defined reduction object
4343
q.parallel_for(nd_range<1>{N, B}, reduction_object, [=](nd_item<1> item, auto& temp) {

DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/lab/sum_oneapi_reduction_buffers.cpp renamed to DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/lab/sum_reduction_buffers.cpp

File renamed without changes.

DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/lab/sum_oneapi_reduction_usm.cpp renamed to DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/lab/sum_reduction_usm.cpp

File renamed without changes.

DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/run_oneapi_reduction_custom_operator.sh renamed to DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/run_reduction_custom_operator.sh

File renamed without changes.

DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/run_sum_oneapi_reduction_buffers.sh renamed to DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/run_sum_reduction_buffers.sh

File renamed without changes.

DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/run_sum_oneapi_reduction_usm.sh renamed to DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/run_sum_reduction_usm.sh

File renamed without changes.

‎DirectProgramming/DPC++/Jupyter/oneapi-essentials-training/08_DPCPP_Reduction/src/multiple_reductions_buffers.cpp‎

Lines changed: 3 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -28,14 +28,11 @@ int main() {
2828
q.submit([&](handler& h) {
2929
//# create accessors for data and results
3030
accessor acc_data(buf_data, h, read_only);
31-
accessor acc_sum(buf_sum, h);
32-
accessor acc_min(buf_min, h);
33-
accessor acc_max(buf_max, h);
3431

3532
//# define reduction objects for sum, min, max reduction
36-
auto reduction_sum = ext::oneapi::reduction(acc_sum, 0, plus<>());
37-
auto reduction_min = ext::oneapi::reduction(acc_min, 0, minimum<>());
38-
auto reduction_max = ext::oneapi::reduction(acc_max, 0, maximum<>());
33+
auto reduction_sum = reduction(buf_sum, h, plus<>());
34+
auto reduction_min = reduction(buf_min, h, minimum<>());
35+
auto reduction_max = reduction(buf_max, h, maximum<>());
3936

4037
//# parallel_for with multiple reduction objects
4138
h.parallel_for(nd_range<1>{N, B}, reduction_sum, reduction_min, reduction_max, [=](nd_item<1> it, auto& temp_sum, auto& temp_min, auto& temp_max) {

0 commit comments

Comments
 (0)