Skip to content

Commit 7372de4

Browse files
TejaX-AlaghariTejaX-Alaghari
committed
Refactored & added lit tests
1 parent a619d05 commit 7372de4

9 files changed

Lines changed: 2769 additions & 155 deletions

File tree

‎clang/test/dpct/nvshmem.cu‎

Lines changed: 0 additions & 155 deletions
This file was deleted.

‎clang/test/dpct/nvshmem/coll_ops.cu‎

Lines changed: 16 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,16 @@
1+
// REQUIRES: system-linux
2+
// UNSUPPORTED: cuda-8.0, cuda-9.0, cuda-9.1, cuda-9.2, cuda-10.0, cuda-10.1
3+
// RUN: dpct --format-range=none -out-root %T/nvshmem %s --cuda-include-path="%cuda-path/include"
4+
// RUN: FileCheck %s --match-full-lines --input-file %T/nvshmem/coll_ops.dp.cpp
5+
// RUN: %if build_lit %{icpx -c -fsycl -DNO_BUILD_TEST %T/nvshmem/coll_ops.dp.cpp -o %T/nvshmem/coll_ops.dp.o %}
6+
#include <nvshmem.h>
7+
#include <nvshmemx.h>
8+
9+
int main() {
10+
cudaStream_t stream = 0;
11+
12+
// CHECK: ishmemx_barrier_all_on_queue(*stream);
13+
nvshmemx_barrier_all_on_stream(stream);
14+
15+
return 0;
16+
}

‎clang/test/dpct/nvshmem/library_setup_exit_query.cu‎

Lines changed: 104 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,104 @@
1+
// REQUIRES: system-linux
2+
// UNSUPPORTED: cuda-8.0, cuda-9.0, cuda-9.1, cuda-9.2, cuda-10.0, cuda-10.1
3+
// RUN: dpct --format-range=none -out-root %T/nvshmem %s --cuda-include-path="%cuda-path/include"
4+
// RUN: FileCheck %s --match-full-lines --input-file %T/nvshmem/library_setup_exit_query.dp.cpp
5+
// RUN: %if build_lit %{icpx -c -fsycl -DNO_BUILD_TEST %T/nvshmem/library_setup_exit_query.dp.cpp -o %T/nvshmem/library_setup_exit_query.dp.o %}
6+
#include <nvshmem.h>
7+
#include <nvshmemx.h>
8+
9+
int main() {
10+
// CHECK: ishmemx_attr_t attr;
11+
nvshmemx_init_attr_t attr;
12+
#ifndef NO_BUILD_TEST
13+
// CHECK: /*
14+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_init_args_t is not supported.
15+
// CHECK-NEXT: */
16+
nvshmemx_init_args_t args;
17+
// CHECK: /*
18+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_uniqueid_t is not supported.
19+
// CHECK-NEXT: */
20+
nvshmemx_uniqueid_t uid;
21+
// CHECK: /*
22+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_uniqueid_args_t is not supported.
23+
// CHECK-NEXT: */
24+
nvshmemx_uniqueid_args_t uid_args;
25+
26+
// CHECK: /*
27+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_uniqueid_t::internal is not supported.
28+
// CHECK-NEXT: */
29+
char *internal = uid.internal;
30+
31+
// CHECK: /*
32+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_uniqueid_args_t::id is not supported.
33+
// CHECK-NEXT: */
34+
uid_args.id = &uid;
35+
// CHECK: /*
36+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_uniqueid_args_t::myrank is not supported.
37+
// CHECK-NEXT: */
38+
uid_args.myrank = 0;
39+
// CHECK: /*
40+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_uniqueid_args_t::nranks is not supported.
41+
// CHECK-NEXT: */
42+
uid_args.nranks = 0;
43+
44+
// CHECK: /*
45+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_init_args_t::uid_args is not supported.
46+
// CHECK-NEXT: */
47+
args.uid_args = uid_args;
48+
// CHECK: /*
49+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_init_args_t::content is not supported.
50+
// CHECK-NEXT: */
51+
char *content = args.content;
52+
53+
attr.mpi_comm = NULL;
54+
// CHECK: /*
55+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_init_attr_t::args is not supported.
56+
// CHECK-NEXT: */
57+
attr.args = args;
58+
#endif // NO_BUILD_TEST
59+
60+
// CHECK: (&attr)->runtime = ISHMEMX_RUNTIME_MPI;
61+
// CHECK-NEXT: ishmemx_init_attr(&attr);
62+
nvshmemx_init_attr(NVSHMEMX_INIT_WITH_MPI_COMM, &attr);
63+
64+
#ifndef NO_BUILD_TEST
65+
// CHECK: /*
66+
// CHECK-NEXT: DPCT1007:{{[0-9]+}}: Migration of nvshmemx_init_attr is not supported.
67+
// CHECK-NEXT: */
68+
nvshmemx_init_attr(NVSHMEMX_INIT_WITH_MPI_COMM | NVSHMEMX_INIT_WITH_SHMEM, &attr);
69+
#endif // NO_BUILD_TEST
70+
71+
// CHECK: unsigned int rt = ISHMEMX_RUNTIME_MPI;
72+
// CHECK-NEXT: (&attr)->runtime = static_cast<ishmemx_runtime_type_t>(rt);
73+
// CHECK-NEXT: ishmemx_init_attr(&attr);
74+
unsigned int rt = NVSHMEMX_INIT_WITH_MPI_COMM;
75+
nvshmemx_init_attr(rt, &attr);
76+
77+
// CHECK: rt = ISHMEMX_RUNTIME_OPENSHMEM;
78+
// CHECK-NEXT: (&attr)->runtime = static_cast<ishmemx_runtime_type_t>(rt);
79+
// CHECK-NEXT: ishmemx_init_attr(&attr);
80+
rt = NVSHMEMX_INIT_WITH_SHMEM;
81+
nvshmemx_init_attr(rt, &attr);
82+
83+
// CHECK: ishmem_init();
84+
nvshmemx_init_attr(0, &attr);
85+
86+
// CHECK: ishmem_init();
87+
nvshmem_init();
88+
89+
// CHECK: int my_pe = ishmem_my_pe();
90+
// CHECK-NEXT: int n_pes = ishmem_n_pes();
91+
int my_pe = nvshmem_my_pe();
92+
int n_pes = nvshmem_n_pes();
93+
94+
int major, minor;
95+
// CHECK: ishmem_info_get_version(&major, &minor);
96+
nvshmem_info_get_version(&major, &minor);
97+
98+
char* name;
99+
// CHECK: ishmem_info_get_name(name);
100+
nvshmem_info_get_name(name);
101+
102+
// CHECK: ishmem_finalize();
103+
nvshmem_finalize();
104+
}

