Skip to content

New 6.0 test for target access alloc using single #860

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.

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

Open
wants to merge 11 commits into
base: master
Choose a base branch
from
78 changes: 78 additions & 0 deletions tests/6.0/target/test_target_access.c
Original file line number Diff line number Diff line change
@@ -0,0 +1,78 @@
//===--- test_target_access_single.c --------------------===//
//
// OpenMP API Version 6.0
// Tests that the target_access single allocator trait works
// as intended for basic device-host operations and multiple
// device access.
//
//===----------------------------------------------------===//
#include <omp.h>
#include <stdio.h>
#include "ompvv.h"
#define N 16

int test_target_access_single() {
int errors = 0;
int *array;
int num_devices = omp_get_num_devices();

OMPVV_WARNING_IF(num_devices < 2, "Test requires at least 2 devices, but only %d found.", num_devices);
Copy link
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Suggested change
OMPVV_WARNING_IF(num_devices < 2, "Test requires at least 2 devices, but only %d found.", num_devices);
OMPVV_WARNING_IF(num_devices < 2, "Test requires at least 2 devices, but only %d found. \n This test will be skipped.\n", num_devices);

if (num_devices < 2) {
return errors;
}

int first_dev = 0;
int second_dev = 1;

omp_alloctrait_t trait = { omp_atk_target_access, omp_atv_single };
omp_allocator_handle_t single_allocator = omp_init_allocator(omp_default_mem_space, 1, &trait);

array = (int*)omp_alloc(N * sizeof(int), single_allocator);

for (int i = 0; i < N; i++) {
array[i] = i;
}

// Using first device
#pragma omp target enter data map(to: array[:N]) device(first_dev)

#pragma omp target is_device_ptr(array) device(first_dev)
{
for (int i = 0; i < N; i++) {
array[i] = i * N;
}
}

#pragma omp target exit data map(from: array[:N]) device(first_dev)

// Using second device
#pragma omp target enter data map(to: array[:N]) device(second_dev)

#pragma omp target is_device_ptr(array) device(second_dev)
{
for (int i = 0; i < N; i++) {
array[i] += 5;
}
}

#pragma omp target exit data map(from: array[:N]) device(second_dev)

for (int i = 0; i < N; i++) {
OMPVV_TEST_AND_SET_VERBOSE(errors, array[i] != i * N + 5);
}

omp_free(array, single_allocator);
omp_destroy_allocator(single_allocator);

return errors;
}

int main() {
OMPVV_TEST_OFFLOADING;

int errors = 0;

OMPVV_TEST_AND_SET_VERBOSE(errors, test_target_access_single() != 0);

OMPVV_REPORT_AND_RETURN(errors);
}