gcc/brig/brigfrontend/phsa.h

   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