[Openmp-dev] The purpose of __tgt_offload_entry and how compiler generates the content for the offloading entry table and omp_offloading_entries section

[Yonghong Yan via Openmp-dev]

Hi Friends, hope you all are cool and safe.

We are experimenting with the libomptarget library to directly use with
source code. I read https://reviews.llvm.org/D64943 about how to create a
fat binary using clang-offload-wrapper utility. Given a CUDA program, I am
able to create a source file for what the clang-offload-wrapper can
produce. I however am not sure what to initialize the
__start_omp_offloading_entries and __stop_omp_offloading_entries symbols.
They are implicitly created by the linker when linking and creating the
omp_offloading_entries section. What I am not sure are what is the purpose
of the offload entry table and what could be put in the entry table by the
compiler/linker (the content of the omp_offloading_entries section)? From
what I read of the libomptarget design document, it seems they are used to
store global variables and functions that could be mapped to devices and
they are used at the runtime to help the mapping. How far of my
understanding is to what it really is?

For example, for the following axpy omp offloading, what could be put in
the offloading entry table? I built the program and did an objdump to see
the content of the section, but not much clue what they are.

void axpy(int N, float *Y, float *X, float a) {
   int i,j;
   #pragma omp target map(to:X[0:N], N) map(tofrom:Y[0:N])
   #pragma omp parallel for
   for (i = 0; i < N; ++i) {
        Y[i] += a * X[i];
   }
}

Thank you
Yonghong
-------------- next part --------------
An HTML attachment was scrubbed...
URL: <http://lists.llvm.org/pipermail/openmp-dev/attachments/20200715/8fcbddb2/attachment.html>