From patchwork Mon Aug 31 17:39:46 2015 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 7bit X-Patchwork-Submitter: Nathan Sidwell X-Patchwork-Id: 512555 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 13C521401CD for ; Tue, 1 Sep 2015 03:40:01 +1000 (AEST) Authentication-Results: ozlabs.org; dkim=pass (1024-bit key; unprotected) header.d=gcc.gnu.org header.i=@gcc.gnu.org header.b=onUB4mM3; 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 :from:subject:message-id:date:mime-version:content-type; q=dns; s=default; b=bVwQvF5ulLnw6JZ+SuoaZ8B59u4MT7LWC1+tXWhSZFIMUK+xwB RBSK+gDCPmTgt1iJVOZrVUuZ0O8ZJ8WVJZCMH7qP+i6I5JSwgBSA2eH2Hk18k47G AZsbUnAFUA6fWVRTbDgCFdSHmKO1VdKloqM10z1IPhnDXySsCenx6IjyA= 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 :from:subject:message-id:date:mime-version:content-type; s= default; bh=1imuVuaY2QTm7Tlh0AcSX6UuyOk=; b=onUB4mM3uS+ieIAiUMWP ou8ZJNFUUUXekfV8PwKCGfcsnfYkem/1DAnGcHaO7+s1/X8J5c77qSN6orbqfd0d mTb25YOuiIEuWhecixW3sw7evMqBeoqTnn03vNtDvIf0vPQ6El9Q/PtngMszb+QR P2akeCJtIg9B+kBuySMDYS0= Received: (qmail 13307 invoked by alias); 31 Aug 2015 17:39:53 -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 13298 invoked by uid 89); 31 Aug 2015 17:39:53 -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 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; Mon, 31 Aug 2015 17:39:50 +0000 Received: by qkbp67 with SMTP id p67so7898412qkb.3 for ; Mon, 31 Aug 2015 10:39:48 -0700 (PDT) X-Received: by 10.55.18.6 with SMTP id c6mr11283224qkh.96.1441042788176; Mon, 31 Aug 2015 10:39:48 -0700 (PDT) Received: from ?IPv6:2601:181:c000:c497:a2a8:cdff:fe3e:b48? ([2601:181:c000:c497:a2a8:cdff:fe3e:b48]) by smtp.googlemail.com with ESMTPSA id 44sm9195967qgx.25.2015.08.31.10.39.47 (version=TLSv1.2 cipher=ECDHE-RSA-AES128-GCM-SHA256 bits=128/128); Mon, 31 Aug 2015 10:39:47 -0700 (PDT) To: GCC Patches From: Nathan Sidwell Subject: [gomp4] expunge shared_size from launch API Message-ID: <55E49162.70202@acm.org> Date: Mon, 31 Aug 2015 13:39:46 -0400 User-Agent: Mozilla/5.0 (X11; Linux x86_64; rv:38.0) Gecko/20100101 Thunderbird/38.1.0 MIME-Version: 1.0 I've committed this to the gomp4 branch. It expunges the shared_size parameter, which is no longer needed with Cesar's reduction work. shared_size never made it to trunk, so nothing to port there. nathan 2015-08-31 Nathan Sidwell libgomp/ * oacc-parallel.c (__goacc_host_ganglocal_ptr): Delete. (GOACC_get_ganglocal_ptr): Delete. (alloc_host_shared_mem, free_host_shared_mem): Delete. (GOACC_parallel_keyed): Remove shared_size param. (GOACC_parallel): Likewise. * libgomp_g.h (GOACC_parallel_keyed): Adjust prototype. * plugin/plugin-nvptx.c (nvptx_exec): Lose shared_size parameter. (GOMP_OFFLOAD_openacc_parallel): Likewise. * libgomp.map (GOACC_get_ganglocal_ptr): Remove. * libgomp.h (struct acc_dispatch_t): Adjust exec_func prototype. * oacc-host.c (host_openacc_exec): Lose shared_size parameter. gcc/ * omp-low.c (expand_omp_target): Lose shared_size parameter. * omp-builtins.def (GOACC_KERNELS, GOACC_PARALLEL): Lose shared_size parameter. * builtin-types.def (DEF_FUNCTION_TYPE_VAR_6): Define. fortran/ * f95-lang.c (DEF_FUNCTION_TYPE_VAR_6): Define. * types.def (BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_VAR): Declare. (BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_SIZE_VAR): Delete. lto/ * lto-lang.c (DEF_FUNCTION_TYPE_VAR_6): Define. c-family/ * c-common.c (DEF_FUNCTION_TYPE_VAR_6): Define. Index: libgomp/libgomp_g.h =================================================================== --- libgomp/libgomp_g.h (revision 227269) +++ libgomp/libgomp_g.h (working copy) @@ -224,13 +224,9 @@ extern void GOACC_data_end (void); extern void GOACC_enter_exit_data (int, size_t, void **, size_t *, unsigned short *, int, int, ...); extern void GOACC_parallel_keyed (int, void (*) (void *), size_t, - void **, size_t *, unsigned short *, - size_t, ...); + void **, size_t *, unsigned short *, ...); extern void GOACC_update (int, size_t, void **, size_t *, unsigned short *, int, int, ...); extern void GOACC_wait (int, int, ...); -extern int GOACC_get_num_threads (int, int, int); -extern int GOACC_get_thread_num (int, int, int); -extern void *GOACC_get_ganglocal_ptr (void); #endif /* LIBGOMP_G_H */ Index: libgomp/libgomp.map =================================================================== --- libgomp/libgomp.map (revision 227352) +++ libgomp/libgomp.map (working copy) @@ -336,14 +336,12 @@ GOACC_2.0 { GOACC_2.0.GOMP_4_BRANCH { global: GOACC_deviceptr; - GOACC_get_ganglocal_ptr; GOACC_parallel_keyed; GOACC_register_static; GOMP_set_offload_targets; Index: libgomp/libgomp.h =================================================================== --- libgomp/libgomp.h (revision 227269) +++ libgomp/libgomp.h (working copy) @@ -694,7 +694,7 @@ typedef struct acc_dispatch_t struct target_mem_desc *data_environ; /* Execute. */ - void (*exec_func) (void (*) (void *), size_t, void **, void **, size_t, int, + void (*exec_func) (void (*) (void *), size_t, void **, void **, int, unsigned *, void *); /* Async cleanup callback registration. */ Index: libgomp/plugin/plugin-nvptx.c =================================================================== --- libgomp/plugin/plugin-nvptx.c (revision 227269) +++ libgomp/plugin/plugin-nvptx.c (working copy) @@ -940,8 +940,7 @@ event_add (enum ptx_event_type type, CUe void nvptx_exec (void (*fn), size_t mapnum, void **hostaddrs, void **devaddrs, - size_t shared_size, int async, unsigned dims[GOMP_DIM_MAX], - void *targ_mem_desc) + int async, unsigned dims[GOMP_DIM_MAX], void *targ_mem_desc) { struct targ_fn_descriptor *targ_fn = (struct targ_fn_descriptor *) fn; CUfunction function; @@ -983,9 +982,9 @@ nvptx_exec (void (*fn), size_t mapnum, v GOMP_PLUGIN_fatal ("cuMemcpy failed: %s", cuda_error (r)); GOMP_PLUGIN_debug (0, " %s: kernel %s: launch" - " gangs=%u, workers=%u, vectors=%u, shared=%u\n", + " gangs=%u, workers=%u, vectors=%u\n", __FUNCTION__, targ_fn->launch->fn, - dims[0], dims[1], dims[2], (unsigned)shared_size); + dims[0], dims[1], dims[2]); // OpenACC CUDA // @@ -997,7 +996,7 @@ nvptx_exec (void (*fn), size_t mapnum, v r = cuLaunchKernel (function, dims[GOMP_DIM_GANG], 1, 1, dims[GOMP_DIM_VECTOR], dims[GOMP_DIM_WORKER], 1, - shared_size, dev_str->stream, kargs, 0); + 0, dev_str->stream, kargs, 0); if (r != CUDA_SUCCESS) GOMP_PLUGIN_fatal ("cuLaunchKernel error: %s", cuda_error (r)); @@ -1719,10 +1718,9 @@ void (*device_run) (int n, void *fn_ptr, void GOMP_OFFLOAD_openacc_parallel (void (*fn) (void *), size_t mapnum, void **hostaddrs, void **devaddrs, - size_t shared_size, int async, unsigned *dims, void *targ_mem_desc) { - nvptx_exec (fn, mapnum, hostaddrs, devaddrs, shared_size, + nvptx_exec (fn, mapnum, hostaddrs, devaddrs, async, dims, targ_mem_desc); } Index: libgomp/oacc-host.c =================================================================== --- libgomp/oacc-host.c (revision 227269) +++ libgomp/oacc-host.c (working copy) @@ -135,7 +135,6 @@ host_openacc_exec (void (*fn) (void *), size_t mapnum __attribute__ ((unused)), void **hostaddrs, void **devaddrs __attribute__ ((unused)), - size_t shared_size __attribute__ ((unused)), int async __attribute__ ((unused)), unsigned *dims __attribute__ ((unused)), void *targ_mem_desc __attribute__ ((unused))) Index: libgomp/oacc-parallel.c =================================================================== --- libgomp/oacc-parallel.c (revision 227269) +++ libgomp/oacc-parallel.c (working copy) @@ -57,31 +57,6 @@ find_pointer (int pos, size_t mapnum, un return 0; } -static void *__goacc_host_ganglocal_ptr; - -void * -GOACC_get_ganglocal_ptr (void) -{ - return __goacc_host_ganglocal_ptr; -} - -static void -alloc_host_shared_mem (size_t shared_size) -{ - if (shared_size > 0) - __goacc_host_ganglocal_ptr = malloc (shared_size); -} - -static void -free_host_shared_mem (void) -{ - if (__goacc_host_ganglocal_ptr) - { - free (__goacc_host_ganglocal_ptr); - __goacc_host_ganglocal_ptr = NULL; - } -} - static void alloc_ganglocal_addrs (size_t mapnum, void **hostaddrs, size_t *sizes, unsigned short *kinds) @@ -176,7 +151,7 @@ static void goacc_wait (int async, int n void GOACC_parallel_keyed (int device, void (*fn) (void *), size_t mapnum, void **hostaddrs, size_t *sizes, unsigned short *kinds, - size_t shared_size, ...) + ...) { bool host_fallback = device == GOMP_DEVICE_HOST_FALLBACK; va_list ap; @@ -194,15 +169,11 @@ GOACC_parallel_keyed (int device, void ( memset (dims, 0, sizeof (dims)); #ifdef HAVE_INTTYPES_H - gomp_debug (0, "%s: mapnum=%"PRIu64", hostaddrs=%p, sizes=%p, kinds=%p, " - "shared_size=%"PRIu64"\n", - __FUNCTION__, (uint64_t) mapnum, hostaddrs, sizes, kinds, - (uint64_t) shared_size); + gomp_debug (0, "%s: mapnum=%"PRIu64", hostaddrs=%p, sizes=%p, kinds=%p\n", + __FUNCTION__, (uint64_t) mapnum, hostaddrs, sizes, kinds); #else - gomp_debug (0, "%s: mapnum=%lu, hostaddrs=%p, sizes=%p, kinds=%p, " - "shared_size=%lu\n", - __FUNCTION__, (unsigned long) mapnum, hostaddrs, sizes, kinds, - (unsigned long) shared_size); + gomp_debug (0, "%s: mapnum=%lu, hostaddrs=%p, sizes=%p, kinds=%p\n", + __FUNCTION__, (unsigned long) mapnum, hostaddrs, sizes, kinds); #endif alloc_ganglocal_addrs (mapnum, hostaddrs, sizes, kinds); @@ -233,21 +204,17 @@ GOACC_parallel_keyed (int device, void ( if (host_fallback) { goacc_save_and_set_bind (acc_device_host); - alloc_host_shared_mem (shared_size); fn (hostaddrs); - free_host_shared_mem (); goacc_restore_bind (); return; } else if (acc_device_type (acc_dev->type) == acc_device_host) { - alloc_host_shared_mem (shared_size); fn (hostaddrs); - free_host_shared_mem (); return; } - va_start (ap, shared_size); + va_start (ap, kinds); /* TODO: This will need amending when device_type is implemented. */ while (GOMP_LAUNCH_PACK (GOMP_LAUNCH_END, 0, 0) != (tag = va_arg (ap, unsigned))) @@ -319,7 +286,7 @@ GOACC_parallel_keyed (int device, void ( } acc_dev->openacc.exec_func (tgt_fn, mapnum, hostaddrs, devaddrs, - shared_size, async, dims, tgt); + async, dims, tgt); /* If running synchronously, unmap immediately. */ if (async < acc_async_noval) @@ -339,7 +306,6 @@ void GOACC_parallel (int device, void (*fn) (void *), size_t mapnum, void **hostaddrs, size_t *sizes, unsigned short *kinds, int num_gangs, int num_workers, int vector_length, - size_t shared_size, int async, int num_waits, ...) { int waits[9]; @@ -356,7 +322,6 @@ GOACC_parallel (int device, void (*fn) ( waits[ix] = GOMP_LAUNCH_PACK (GOMP_LAUNCH_END, 0, 0); GOACC_parallel_keyed (device, fn, mapnum, hostaddrs, sizes, kinds, - shared_size, GOMP_LAUNCH_PACK (GOMP_LAUNCH_DIM, 0, GOMP_DIM_MASK (GOMP_DIM_MAX) - 1), num_gangs, num_workers, vector_length, Index: gcc/fortran/f95-lang.c =================================================================== --- gcc/fortran/f95-lang.c (revision 227269) +++ gcc/fortran/f95-lang.c (working copy) @@ -660,6 +660,8 @@ gfc_init_builtin_functions (void) ARG6, ARG7, ARG8) NAME, #define DEF_FUNCTION_TYPE_VAR_0(NAME, RETURN) NAME, #define DEF_FUNCTION_TYPE_VAR_2(NAME, RETURN, ARG1, ARG2) NAME, +#define DEF_FUNCTION_TYPE_VAR_6(NAME, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ + ARG6) NAME, #define DEF_FUNCTION_TYPE_VAR_7(NAME, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ ARG6, ARG7) NAME, #define DEF_POINTER_TYPE(NAME, TYPE) NAME, @@ -676,6 +678,7 @@ gfc_init_builtin_functions (void) #undef DEF_FUNCTION_TYPE_8 #undef DEF_FUNCTION_TYPE_VAR_0 #undef DEF_FUNCTION_TYPE_VAR_2 +#undef DEF_FUNCTION_TYPE_VAR_6 #undef DEF_FUNCTION_TYPE_VAR_7 #undef DEF_POINTER_TYPE BT_LAST @@ -1117,6 +1120,17 @@ gfc_init_builtin_functions (void) builtin_types[(int) ARG1], \ builtin_types[(int) ARG2], \ NULL_TREE); +#define DEF_FUNCTION_TYPE_VAR_6(ENUM, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ + ARG6) \ + builtin_types[(int) ENUM] \ + = build_varargs_function_type_list (builtin_types[(int) RETURN], \ + builtin_types[(int) ARG1], \ + builtin_types[(int) ARG2], \ + builtin_types[(int) ARG3], \ + builtin_types[(int) ARG4], \ + builtin_types[(int) ARG5], \ + builtin_types[(int) ARG6], \ + NULL_TREE); #define DEF_FUNCTION_TYPE_VAR_7(ENUM, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ ARG6, ARG7) \ builtin_types[(int) ENUM] \ Index: gcc/fortran/types.def =================================================================== --- gcc/fortran/types.def (revision 227269) +++ gcc/fortran/types.def (working copy) @@ -218,10 +218,11 @@ DEF_FUNCTION_TYPE_VAR_0 (BT_FN_VOID_VAR, DEF_FUNCTION_TYPE_VAR_2 (BT_FN_VOID_INT_INT_VAR, BT_VOID, BT_INT, BT_INT) +DEF_FUNCTION_TYPE_VAR_6 (BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_VAR, + BT_VOID, BT_INT, BT_PTR_FN_VOID_PTR, BT_SIZE, + BT_PTR, BT_PTR, BT_PTR) + DEF_FUNCTION_TYPE_VAR_7 (BT_FN_VOID_INT_SIZE_PTR_PTR_PTR_INT_INT_VAR, BT_VOID, BT_INT, BT_SIZE, BT_PTR, BT_PTR, BT_PTR, BT_INT, BT_INT) -DEF_FUNCTION_TYPE_VAR_7 (BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_SIZE_VAR, - BT_VOID, BT_INT, BT_PTR_FN_VOID_PTR, BT_SIZE, - BT_PTR, BT_PTR, BT_PTR, BT_SIZE) Index: gcc/lto/lto-lang.c =================================================================== --- gcc/lto/lto-lang.c (revision 227269) +++ gcc/lto/lto-lang.c (working copy) @@ -160,6 +160,8 @@ enum lto_builtin_type #define DEF_FUNCTION_TYPE_VAR_4(NAME, RETURN, ARG1, ARG2, ARG3, ARG4) NAME, #define DEF_FUNCTION_TYPE_VAR_5(NAME, RETURN, ARG1, ARG2, ARG3, ARG4, ARG6) \ NAME, +#define DEF_FUNCTION_TYPE_VAR_6(NAME, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ + ARG6) NAME, #define DEF_FUNCTION_TYPE_VAR_7(NAME, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ ARG6, ARG7) NAME, #define DEF_POINTER_TYPE(NAME, TYPE) NAME, @@ -180,6 +182,7 @@ enum lto_builtin_type #undef DEF_FUNCTION_TYPE_VAR_3 #undef DEF_FUNCTION_TYPE_VAR_4 #undef DEF_FUNCTION_TYPE_VAR_5 +#undef DEF_FUNCTION_TYPE_VAR_6 #undef DEF_FUNCTION_TYPE_VAR_7 #undef DEF_POINTER_TYPE BT_LAST @@ -665,6 +668,9 @@ lto_define_builtins (tree va_list_ref_ty def_fn_type (ENUM, RETURN, 1, 4, ARG1, ARG2, ARG3, ARG4); #define DEF_FUNCTION_TYPE_VAR_5(ENUM, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5) \ def_fn_type (ENUM, RETURN, 1, 5, ARG1, ARG2, ARG3, ARG4, ARG5); +#define DEF_FUNCTION_TYPE_VAR_6(ENUM, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ + ARG6) \ + def_fn_type (ENUM, RETURN, 1, 6, ARG1, ARG2, ARG3, ARG4, ARG5, ARG6); #define DEF_FUNCTION_TYPE_VAR_7(ENUM, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ ARG6, ARG7) \ def_fn_type (ENUM, RETURN, 1, 7, ARG1, ARG2, ARG3, ARG4, ARG5, ARG6, ARG7); @@ -689,6 +695,7 @@ lto_define_builtins (tree va_list_ref_ty #undef DEF_FUNCTION_TYPE_VAR_3 #undef DEF_FUNCTION_TYPE_VAR_4 #undef DEF_FUNCTION_TYPE_VAR_5 +#undef DEF_FUNCTION_TYPE_VAR_6 #undef DEF_FUNCTION_TYPE_VAR_7 #undef DEF_POINTER_TYPE builtin_types[(int) BT_LAST] = NULL_TREE; Index: gcc/omp-low.c =================================================================== --- gcc/omp-low.c (revision 227269) +++ gcc/omp-low.c (working copy) @@ -9949,8 +9949,6 @@ expand_omp_target (struct omp_region *re case BUILT_IN_GOACC_KERNELS_INTERNAL: case BUILT_IN_GOACC_PARALLEL: { - tree shared_memory = build_int_cst (integer_type_node, 0); - args.quick_push (shared_memory); set_oacc_fn_attrib (child_fn, clauses, &args); tagging = true; } Index: gcc/omp-builtins.def =================================================================== --- gcc/omp-builtins.def (revision 227269) +++ gcc/omp-builtins.def (working copy) @@ -45,11 +45,11 @@ DEF_GOACC_BUILTIN_FNSPEC (BUILT_IN_GOACC ATTR_NOTHROW_LIST, "...rrr") DEF_GOACC_BUILTIN_FNSPEC (BUILT_IN_GOACC_KERNELS_INTERNAL, "GOACC_kernels_internal", - BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_SIZE_VAR, - ATTR_FNSPEC_DOT_DOT_DOT_DOT_r_r_r_NOTHROW_LIST, - ATTR_NOTHROW_LIST, "....rrr") + BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_VAR, + ATTR_FNSPEC_DOT_DOT_DOT_r_r_r_NOTHROW_LIST, + ATTR_NOTHROW_LIST, "...rrr") DEF_GOACC_BUILTIN (BUILT_IN_GOACC_PARALLEL, "GOACC_parallel_keyed", - BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_SIZE_VAR, + BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_VAR, ATTR_NOTHROW_LIST) DEF_GOACC_BUILTIN_FNSPEC (BUILT_IN_GOACC_UPDATE, "GOACC_update", BT_FN_VOID_INT_SIZE_PTR_PTR_PTR_INT_INT_VAR, Index: gcc/c-family/c-common.c =================================================================== --- gcc/c-family/c-common.c (revision 227269) +++ gcc/c-family/c-common.c (working copy) @@ -5546,6 +5546,8 @@ enum c_builtin_type #define DEF_FUNCTION_TYPE_VAR_4(NAME, RETURN, ARG1, ARG2, ARG3, ARG4) NAME, #define DEF_FUNCTION_TYPE_VAR_5(NAME, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5) \ NAME, +#define DEF_FUNCTION_TYPE_VAR_6(NAME, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ + ARG6) NAME, #define DEF_FUNCTION_TYPE_VAR_7(NAME, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ ARG6, ARG7) NAME, #define DEF_POINTER_TYPE(NAME, TYPE) NAME, @@ -5566,6 +5568,7 @@ enum c_builtin_type #undef DEF_FUNCTION_TYPE_VAR_3 #undef DEF_FUNCTION_TYPE_VAR_4 #undef DEF_FUNCTION_TYPE_VAR_5 +#undef DEF_FUNCTION_TYPE_VAR_6 #undef DEF_FUNCTION_TYPE_VAR_7 #undef DEF_POINTER_TYPE BT_LAST @@ -5659,6 +5662,9 @@ c_define_builtins (tree va_list_ref_type def_fn_type (ENUM, RETURN, 1, 4, ARG1, ARG2, ARG3, ARG4); #define DEF_FUNCTION_TYPE_VAR_5(ENUM, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5) \ def_fn_type (ENUM, RETURN, 1, 5, ARG1, ARG2, ARG3, ARG4, ARG5); +#define DEF_FUNCTION_TYPE_VAR_6(ENUM, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ + ARG6) \ + def_fn_type (ENUM, RETURN, 1, 6, ARG1, ARG2, ARG3, ARG4, ARG5, ARG6); #define DEF_FUNCTION_TYPE_VAR_7(ENUM, RETURN, ARG1, ARG2, ARG3, ARG4, ARG5, \ ARG6, ARG7) \ def_fn_type (ENUM, RETURN, 1, 7, ARG1, ARG2, ARG3, ARG4, ARG5, ARG6, ARG7); @@ -5683,6 +5689,7 @@ c_define_builtins (tree va_list_ref_type #undef DEF_FUNCTION_TYPE_VAR_3 #undef DEF_FUNCTION_TYPE_VAR_4 #undef DEF_FUNCTION_TYPE_VAR_5 +#undef DEF_FUNCTION_TYPE_VAR_6 #undef DEF_FUNCTION_TYPE_VAR_7 #undef DEF_POINTER_TYPE builtin_types[(int) BT_LAST] = NULL_TREE; Index: gcc/builtin-types.def =================================================================== --- gcc/builtin-types.def (revision 227269) +++ gcc/builtin-types.def (working copy) @@ -592,14 +592,14 @@ DEF_FUNCTION_TYPE_VAR_5 (BT_FN_INT_STRIN DEF_FUNCTION_TYPE_VAR_5 (BT_FN_INT_INT_INT_INT_INT_INT_VAR, BT_INT, BT_INT, BT_INT, BT_INT, BT_INT, BT_INT) +DEF_FUNCTION_TYPE_VAR_6 (BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_VAR, + BT_VOID, BT_INT, BT_PTR_FN_VOID_PTR, BT_SIZE, + BT_PTR, BT_PTR, BT_PTR) + DEF_FUNCTION_TYPE_VAR_7 (BT_FN_VOID_INT_SIZE_PTR_PTR_PTR_INT_INT_VAR, BT_VOID, BT_INT, BT_SIZE, BT_PTR, BT_PTR, BT_PTR, BT_INT, BT_INT) -DEF_FUNCTION_TYPE_VAR_7 (BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_SIZE_VAR, - BT_VOID, BT_INT, BT_PTR_FN_VOID_PTR, BT_SIZE, - BT_PTR, BT_PTR, BT_PTR, BT_SIZE) - DEF_POINTER_TYPE (BT_PTR_FN_VOID_VAR, BT_FN_VOID_VAR) DEF_FUNCTION_TYPE_3 (BT_FN_PTR_PTR_FN_VOID_VAR_PTR_SIZE, BT_PTR, BT_PTR_FN_VOID_VAR, BT_PTR, BT_SIZE)