Skip to content

Commit e715eb4

Browse files
intwanghaointwanghao
committed
fix
Signed-off-by: intwanghao <hao3.wang@intel.com>
1 parent 37eddd9 commit e715eb4

6 files changed

Lines changed: 72 additions & 0 deletions

File tree

‎clang/examples/DPCT/CUB/cub$$BlockRadixSort::Sort.cu‎

Lines changed: 10 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,10 @@
1+
// clang-format off
2+
#include <cstddef>
3+
#include <cub/cub.cuh>
4+
5+
__device__ void test(int (&thread_data)[4]) {
6+
// Start
7+
__shared__ typename cub::BlockRadixSort<int, 128, 4>::TempStorage temp_storage;
8+
cub::BlockRadixSort<int, 128, 4>(temp_storage).Sort(thread_data/*int(&)[4]*/);
9+
// End
10+
}

‎clang/examples/DPCT/CUB/cub$$BlockRadixSort::SortBlockedToStriped.cu‎

Lines changed: 10 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,10 @@
1+
// clang-format off
2+
#include <cstddef>
3+
#include <cub/cub.cuh>
4+
5+
__device__ void test(int (&thread_data)[4]) {
6+
// Start
7+
__shared__ typename cub::BlockRadixSort<int, 128, 4>::TempStorage temp_storage;
8+
cub::BlockRadixSort<int, 128, 4>(temp_storage).SortBlockedToStriped(thread_data/*int(&)[4]*/);
9+
// End
10+
}

‎clang/examples/DPCT/CUB/cub$$BlockRadixSort::SortDescending.cu‎

Lines changed: 10 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,10 @@
1+
// clang-format off
2+
#include <cstddef>
3+
#include <cub/cub.cuh>
4+
5+
__device__ void test(int (&thread_data)[4]) {
6+
// Start
7+
__shared__ typename cub::BlockRadixSort<int, 128, 4>::TempStorage temp_storage;
8+
cub::BlockRadixSort<int, 128, 4>(temp_storage).SortDescending(thread_data/*int(&)[4]*/);
9+
// End
10+
}

‎clang/examples/DPCT/CUB/cub$$BlockRadixSort::SortDescendingBlockedToStriped.cu‎

Lines changed: 10 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,10 @@
1+
// clang-format off
2+
#include <cstddef>
3+
#include <cub/cub.cuh>
4+
5+
__device__ void test(int (&thread_data)[4]) {
6+
// Start
7+
__shared__ typename cub::BlockRadixSort<int, 128, 4>::TempStorage temp_storage;
8+
cub::BlockRadixSort<int, 128, 4>(temp_storage).SortDescendingBlockedToStriped(thread_data/*int(&)[4]*/);
9+
// End
10+
}

‎clang/test/dpct/query_api_mapping/CUB/cub_block.cu‎

