Commit c82a3b28 authored by Huber, Thomas's avatar Huber, Thomas
Browse files

Demonstration of basic mechanism for issuing device runtime messages

parent a59e21d4

openmp/libomptarget/include/OpenACCDebug.h

0 → 100644
+14 −0
Original line number Diff line number Diff line 
// Header file to demonstrate where unique OpenACC may reside



/// Print a generic message string for OpenACC from libomptarget or a plugin RTL

#define OPENACC_MESSAGE0(_str)                                                         \

  do {                                                                         \

    fprintf(stderr, "OpenACC (via Libomptarget) message: %s\n", _str);              \

  } while (0)



// Print a printf formatting string message for OpenACC from libomptarget or a plugin RTL

#define OPENACC_MESSAGE(_str, ...)                                                     \

  do {                                                                         \

    fprintf(stderr, "OpenACC (via Libomptarget) message: " _str "\n", __VA_ARGS__); \

  } while (0)

openmp/libomptarget/src/device.cpp

+9 −2
Original line number Diff line number Diff line
@@ -319,14 +319,21 @@ void *DeviceTy::getOrAllocTgtPtr(void *HstPtrBegin, void *HstPtrBase, 
       DPxPTR(HstPtrBegin), Size, DPxPTR(lr.Entry->HstPtrBegin),

       lr.Entry->HstPtrEnd - lr.Entry->HstPtrBegin);

    // Explicit extension of mapped data - not allowed.

    if (HasPresentModifier || !HasNoAllocModifier)

    int openacc_error_flag = 1;

    if ((HasPresentModifier && openacc_error_flag == 0) || (!HasNoAllocModifier && openacc_error_flag == 0))

      MESSAGE("explicit extension not allowed: host address specified is "

              DPxMOD " (%" PRId64

              " bytes), but device allocation maps to host at " DPxMOD

              " (%" PRId64 " bytes)",

            DPxPTR(HstPtrBegin), Size, DPxPTR(lr.Entry->HstPtrBegin),

            lr.Entry->HstPtrEnd - lr.Entry->HstPtrBegin);

    if (HasPresentModifier)

    if ((HasPresentModifier && openacc_error_flag == 1) || (!HasNoAllocModifier && openacc_error_flag == 0)) {

      OPENACC_MESSAGE( "var lives at " DPxMOD " size %" PRId64 

	      " partially present", 

	      DPxPTR(HstPtrBegin), Size);

      OPENACC_MESSAGE0("FATAL ERROR: variable in data clause is partially present on the device");

    }

    if (HasPresentModifier && openacc_error_flag == 0)

      MESSAGE("device mapping required by 'present' map type modifier does not "

              "exist for host address " DPxMOD " (%" PRId64 " bytes)",

              DPxPTR(HstPtrBegin), Size);

openmp/libomptarget/src/private.h

+1 −0
Original line number Diff line number Diff line
@@ -15,6 +15,7 @@ 


#include "device.h"

#include <Debug.h>

#include <OpenACCDebug.h>

#include <SourceInfo.h>

#include <omptarget.h>

present_simple.c

0 → 100644
+26 −0
Original line number Diff line number Diff line 
#include <stdio.h>

#define N 1000





int main() {



   int a[N];

   int b[N];

   int i;



   for (i = 0; i < N; i++) {

      a[i] = 5;

      b[i] = 0;

   }



   #pragma acc enter data copyin(a[0:500],b)



   #pragma acc parallel present(a) copyout(b)

   { 

      b[10] = a[10];

   }



   printf("The value of b[10] should be 5, it is %d", b[10]);

        

   return 0;

}