1 /* phsa.h -- interfacing between the gcc BRIG FE and the phsa runtime 2 Copyright (C) 2016-2020 Free Software Foundation, Inc. 3 Contributed by Pekka Jaaskelainen <pekka.jaaskelainen@parmance.com> 4 for General Processor Tech. 5 6 This file is part of GCC. 7 8 GCC is free software; you can redistribute it and/or modify it under 9 the terms of the GNU General Public License as published by the Free 10 Software Foundation; either version 3, or (at your option) any later 11 version. 12 13 GCC is distributed in the hope that it will be useful, but WITHOUT ANY 14 WARRANTY; without even the implied warranty of MERCHANTABILITY or 15 FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License 16 for more details. 17 18 You should have received a copy of the GNU General Public License 19 along with GCC; see the file COPYING3. If not see 20 <http://www.gnu.org/licenses/>. */ 21 22 #ifndef PHSA_H 23 #define PHSA_H 24 25 #include <stdint.h> 26 27 /* This struct is used to pass information from the BRIG FE to the 28 runtime of the finalizer kernel, its control directives etc. 29 The data is passed raw in a special ELF section named 30 phsa.kerneldesc.kernel_function_name. */ 31 32 typedef struct __attribute__((__packed__)) 33 { 34 /* Set to 1 in case the function is a kernel. */ 35 uint8_t is_kernel; 36 /* The size of the group segment used by the kernel. */ 37 uint32_t group_segment_size; 38 /* Size of the private segment used by a single work-item. */ 39 uint32_t private_segment_size; 40 /* Total size of the kernel arguments. */ 41 uint32_t kernarg_segment_size; 42 /* Maximum alignment of a kernel argument variable. */ 43 uint16_t kernarg_max_align; 44 /* Maximum size (in bytes) of dynamic group memory. */ 45 uint32_t max_dynamic_group_size; 46 /* Max number of work-items used to launch the kernel. */ 47 uint64_t max_flat_grid_size; 48 /* Max number of work-items in a work-group used to launch the kernel. */ 49 uint32_t max_flat_workgroup_size; 50 /* The grid size required by the kernel. */ 51 uint64_t required_grid_size[3]; 52 /* The work group size required by the kernel. */ 53 uint32_t required_workgroup_size[3]; 54 /* The number of dimensions required by the kernel. */ 55 uint8_t required_dim; 56 57 } phsa_descriptor; 58 59 /* The prefix to use in the ELF section containing descriptor for 60 a function. */ 61 62 #define PHSA_DESC_SECTION_PREFIX "phsa.desc." 63 #define PHSA_HOST_DEF_PTR_PREFIX "__phsa.host_def." 64 65 /* The frontend error messages are parsed by the host runtime. Known 66 prefix strings are used to separate the different runtime error 67 codes. */ 68 69 #define PHSA_ERROR_PREFIX_INCOMPATIBLE_MODULE "Incompatible module: " 70 #define PHSA_ERROR_PREFIX_CORRUPTED_MODULE "Corrupted module: " 71 72 /* Offsets of attributes in the PHSA context structs. 73 Used by -fphsa-wi-context-opt. */ 74 #define PHSA_CONTEXT_OFFS_WI_IDS 0 75 #define PHSA_CONTEXT_OFFS_WG_IDS (PHSA_CONTEXT_OFFS_WI_IDS + 3 * 4) 76 #define PHSA_CONTEXT_WG_SIZES (PHSA_CONTEXT_OFFS_WG_IDS + 3 * 4) 77 #define PHSA_CONTEXT_CURRENT_WG_SIZES (PHSA_CONTEXT_WG_SIZES + 3 * 4) 78 79 #endif 80