<html>
<head>
<meta http-equiv="Content-Type" content="text/html; charset=us-ascii">
<style type="text/css" style="display:none;"> P {margin-top:0;margin-bottom:0;} </style>
</head>
<body dir="ltr">
<div style="font-family: Calibri, Arial, Helvetica, sans-serif; font-size: 12pt; color: rgb(0, 0, 0);">
Kumbhar,<br>
<br>
As ye said, the current support for linking static libraries is a little flaky. I'm currently working on a series of patches to hopefully address this problem that is under review at
<a href="https://reviews.llvm.org/D116541" id="LPNoLPOWALinkPreview">https://reviews.llvm.org/D116541</a> and its child revisions. I can currently link and run your tests using my development branch after adding some additional support to search the library
paths for static libraries. This should solve your problem once these patches are ready to land upstream.<br>
<br>
Thanks,<br>
Joseph Huber</div>
<div class="_Entity _EType_OWALinkPreview _EId_OWALinkPreview _EReadonly_1"></div>
<div id="appendonsend"></div>
<hr style="display:inline-block;width:98%" tabindex="-1">
<div id="divRplyFwdMsg" dir="ltr"><font face="Calibri, sans-serif" style="font-size:11pt" color="#000000"><b>From:</b> Openmp-dev <openmp-dev-bounces@lists.llvm.org> on behalf of pramod kumbhar via Openmp-dev <openmp-dev@lists.llvm.org><br>
<b>Sent:</b> Tuesday, January 4, 2022 6:59 PM<br>
<b>To:</b> openmp-dev@lists.llvm.org <openmp-dev@lists.llvm.org><br>
<b>Subject:</b> [EXTERNAL] [Openmp-dev] OpenMP offload: Using global variable with a library result in "CUDA error: Loading global 'xxxx' Failed" or "nvlink error : Undefined reference to ..."</font>
<div> </div>
</div>
<div>
<div dir="ltr">Dear All,
<div><br>
</div>
<div>I have two questions regarding the usage of global variables with OpenMP offload:</div>
<div><br>
</div>
<div>1. When I have global variable usage in the code from which I create a library then I get "CUDA error: Loading global 'x' Failed" error. Here is a simple reproducer showing the issue:</div>
<div><br>
</div>
<blockquote style="margin:0 0 0 40px; border:none; padding:0px">
<div><font color="#20124d">$ cat test.sh</font></div>
<div><font color="#20124d"><br>
</font></div>
<div><font color="#20124d">CXX=clang++</font></div>
<div><font color="#20124d">CXXFLAGS="-fopenmp -fopenmp-targets=nvptx64-nvidia-cuda -g -O2"</font></div>
<div><font color="#20124d">${CXX} ${CXXFLAGS} -c test.cpp</font></div>
<div><font color="#20124d">ar cq libtest.a test.o</font></div>
<div><font color="#20124d">${CXX} ${CXXFLAGS} -o test1 main.cpp -L. -ltest</font></div>
<div><font color="#20124d">${CXX} ${CXXFLAGS} -o test2 main.cpp test.o</font></div>
<div><font color="#20124d"><br>
</font></div>
<div><font color="#0b5394">$ cat test.cpp</font></div>
<div><font color="#0b5394"><br>
#pragma omp declare target<br>
int y;<br>
#pragma omp end declare target<br>
<br>
int test() {<br>
y = 24;<br>
#pragma omp target update to(y)<br>
y = 42;<br>
<br>
int x;<br>
#pragma omp target map(from:x)<br>
{<br>
x = y;<br>
}<br>
return x;<br>
}</font></div>
<div><font color="#0b5394"><br>
$ cat main.cpp<br>
extern int test();<br>
<br>
int main() {<br>
return test();<br>
}</font><font color="#20124d"><br>
</font></div>
</blockquote>
<div><br>
</div>
Running the ./<i>test2</i> works as expected as I am not using static library but the ./test1 fails with an error shown below:
<div><br>
<blockquote style="margin:0 0 0 40px; border:none; padding:0px">
<div><font color="#351c75">$ ./test2</font></div>
<div><font color="#351c75">$ echo $?</font></div>
<div><font color="#351c75">24</font></div>
<div><font color="#351c75">$ ./test1</font></div>
<div><font color="#351c75">CUDA error: Loading global 'y' Failed</font></div>
<div><font color="#351c75">CUDA error: named symbol not found</font></div>
<div><font color="#351c75">Libomptarget error: Unable to generate entries table for device id 0.</font></div>
<div><font color="#351c75">Libomptarget error: Failed to init globals on device 0</font></div>
<div><font color="#351c75">Libomptarget error: Run with LIBOMPTARGET_INFO=4 to dump host-target pointer mappings.</font></div>
<div><font color="#351c75">test.cpp:7:3: Libomptarget fatal error 1: failure of target construct while offloading is mandatory</font></div>
<div><font color="#351c75">Aborted</font></div>
</blockquote>
<div>
<blockquote style="margin:0 0 0 40px; border:none; padding:0px">
<div><br>
</div>
</blockquote>
Is this expected behaviour? Is there any workaround? I have tested this with PGI/NVHPC compiler and it works there.</div>
<div><br>
</div>
<div>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:</div>
<div><br>
</div>
<blockquote style="margin:0 0 0 40px; border:none; padding:0px">
<div><font color="#20124d">$ cat main.cpp</font></div>
<div><font color="#20124d">extern int test();</font></div>
<div><font color="#20124d"><br>
</font></div>
<div><font color="#20124d">#include <cstdio></font></div>
<div><font color="#20124d"><br>
</font></div>
<div><font color="#20124d">#pragma omp declare target</font></div>
<div><font color="#20124d">extern int y;</font></div>
<div><font color="#20124d">#pragma omp end declare target</font></div>
<div><font color="#20124d"><br>
</font></div>
<div><font color="#20124d">int main() {</font></div>
<div><font color="#20124d"> #pragma omp target teams distribute parallel for</font></div>
<div><font color="#20124d"> for(int i=0; i<5; i++) {</font></div>
<div><font color="#20124d"> printf("--> %d \n", y + i);</font></div>
<div><font color="#20124d"> }</font></div>
<div><font color="#20124d"> return test();</font></div>
<div><font color="#20124d">}</font></div>
<div><font color="#20124d"><br>
</font></div>
</blockquote>
<font color="#20124d">This now fails to compile with a library:</font>
<div><font color="#20124d"><br>
</font></div>
<blockquote style="margin:0 0 0 40px; border:none; padding:0px">
<div><font color="#351c75">+ clang++ -fopenmp -fopenmp-targets=nvptx64-nvidia-cuda -g -O2 -c test.cpp</font></div>
<div><font color="#351c75">+ ar cq libtest.a test.o</font></div>
<div><font color="#351c75">+ clang++ -fopenmp -fopenmp-targets=nvptx64-nvidia-cuda -g -O2 -o test1 main.cpp -L. -ltest</font></div>
<div><font color="#351c75">nvlink error : Undefined reference to 'y' in '/gpfs/<a href="http://bbp.cscs.ch/ssd/slurmTmpFS/kumbhar/151151/main-9ea2c6.cubin">bbp.cscs.ch/ssd/slurmTmpFS/kumbhar/151151/main-9ea2c6.cubin</a>'</font></div>
<div><font color="#351c75">clang-13: error: nvlink command failed with exit code 255 (use -v to see invocation)</font></div>
<div><font color="#351c75">clang version 13.0.0</font></div>
<div><font color="#351c75">Target: x86_64-unknown-linux-gnu</font></div>
<div><font color="#351c75">Thread model: posix</font></div>
<div><font color="#351c75">InstalledDir: /gpfs/<a href="http://bbp.cscs.ch/ssd/apps/bsd/2021-11/stage_externals/install_gcc-11.2.0-skylake/llvm-13.0.0-lvcrm6/bin">bbp.cscs.ch/ssd/apps/bsd/2021-11/stage_externals/install_gcc-11.2.0-skylake/llvm-13.0.0-lvcrm6/bin</a></font></div>
<div><font color="#351c75"><br>
</font></div>
</blockquote>
<font color="#000000"><span style="">In a big application, we build & ship libraries and it's not easy/convenient to use objects for linking.
</span></font><span style="color:rgb(0,0,0)">Do you have any recommendations to solve/workaround issues for this use case?</span>
<div>
<div><br>
</div>
<div>Thank you very much!</div>
<div><br>
</div>
<div>Regards,</div>
<div>Pramod Kumbhar</div>
</div>
</div>
</div>
</div>
</body>
</html>