This is the mail archive of the mailing list for the GCC project.

Index Nav: [Date Index] [Subject Index] [Author Index] [Thread Index]
Message Nav: [Date Prev] [Date Next] [Thread Prev] [Thread Next]
Other format: [Raw text]

Re: [gomp4] Tweak GOMP_target{,_data,_update} arguments

On 19 Sep 11:23, Jakub Jelinek wrote:
> that.  Another complication is dependent shared libraries.
> Consider
> liba.c:
> #pragma omp declare target
> int i;
> int foo (void)
> {
>   return ++i;
> }
> #pragma omp end declare target
> main.c:
> #pragma omp declare target
> extern int i;
> extern int foo (void);
> #pragma omp end declare target
> int main ()
> {
>   int j;
>   #pragma omp target
>     {
>       j = i;
>       j += foo ();
>     }
>   if (j != 1)
>     abort ();
>   return 0;
> }
> gcc -shared -O2 -fpic -fopenmp -o -Wl,-soname, liba.c
> gcc -O2 -fopenmp -o main main.c -L. -la
> ./main
> Perhaps the linker plugin can extract the target shared libraries from
> the embedded sections of dependent shared libraries (if any), and link the
> "main" shared library against that, but GOMP_target will need to know that
> it can't just offload, but also has to offload the dependent
> (and of course from the libgomp plugin).
> What does ICC do in this case?
> 	Jakub

Hi Jakub,

Here's what ICC does.
Suppose we have liba.c and main.c, both with target regions:

1. Building liba.c ->
A call to offload-runtime library is inserted into _init of
Target region is compiled into, and placed into .rodata of

2. Building main.c -> main.exe.
Similarly, a call to offload-runtime library is inserted into _init of main.exe.
Target region is compiled into, and placed into .rodata of

3. Runtime.
So, when and main.exe are loaded at host-side, the runtime library
knows, that it should transfer and to the
target-side.  Then, main.exe starts execution.  At every entry point to the
target region, runtime library checks whether it should perform an
initialization.  If target is not initialized, runtime library calls
COIProcessCreateFromMemory(main_target.exe), that transfers some standard
main_target.exe to the target and starts it.  Then, runtime library calls
COIProcessLoadLibraryFromMemory(,, that transfers
these libraries to the target and loads them into the main_target.exe.
The target-side functions are called from host through
COIProcessGetFunctionHandles("f_name") and COIPipelineRunFunction(handle). The
addresses of target-side functions are obtained from * by dlsym().
So, the host-side knows nothing about target addresses.

What do you think, how will such an approach work with other target
architectures, and with current implementation of GOMP_target{,_data,_update}?

  -- Ilya

Index Nav: [Date Index] [Subject Index] [Author Index] [Thread Index]
Message Nav: [Date Prev] [Date Next] [Thread Prev] [Thread Next]