From patchwork Wed May 25 12:29:11 2016 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 7bit X-Patchwork-Submitter: Nathan Sidwell X-Patchwork-Id: 626153 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@bilbo.ozlabs.org Received: from sourceware.org (server1.sourceware.org [209.132.180.131]) (using TLSv1.2 with cipher ECDHE-RSA-AES256-GCM-SHA384 (256/256 bits)) (No client certificate requested) by ozlabs.org (Postfix) with ESMTPS id 3rFBRg4Vltz9s5l for ; Wed, 25 May 2016 22:29:35 +1000 (AEST) Authentication-Results: ozlabs.org; dkim=pass (1024-bit key; unprotected) header.d=gcc.gnu.org header.i=@gcc.gnu.org header.b=EODiY1w5; dkim-atps=neutral DomainKey-Signature: a=rsa-sha1; c=nofws; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender:to:cc :from:subject:message-id:date:mime-version:content-type; q=dns; s=default; b=feqAhdIhB2xWixRnne/vBkNOjjnauYC/2Au5mKg8N0JOMDpHBE kqKD2MQCHZk7yvthaXSBuiWCRAPKv5WAANNj5rL+/IgknLod+cKJ+V6Zf94NLwFX +lY1F0wm+E4QluEPVqQPUo/kGIhY8rQGU/DSeuOt1iWHShbw3eNEVRjZQ= DKIM-Signature: v=1; a=rsa-sha1; c=relaxed; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender:to:cc :from:subject:message-id:date:mime-version:content-type; s= default; bh=APUl8RXvy2KKTmyc4/07t8Qk5n0=; b=EODiY1w5vYhPPEn51SgZ pe5KDDYLCSieNZevCrOnV1+/XGpQpyrrW38JSQ3NigdYDN7J0IEkHMcgx2aNoDw2 xToUqnsCXLs0yZzUW7gHhcwKFMYPOaXoeAXs+faTKF2gFIIpGs8ObEghu+RO+NPx LAj6PlTFdYM5v2EKacD4+v8= Received: (qmail 91907 invoked by alias); 25 May 2016 12:29:25 -0000 Mailing-List: contact gcc-patches-help@gcc.gnu.org; run by ezmlm Precedence: bulk List-Id: List-Unsubscribe: List-Archive: List-Post: List-Help: Sender: gcc-patches-owner@gcc.gnu.org Delivered-To: mailing list gcc-patches@gcc.gnu.org Received: (qmail 91893 invoked by uid 89); 25 May 2016 12:29:25 -0000 Authentication-Results: sourceware.org; auth=none X-Virus-Found: No X-Spam-SWARE-Status: No, score=-1.8 required=5.0 tests=BAYES_00, FREEMAIL_FROM, KAM_ASCII_DIVIDERS, RCVD_IN_DNSWL_LOW, SPF_PASS autolearn=no version=3.3.2 spammy=nonexistent, H*MI:acm, UD:as, H*MI:d896 X-HELO: mail-qk0-f180.google.com Received: from mail-qk0-f180.google.com (HELO mail-qk0-f180.google.com) (209.85.220.180) by sourceware.org (qpsmtpd/0.93/v0.84-503-g423c35a) with (AES128-GCM-SHA256 encrypted) ESMTPS; Wed, 25 May 2016 12:29:14 +0000 Received: by mail-qk0-f180.google.com with SMTP id y126so32795680qke.1 for ; Wed, 25 May 2016 05:29:14 -0700 (PDT) X-Google-DKIM-Signature: v=1; a=rsa-sha256; c=relaxed/relaxed; d=1e100.net; s=20130820; h=x-gm-message-state:sender:to:cc:from:subject:message-id:date :user-agent:mime-version; bh=oiBJ00xKrqzYuHt71C/uaS0ubT60W+BG+oG7ljlO8eU=; b=czj5l4kfJlbojaiSAS+FPHeVm+2dBLxJ6ADcVEgYQT+zGW2hhNc0ak5Rf8sQeadaED u8R8R3Xc987NIPGnNOkvUGrXqvxz6aIjqhOitqWOVoUE4HXXjWhsRp/nLXdXYKKLFJZg qnTLTPHlKkMcjMzgJgY50PiaYcTHoJJke98iCdtAvvpANCSLwlRq7ZwC8GP3iygm47ML p2N232KrJb2ux3NtIQfLACY3WtU31E4410uMgUJrMiOUP/HU/pMVFxJ+Kxg/QSZADR+9 JOObLlJy/VB39pfYkU+0DRi4gCgQCh3+RMxcQKw8TxeTekySAok2TLkCugXh0tjezVx+ YDqg== X-Gm-Message-State: ALyK8tIiT4/zEHzVpCTqdXl/arFRe5NDW4pnSZ3/MMwO0HCl5RwKM1T8SmKn+XbHX9yIlQ== X-Received: by 10.55.99.87 with SMTP id x84mr3104640qkb.163.1464179352751; Wed, 25 May 2016 05:29:12 -0700 (PDT) Received: from ?IPv6:2601:181:c003:1930:a2a8:cdff:fe3e:b48? ([2601:181:c003:1930:a2a8:cdff:fe3e:b48]) by smtp.googlemail.com with ESMTPSA id b189sm2271341qkf.10.2016.05.25.05.29.11 (version=TLSv1/SSLv3 cipher=OTHER); Wed, 25 May 2016 05:29:11 -0700 (PDT) To: GCC Patches Cc: Alexander Monakov From: Nathan Sidwell Subject: [PTX] crt0 Message-ID: <1adcf079-7ac8-544d-d896-d6d246e689e1@acm.org> Date: Wed, 25 May 2016 08:29:11 -0400 User-Agent: Mozilla/5.0 (X11; Linux x86_64; rv:45.0) Gecko/20100101 Thunderbird/45.1.0 MIME-Version: 1.0 This patch reimplements crt0 as C rather than assembly. That means it;ll be good with 32 bit addresses. I've also completely moved abort and exit into newlib (a fork on github), which had it's own copies that failed to set the exit code. Things happened to work because they were never found, being hidden by the ones provided in crt0.s. You'll need to update your newlib sources for exit codes to continue to work. Alex, this should make it simple to add a check in __main for a single thread soft stack allocation. I imagine a check on a weakly declared __soft_stack symbol or something? Or is that something nvptx-run should check for and initialize? nathan 2016-05-25 Nathan Sidwell libgcc/ * config/nvptx/crt0.s: Delete. * config/nvptx/crt0.c: New. * t-nvptx: Update. gcc/testsuite/ * gcc.c-torture/execute/921110-1.c: Fix abort decl. Index: gcc/testsuite/gcc.c-torture/execute/921110-1.c =================================================================== --- gcc/testsuite/gcc.c-torture/execute/921110-1.c (revision 236531) +++ gcc/testsuite/gcc.c-torture/execute/921110-1.c (working copy) @@ -1,7 +1,8 @@ -extern int abort(); -typedef int (*frob)(); +extern void abort(void); +typedef void (*frob)(); frob f[] = {abort}; -main() + +int main(void) { exit(0); } Index: libgcc/config/nvptx/crt0.c =================================================================== --- libgcc/config/nvptx/crt0.c (nonexistent) +++ libgcc/config/nvptx/crt0.c (working copy) @@ -0,0 +1,37 @@ +/* Copyright (C) 2014-2016 Free Software Foundation, Inc. + + This file is free software; you can redistribute it and/or modify it + under the terms of the GNU General Public License as published by the + Free Software Foundation; either version 3, or (at your option) any + later version. + + This file is distributed in the hope that it will be useful, but + WITHOUT ANY WARRANTY; without even the implied warranty of + MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the GNU + General Public License for more details. + + Under Section 7 of GPL version 3, you are granted additional + permissions described in the GCC Runtime Library Exception, version + 3.1, as published by the Free Software Foundation. + + You should have received a copy of the GNU General Public License and + a copy of the GCC Runtime Library Exception along with this program; + see the files COPYING3 and COPYING.RUNTIME respectively. If not, see + . */ + +int *__exitval_ptr; + +extern void __attribute__((noreturn)) exit (int status); +extern int main (int, void **); + +void __attribute__((kernel)) +__main (int *rval_ptr, int argc, void **argv) +{ + __exitval_ptr = rval_ptr; + /* Store something non-zero, so the host knows something went wrong, + if we fail to reach exit properly. */ + if (rval_ptr) + *rval_ptr = 255; + + exit (main (argc, argv)); +} Index: libgcc/config/nvptx/crt0.s =================================================================== --- libgcc/config/nvptx/crt0.s (revision 236531) +++ libgcc/config/nvptx/crt0.s (nonexistent) @@ -1,45 +0,0 @@ - .version 3.1 - .target sm_30 - .address_size 64 - -.global .u64 %__exitval; -// BEGIN GLOBAL FUNCTION DEF: abort -.visible .func abort -{ - .reg .u64 %rd1; - ld.global.u64 %rd1,[%__exitval]; - st.u32 [%rd1], 255; - exit; -} -// BEGIN GLOBAL FUNCTION DEF: exit -.visible .func exit (.param .u32 %arg) -{ - .reg .u64 %rd1; - .reg .u32 %val; - ld.param.u32 %val,[%arg]; - ld.global.u64 %rd1,[%__exitval]; - st.u32 [%rd1], %val; - exit; -} - -.extern .func (.param.u32 retval) main (.param.u32 argc, .param.u64 argv); - -.visible .entry __main (.param .u64 __retval, .param.u32 __argc, .param.u64 __argv) -{ - .reg .u32 %r<3>; - .reg .u64 %rd<3>; - .param.u32 %argc; - .param.u64 %argp; - .param.u32 %mainret; - ld.param.u64 %rd0, [__retval]; - st.global.u64 [%__exitval], %rd0; - - ld.param.u32 %r1, [__argc]; - ld.param.u64 %rd1, [__argv]; - st.param.u32 [%argc], %r1; - st.param.u64 [%argp], %rd1; - call.uni (%mainret), main, (%argc, %argp); - ld.param.u32 %r1,[%mainret]; - st.s32 [%rd0], %r1; - exit; -} Index: libgcc/config/nvptx/t-nvptx =================================================================== --- libgcc/config/nvptx/t-nvptx (revision 236531) +++ libgcc/config/nvptx/t-nvptx (working copy) @@ -6,8 +6,8 @@ LIB2ADD=$(srcdir)/config/nvptx/malloc.as LIB2ADDEH= LIB2FUNCS_EXCLUDE=__main -crt0.o: $(srcdir)/config/nvptx/crt0.s - cp $< $@ +crt0.o: $(srcdir)/config/nvptx/crt0.c + $(crt_compile) -c $< # Prevent building "advanced" stuff (for example, gcov support). We don't # support it, and it may cause the build to fail, because of alloca usage, for