OpenMP offload: Using global variable with a library

[Christos_Kotsalos]

Hello,

I have two questions regarding the usage of global variables with OpenMP offload:

1. When I have global variable usage in the code from which I create a library then I get a runtime error. Here is a simple reproducer showing the issue:

-> qsub -I -l nodes=1:gpu:ppn=2 -d . (Intel DevCloud)

$ cat test.sh

export OMP_TARGET_OFFLOAD=MANDATORY

CXX=icpx
CXXFLAGS="-qopenmp -fopenmp-targets=spir64"
${CXX} ${CXXFLAGS} -c test.cpp
ar cq libtest.a test.o
${CXX} ${CXXFLAGS} -o test1_ main.cpp -L. -ltest
${CXX} ${CXXFLAGS} -o test2_ main.cpp test.o

echo "*****>intel"
if [ -f test1_ ] ; then ./test1_; echo $?; fi
if [ -f test2_ ] ; then ./test2_; echo $?; fi
rm test*_ *.o *.a
echo "*****<"

$ cat test.cpp

#pragma omp declare target
int y;
#pragma omp end declare target

int test() {
  y = 24;
  #pragma omp target update to(y)
  y = 42;

  int x;
  #pragma omp target map(from:x)
  {
    x = y;
  }
  return x;
}

$ cat main.cpp

extern int test();

int main() {
  return test();
}

Running the ./test2_ works as expected as I am not using static library but the ./test1_ fails with an error shown below:

LEVEL0 error: Failed to invoke deleted kernel.
Libomptarget error: Executing target region abort target.
Libomptarget error: Run with
Libomptarget error: LIBOMPTARGET_DEBUG=1 to display basic debug information.
Libomptarget error: LIBOMPTARGET_DEBUG=2 to display calls to the compute runtime.
Libomptarget error: LIBOMPTARGET_INFO=4 to dump host-target pointer mappings.
Libomptarget error: Source location information not present. Compile with -g or -gline-tables-only.
Libomptarget fatal error 1: failure of target construct while offloading is mandatory
Aborted
134

FYI, the correct output is 24 (check ./test2_).

Is this expected behaviour? Is there any workaround? I have tested this with PGI/NVHPC compiler and it works there.

2. The second scenario is similar but now I am trying to use a global variable from the library into the offload region in main.cpp i.e. modified main.cpp looks as:

$ cat main.cpp

extern int test();

#include <cstdio>

#pragma omp declare target
extern int y;
#pragma omp end declare target

int main() {
  #pragma omp target teams distribute parallel for
  for(int i=0; i<5; i++) {
    printf("--> %d \n", y + i);
  }
  return test();
}

While it compiles for both tests, test2_ fails with the following error:

if [ -f test1_ ] ; then ./test1_; echo $?; fi
--> 0
--> 1
--> 2
--> 3
--> 4
LEVEL0 error: Failed to invoke deleted kernel.
Libomptarget error: Executing target region abort target.
Libomptarget error: Run with
Libomptarget error: LIBOMPTARGET_DEBUG=1 to display basic debug information.
Libomptarget error: LIBOMPTARGET_DEBUG=2 to display calls to the compute runtime.
Libomptarget error: LIBOMPTARGET_INFO=4 to dump host-target pointer mappings.
unknown:11:11: Libomptarget fatal error 1: failure of target construct while offloading is mandatory
Aborted
134

Any feedback would be really helpful and appreciated.

Thank you very much.

Best,

Christos

[SantoshY_Intel]

Hi,

Thank you for posting in the Intel forums.

Thanks for providing the sample reproducer code along with the steps to reproduce your issue. We were able to reproduce your issue on Intel DevCloud using the ICPX compiler. We are working on your issue and we will get back to you soon.

Thanks & Regards,

Santosh

[Christos_Kotsalos]

Hi Santosh,

Thank you very much for your quick response.

I am looking forward to hearing from you regarding this compiler issue.

Best,

Christos

[SantoshY_Intel · Accepted solution]

Hi Christos,

We have reported this issue to the concerned development team. Your issue will be fixed in the next future release.

Meanwhile, could you please use the below workaround?

Use the below command which references the full static library name(libtest.a in this case) instead of -lname(ltest in this case):

icx -fiopenmp -fopenmp-targets=spir64 -o test1_ main.cpp -L. libtest.a

We tried this workaround at our end and it is working fine without any errors as shown below.

$ cat test.sh

export OMP_TARGET_OFFLOAD=MANDATORY
CXX=icpx
CXXFLAGS="-qopenmp -fopenmp-targets=spir64"
${CXX} ${CXXFLAGS} -c test.cpp
ar cq libtest.a test.o
#${CXX} ${CXXFLAGS} -o test1_ main.cpp -L. -ltest
${CXX} ${CXXFLAGS} -o test1_ main.cpp -L. libtest.a
${CXX} ${CXXFLAGS} -o test2_ main.cpp test.o

echo "*****>intel"
if [ -f test1_ ] ; then ./test1_; echo $?; fi
if [ -f test2_ ] ; then ./test2_; echo $?; fi

rm test*_ *.o *.a
echo "*****<"

The output for scenario-1 would be:

u67125@s001-n141:~/offload$ sh test.sh
*****>intel
24
24
*****<
u67125@s001-n141:~/offload$

The output for scenario-2 would be:

u67125@s001-n141:~/offload$ sh test.sh
*****>intel
--> 0
--> 1
--> 2
--> 3
--> 4
24
--> 0
--> 1
--> 2
--> 3
--> 4
24
*****<
u67125@s001-n141:~/offload$

Thanks & Regards,

Santosh

[Christos_Kotsalos]

Hi Santosh,

Thank you very much for the workaround and for letting me know about the future release.

FYI, I have tried the workaround and everything works fine.

Thank you for resolving the issue.

Best,

Christos

[SantoshY_Intel]

Hi Christos,

Thanks for accepting our solution. If you need any additional information, please post a new question as this thread will no longer be monitored by Intel. 

Thanks & Regards,

Santosh