<html>
<head>
<base href="https://bugs.llvm.org/">
</head>
<body><table border="1" cellspacing="0" cellpadding="8">
<tr>
<th>Bug ID</th>
<td><a class="bz_bug_link
bz_status_NEW "
title="NEW - -O0 built openmp offload for nvidia, occasionally fails trying to copy data"
href="https://bugs.llvm.org/show_bug.cgi?id=51552">51552</a>
</td>
</tr>
<tr>
<th>Summary</th>
<td>-O0 built openmp offload for nvidia, occasionally fails trying to copy data
</td>
</tr>
<tr>
<th>Product</th>
<td>OpenMP
</td>
</tr>
<tr>
<th>Version</th>
<td>unspecified
</td>
</tr>
<tr>
<th>Hardware</th>
<td>Other
</td>
</tr>
<tr>
<th>OS</th>
<td>Linux
</td>
</tr>
<tr>
<th>Status</th>
<td>NEW
</td>
</tr>
<tr>
<th>Severity</th>
<td>normal
</td>
</tr>
<tr>
<th>Priority</th>
<td>P
</td>
</tr>
<tr>
<th>Component</th>
<td>Runtime Library
</td>
</tr>
<tr>
<th>Assignee</th>
<td>unassignedbugs@nondot.org
</td>
</tr>
<tr>
<th>Reporter</th>
<td>rlieberm@amd.com
</td>
</tr>
<tr>
<th>CC</th>
<td>llvm-bugs@lists.llvm.org
</td>
</tr></table>
<p>
<div>
<pre>cmake command to build trunk for nvptx
cmake -DCMAKE_INSTALL_PREFIX=/xtmp/rlieberm/trunkNvptx_1.0 \
-DCMAKE_BUILD_TYPE=Release \
-DLLVM_ENABLE_PROJECTS="clang" \
-DLLVM_TARGETS_TO_BUILD="X86;NVPTX" \
-DLLVM_ENABLE_ASSERTIONS=ON \
-DLLVM_CCACHE_BUILD=$OPTION_CCACHE \
-DLLVM_ENABLE_RUNTIMES="openmp" \
-DLIBOMPTARGET_CMAKE_BUILD_TYPE=debug \
../llvm
compile command
/xtmp/rlieberm/trunkNvptx_1.0//bin/clang -O0 -target x86_64-pc-linux-gnu
-fopenmp -fopenmp-targets=nvptx64-nvidia-cuda
-Xopenmp-target=nvptx64-nvidia-cuda -march=sm_50 nested_loop.c -o nested_loop
-L/usr/local/cuda/targets/x86_64-linux/lib -lcudart
execution, this is intermittent in failure, so it may reproduce , i hope so
$ LIBOMPTARGET_INFO=16 LD_LIBRARY_PATH=/xtmp/rlieberm/trunkNvptx_1.0/lib
./nested_loop
CUDA device 0 info: Device supports up to 65536 CUDA blocks and 1024 threads
with a warp size of 32
CUDA device 0 info: Device heap size is 8388608 Bytes, device stack size is
1024 Bytes per thread
CUDA device 0 info: Launching kernel __omp_offloading_811_181cfa_main_l9 with 1
blocks and 38 threads in Generic mode
CUDA error: an illegal memory access was encountered
Libomptarget error: Copying data from device failed.
Libomptarget error: Call to targetDataEnd failed, abort target.
Libomptarget error: Failed to process data after launching the kernel.
Libomptarget error: Run with 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 (core dumped)
Test case:
#include <stdio.h>
#include "assert.h"
#include <unistd.h>
int t1 = 6;
int main() {
int wayout=0;
#pragma omp target map(tofrom: wayout)
#pragma omp teams num_teams(1 ) thread_limit(t1)
{ wayout =1;
}
printf("wayout=%d\n", wayout);
if(wayout==1){
printf("Succeeded\n");
return 0;
} else {
printf("Failed\n");
return 1;
}
}</pre>
</div>
</p>
<hr>
<span>You are receiving this mail because:</span>
<ul>
<li>You are on the CC list for the bug.</li>
</ul>
</body>
</html>