From mboxrd@z Thu Jan 1 00:00:00 1970 Return-Path: Received: (qmail 20151 invoked by alias); 23 Aug 2013 17:15:46 -0000 Mailing-List: contact gcc-help@gcc.gnu.org; run by ezmlm Precedence: bulk List-Id: List-Archive: List-Post: List-Help: Sender: gcc-owner@gcc.gnu.org Received: (qmail 20141 invoked by uid 89); 23 Aug 2013 17:15:45 -0000 X-Spam-SWARE-Status: No, score=-2.6 required=5.0 tests=BAYES_00,FREEMAIL_FROM,RCVD_IN_DNSWL_LOW,RCVD_IN_HOSTKARMA_YE,SPF_PASS autolearn=ham version=3.3.2 Received: from mail-pb0-f44.google.com (HELO mail-pb0-f44.google.com) (209.85.160.44) by sourceware.org (qpsmtpd/0.84/v0.84-167-ge50287c) with ESMTP; Fri, 23 Aug 2013 17:15:44 +0000 Received: by mail-pb0-f44.google.com with SMTP id xa7so914128pbc.17 for ; Fri, 23 Aug 2013 10:15:43 -0700 (PDT) X-Received: by 10.66.161.99 with SMTP id xr3mr4066389pab.172.1377278143509; Fri, 23 Aug 2013 10:15:43 -0700 (PDT) Received: from msticlxl57.ims.intel.com ([192.55.54.42]) by mx.google.com with ESMTPSA id uw6sm942115pbc.8.1969.12.31.16.00.00 (version=TLSv1 cipher=RC4-SHA bits=128/128); Fri, 23 Aug 2013 10:15:42 -0700 (PDT) Date: Sun, 25 Aug 2013 22:36:00 -0000 From: Ilya Verbin To: Richard Biener , Jakub Jelinek Cc: Uday Khedker , hubicka@ucw.cz, rth@redhat.com, kirill.yukhin@gmail.com, "Michael V. Zolotukhin" , gcc@gcc.gnu.org Subject: Re: Questions about LTO infrastructure and pragma omp target Message-ID: <20130823171514.GB6976@msticlxl57.ims.intel.com> References: <20130815133639.GA42355@msticlxl57.ims.intel.com> <20130815134437.GB1814@tucnak.redhat.com> <4df844f6-f385-4e63-9413-8ea341992b77@email.android.com> <20130823105527.GA6976@msticlxl57.ims.intel.com> <85e37f42-69fe-4bbf-bf1d-f73194e7c444@email.android.com> <20130823123638.GL1814@tucnak.redhat.com> MIME-Version: 1.0 Content-Type: text/plain; charset=us-ascii Content-Disposition: inline In-Reply-To: <20130823123638.GL1814@tucnak.redhat.com> User-Agent: Mutt/1.5.21 (2010-09-15) X-SW-Source: 2013-08/txt/msg00288.txt.bz2 On 23 Aug 13:17, Jakub Jelinek wrote: > I don't think we should stream into more than one target section. > There should be just .gnu.target_lto section (or whatever other suitable > name) and should stream into it: > 1) all functions and variables with "omp declare target" attribute > 2) the outlined bodies of #pragma omp target turned into *.ompfn functions > 3) all the types, symtab etc. needed for that Why having one target section is preferable than multiple sections for each function body? > Then the question is what the plugin should perform with these sections, > whether it will compile each input .gnu.target_lto section hunk separately > (as in non-LTO mode), or with -flto also LTO them together. Yes, it is an important question... To get started it is easier to implement "non-target-lto" mode, however this approach should be general enough to extend it to "target-lto" mode. Does anyone need it? On 23 Aug 14:36, Jakub Jelinek wrote: > On Fri, Aug 23, 2013 at 02:24:42PM +0200, Richard Biener wrote: > > No, as you will refer to the symbol with the target code from the host > > code you need a single unified symtab. > > I really think you want two symtabs rather than a unified symtab, > or just stream a subset of the host symtab into the .gnu.target_lto > section. The thing is, the target code (functions, vars, outlined bodies) > is a strict subset of the host code (because as a fallback, everything > needs to be able to run on the host), but when not compiling originally with > -flto, we IMHO should stream just the target subset, not everything > (and for -flto stream both the target subset into one section and everything > (host code) as we do right now, either with fat or slim lto objects). I also think that having two symtabs looks better. There is no direct refs to the target symbols from the host code. And (as far as I see it) unified symtab will lead to mess in places, where host and target symbols should be handled differently. Thanks, -- Ilya