‎clang/test/dpct/nvshmem/mem_mgmt.cu‎

Lines changed: 24 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,24 @@
1+
// REQUIRES: system-linux
2+
// UNSUPPORTED: cuda-8.0, cuda-9.0, cuda-9.1, cuda-9.2, cuda-10.0, cuda-10.1
3+
// RUN: dpct --format-range=none -out-root %T/nvshmem %s --cuda-include-path="%cuda-path/include"
4+
// RUN: FileCheck %s --match-full-lines --input-file %T/nvshmem/mem_mgmt.dp.cpp
5+
// RUN: %if build_lit %{icpx -c -fsycl -DNO_BUILD_TEST %T/nvshmem/mem_mgmt.dp.cpp -o %T/nvshmem/mem_mgmt.dp.o %}
6+
#include <nvshmem.h>
7+
#include <nvshmemx.h>
8+
9+
int main() {
10+
int my_pe = 0;
11+
size_t size = 1024;
12+
13+
// CHECK: int* array = (int*)ishmem_malloc(size * sizeof(int));
14+
// CHECK-NEXT: array = (int*)ishmem_align(64, size * sizeof(int));
15+
// CHECK-NEXT: array = (int*)ishmem_calloc(size, sizeof(int));
16+
// CHECK-NEXT: void* symmetric_ptr = ishmem_ptr(array, my_pe);
17+
int* array = (int*)nvshmem_malloc(size * sizeof(int));
18+
array = (int*)nvshmem_align(64, size * sizeof(int));
19+
array = (int*)nvshmem_calloc(size, sizeof(int));
20+
void* symmetric_ptr = nvshmem_ptr(array, my_pe);
21+
22+
// CHECK: ishmem_free(array);
23+
nvshmem_free(array);
24+
}

‎clang/test/dpct/nvshmem/mem_order.cu‎

Lines changed: 30 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,30 @@
1+
// REQUIRES: system-linux
2+
// UNSUPPORTED: cuda-8.0, cuda-9.0, cuda-9.1, cuda-9.2, cuda-10.0, cuda-10.1
3+
// RUN: dpct --format-range=none -out-root %T/nvshmem %s --cuda-include-path="%cuda-path/include"
4+
// RUN: FileCheck %s --match-full-lines --input-file %T/nvshmem/mem_order.dp.cpp
5+
// RUN: %if build_lit %{icpx -c -fsycl -DNO_BUILD_TEST %T/nvshmem/mem_order.dp.cpp -o %T/nvshmem/mem_order.dp.o %}
6+
#include <nvshmem.h>
7+
#include <nvshmemx.h>
8+
9+
__host__ __device__ void test() {
10+
// CHECK: ishmem_fence();
11+
nvshmem_fence();
12+
13+
// CHECK: ishmem_quiet();
14+
nvshmem_quiet();
15+
}
16+
17+
int main() {
18+
cudaStream_t stream;
19+
20+
// CHECK: ishmem_fence();
21+
nvshmem_fence();
22+
23+
// CHECK: ishmem_quiet();
24+
nvshmem_quiet();
25+
26+
// ishmemx_quiet_on_queue(*stream);
27+
nvshmemx_quiet_on_stream(stream);
28+
29+
return 0;
30+
}

0 commit comments

Comments
 (0)