offload/test/offloading/memory_manager.cpp

   1 // RUN: %libomptarget-compilexx-run-and-check-generic
   2
   3 // REQUIRES: nvidiagpu
   4
   5 #include <omp.h>
   6
   7 #include <cassert>
   8 #include <iostream>
   9
  10 int main(int argc, char *argv[]) {
  11 #pragma omp parallel for
  12   for (int i = 0; i < 16; ++i) {
  13     for (int n = 1; n < (1 << 13); n <<= 1) {
  14       void *p = omp_target_alloc(n * sizeof(int), 0);
  15       omp_target_free(p, 0);
  16     }
  17   }
  18
  19 #pragma omp parallel for
  20   for (int i = 0; i < 16; ++i) {
  21     for (int n = 1; n < (1 << 13); n <<= 1) {
  22       int *p = (int *)omp_target_alloc(n * sizeof(int), 0);
  23 #pragma omp target teams distribute parallel for is_device_ptr(p)
  24       for (int j = 0; j < n; ++j) {
  25         p[j] = i;
  26       }
  27       int buffer[n];
  28 #pragma omp target teams distribute parallel for is_device_ptr(p)              \
  29     map(from : buffer)
  30       for (int j = 0; j < n; ++j) {
  31         buffer[j] = p[j];
  32       }
  33       for (int j = 0; j < n; ++j) {
  34         assert(buffer[j] == i);
  35       }
  36       omp_target_free(p, 0);
  37     }
  38   }
  39
  40   std::cout << "PASS\n";
  41   return 0;
  42 }
  43
  44 // CHECK: PASS