Lines changed: 28 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -66,3 +66,31 @@
6666
// CHECK_BLOCKSTORE_STORE: auto item_ct1 = sycl::ext::oneapi::this_work_item::get_nd_item<3>();
6767
// CHECK_BLOCKSTORE_STORE: dpct::group::group_store<int, 4, dpct::group::group_store_algorithm::blocked>(temp_storage).store(item_ct1, dst, thread_data);
6868
// CHECK_BLOCKSTORE_STORE: dpct::group::group_store<int, 4, dpct::group::group_store_algorithm::blocked>(temp_storage).store(item_ct1, dst, thread_data, end);
69+
70+
// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cub::BlockRadixSort::Sort | FileCheck %s -check-prefix=CHECK_BLOCKRADIXSORT_SORT
71+
// CHECK_BLOCKRADIXSORT_SORT: CUDA API:
72+
// CHECK_BLOCKRADIXSORT_SORT: __shared__ typename cub::BlockRadixSort<int, 128, 4>::TempStorage temp_storage;
73+
// CHECK_BLOCKRADIXSORT_SORT: cub::BlockRadixSort<int, 128, 4>(temp_storage).Sort(thread_data/*int(&)[4]*/);
74+
// CHECK_BLOCKRADIXSORT_SORT: Is migrated to:
75+
// CHECK_BLOCKRADIXSORT_SORT: dpct::group::group_radix_sort<int, 4>(temp_storage).sort(sycl::ext::oneapi::this_work_item::get_nd_item<3>(), thread_data);
76+
77+
// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cub::BlockRadixSort::SortBlockedToStriped | FileCheck %s -check-prefix=CHECK_BLOCKRADIXSORT_SORTBLOCKEDTOSTRIPED
78+
// CHECK_BLOCKRADIXSORT_SORTBLOCKEDTOSTRIPED: CUDA API:
79+
// CHECK_BLOCKRADIXSORT_SORTBLOCKEDTOSTRIPED: __shared__ typename cub::BlockRadixSort<int, 128, 4>::TempStorage temp_storage;
80+
// CHECK_BLOCKRADIXSORT_SORTBLOCKEDTOSTRIPED: cub::BlockRadixSort<int, 128, 4>(temp_storage).SortBlockedToStriped(thread_data/*int(&)[4]*/);
81+
// CHECK_BLOCKRADIXSORT_SORTBLOCKEDTOSTRIPED: Is migrated to:
82+
// CHECK_BLOCKRADIXSORT_SORTBLOCKEDTOSTRIPED: dpct::group::group_radix_sort<int, 4>(temp_storage).sort_blocked_to_striped(sycl::ext::oneapi::this_work_item::get_nd_item<3>(), thread_data);
83+
84+
// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cub::BlockRadixSort::SortDescending | FileCheck %s -check-prefix=CHECK_BLOCKRADIXSORT_SORTDESCENDING
85+
// CHECK_BLOCKRADIXSORT_SORTDESCENDING: CUDA API:
86+
// CHECK_BLOCKRADIXSORT_SORTDESCENDING: __shared__ typename cub::BlockRadixSort<int, 128, 4>::TempStorage temp_storage;
87+
// CHECK_BLOCKRADIXSORT_SORTDESCENDING: cub::BlockRadixSort<int, 128, 4>(temp_storage).SortDescending(thread_data/*int(&)[4]*/);
88+
// CHECK_BLOCKRADIXSORT_SORTDESCENDING: Is migrated to:
89+
// CHECK_BLOCKRADIXSORT_SORTDESCENDING: dpct::group::group_radix_sort<int, 4>(temp_storage).sort_descending(sycl::ext::oneapi::this_work_item::get_nd_item<3>(), thread_data);
90+
91+
// RUN: dpct --cuda-include-path="%cuda-path/include" --query-api-mapping=cub::BlockRadixSort::SortDescendingBlockedToStriped | FileCheck %s -check-prefix=CHECK_BLOCKRADIXSORT_SORTDESCENDINGBLOCKEDTOSTRIPED
92+
// CHECK_BLOCKRADIXSORT_SORTDESCENDINGBLOCKEDTOSTRIPED: CUDA API:
93+
// CHECK_BLOCKRADIXSORT_SORTDESCENDINGBLOCKEDTOSTRIPED: __shared__ typename cub::BlockRadixSort<int, 128, 4>::TempStorage temp_storage;
94+
// CHECK_BLOCKRADIXSORT_SORTDESCENDINGBLOCKEDTOSTRIPED: cub::BlockRadixSort<int, 128, 4>(temp_storage).SortDescendingBlockedToStriped(thread_data/*int(&)[4]*/);
95+
// CHECK_BLOCKRADIXSORT_SORTDESCENDINGBLOCKEDTOSTRIPED: Is migrated to:
96+
// CHECK_BLOCKRADIXSORT_SORTDESCENDINGBLOCKEDTOSTRIPED: dpct::group::group_radix_sort<int, 4>(temp_storage).sort_descending_blocked_to_striped(sycl::ext::oneapi::this_work_item::get_nd_item<3>(), thread_data);

‎clang/test/dpct/query_api_mapping/test_all.cu‎

Lines changed: 4 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -746,6 +746,10 @@
746746
// CHECK-NEXT: cub::BlockExchange::StripedToBlocked
747747
// CHECK-NEXT: cub::BlockExchange::WarpStripedToBlocked
748748
// CHECK-NEXT: cub::BlockLoad::Load
749+
// CHECK-NEXT: cub::BlockRadixSort::Sort
750+
// CHECK-NEXT: cub::BlockRadixSort::SortBlockedToStriped
751+
// CHECK-NEXT: cub::BlockRadixSort::SortDescending
752+
// CHECK-NEXT: cub::BlockRadixSort::SortDescendingBlockedToStriped
749753
// CHECK-NEXT: cub::BlockStore::Store
750754
// CHECK-NEXT: cub::DeviceReduce::ArgMax
751755
// CHECK-NEXT: cub::DeviceReduce::ArgMin

0 commit comments

Comments
 (0)