CFD/hypre
1
0
Fork 0
Code Issues Pull Requests Packages Projects Releases Wiki Activity
3e56e6e590
hypre/src/struct_mv/_hypre_struct_mv.h
Victor A. P. Magri 3e56e6e590
Fix MSVC build (#978) 
* Allocate buffer on heap memory
* Fix Pragma definition for MSVC
* Fix uninitiliazed variable
* Loop counter cannot be non-negative for MSVC
2023-10-10 08:28:24 -04:00

2736 lines
137 KiB
C
Raw Blame History

/*** DO NOT EDIT THIS FILE DIRECTLY (use 'headers' to generate) ***/
#ifndef hypre_STRUCT_MV_HEADER
#define hypre_STRUCT_MV_HEADER
#include <stdlib.h>
#include <stdio.h>
#include <math.h>
#include "HYPRE_struct_mv.h"
#include "_hypre_utilities.h"
#ifdef __cplusplus
extern "C" {
#endif
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/******************************************************************************
*
* Header info for the Box structures
*
*****************************************************************************/
#ifndef hypre_BOX_HEADER
#define hypre_BOX_HEADER
#ifndef HYPRE_MAXDIM
#define HYPRE_MAXDIM 3
#endif
/*--------------------------------------------------------------------------
* hypre_Index:
* This is used to define indices in index space, or dimension
* sizes of boxes.
*
* The spatial dimensions x, y, and z may be specified by the
* integers 0, 1, and 2, respectively (see the hypre_IndexD macro below).
* This simplifies the code in the hypre_Box class by reducing code
* replication.
*--------------------------------------------------------------------------*/
typedef HYPRE_Int hypre_Index[HYPRE_MAXDIM];
typedef HYPRE_Int *hypre_IndexRef;
/*--------------------------------------------------------------------------
* hypre_Box:
*--------------------------------------------------------------------------*/
typedef struct hypre_Box_struct
{
hypre_Index imin; /* min bounding indices */
hypre_Index imax; /* max bounding indices */
HYPRE_Int ndim; /* number of dimensions */
} hypre_Box;
/*--------------------------------------------------------------------------
* hypre_BoxArray:
* An array of boxes.
* Since size can be zero, need to store ndim separately.
*--------------------------------------------------------------------------*/
typedef struct hypre_BoxArray_struct
{
hypre_Box *boxes; /* Array of boxes */
HYPRE_Int size; /* Size of box array */
HYPRE_Int alloc_size; /* Size of currently alloced space */
HYPRE_Int ndim; /* number of dimensions */
} hypre_BoxArray;
#define hypre_BoxArrayExcess 10
/*--------------------------------------------------------------------------
* hypre_BoxArrayArray:
* An array of box arrays.
* Since size can be zero, need to store ndim separately.
*--------------------------------------------------------------------------*/
typedef struct hypre_BoxArrayArray_struct
{
hypre_BoxArray **box_arrays; /* Array of pointers to box arrays */
HYPRE_Int size; /* Size of box array array */
HYPRE_Int ndim; /* number of dimensions */
} hypre_BoxArrayArray;
/*--------------------------------------------------------------------------
* Accessor macros: hypre_Index
*--------------------------------------------------------------------------*/
#define hypre_IndexD(index, d) (index[d])
/* Avoid using these macros */
#define hypre_IndexX(index) hypre_IndexD(index, 0)
#define hypre_IndexY(index) hypre_IndexD(index, 1)
#define hypre_IndexZ(index) hypre_IndexD(index, 2)
/*--------------------------------------------------------------------------
* Member functions: hypre_Index
*--------------------------------------------------------------------------*/
/*----- Avoid using these Index macros -----*/
#define hypre_SetIndex3(index, ix, iy, iz) \
( hypre_IndexD(index, 0) = ix,\
hypre_IndexD(index, 1) = iy,\
hypre_IndexD(index, 2) = iz )
#define hypre_ClearIndex(index) hypre_SetIndex(index, 0)
/*--------------------------------------------------------------------------
* Accessor macros: hypre_Box
*--------------------------------------------------------------------------*/
#define hypre_BoxIMin(box) ((box) -> imin)
#define hypre_BoxIMax(box) ((box) -> imax)
#define hypre_BoxNDim(box) ((box) -> ndim)
#define hypre_BoxIMinD(box, d) (hypre_IndexD(hypre_BoxIMin(box), d))
#define hypre_BoxIMaxD(box, d) (hypre_IndexD(hypre_BoxIMax(box), d))
#define hypre_BoxSizeD(box, d) \
hypre_max(0, (hypre_BoxIMaxD(box, d) - hypre_BoxIMinD(box, d) + 1))
#define hypre_IndexDInBox(index, d, box) \
( hypre_IndexD(index, d) >= hypre_BoxIMinD(box, d) && \
hypre_IndexD(index, d) <= hypre_BoxIMaxD(box, d) )
/* The first hypre_CCBoxIndexRank is better style because it is similar to
hypre_BoxIndexRank. The second one sometimes avoids compiler warnings. */
#define hypre_CCBoxIndexRank(box, index) 0
#define hypre_CCBoxIndexRank_noargs() 0
#define hypre_CCBoxOffsetDistance(box, index) 0
/*----- Avoid using these Box macros -----*/
#define hypre_BoxSizeX(box) hypre_BoxSizeD(box, 0)
#define hypre_BoxSizeY(box) hypre_BoxSizeD(box, 1)
#define hypre_BoxSizeZ(box) hypre_BoxSizeD(box, 2)
/*--------------------------------------------------------------------------
* Accessor macros: hypre_BoxArray
*--------------------------------------------------------------------------*/
#define hypre_BoxArrayBoxes(box_array) ((box_array) -> boxes)
#define hypre_BoxArrayBox(box_array, i) &((box_array) -> boxes[(i)])
#define hypre_BoxArraySize(box_array) ((box_array) -> size)
#define hypre_BoxArrayAllocSize(box_array) ((box_array) -> alloc_size)
#define hypre_BoxArrayNDim(box_array) ((box_array) -> ndim)
/*--------------------------------------------------------------------------
* Accessor macros: hypre_BoxArrayArray
*--------------------------------------------------------------------------*/
#define hypre_BoxArrayArrayBoxArrays(box_array_array) \
((box_array_array) -> box_arrays)
#define hypre_BoxArrayArrayBoxArray(box_array_array, i) \
((box_array_array) -> box_arrays[(i)])
#define hypre_BoxArrayArraySize(box_array_array) \
((box_array_array) -> size)
#define hypre_BoxArrayArrayNDim(box_array_array) \
((box_array_array) -> ndim)
/*--------------------------------------------------------------------------
* Looping macros:
*--------------------------------------------------------------------------*/
#define hypre_ForBoxI(i, box_array) \
for (i = 0; i < hypre_BoxArraySize(box_array); i++)
#define hypre_ForBoxArrayI(i, box_array_array) \
for (i = 0; i < hypre_BoxArrayArraySize(box_array_array); i++)
#define ZYPRE_BOX_PRIVATE hypre__IN,hypre__JN,hypre__I,hypre__J,hypre__d,hypre__i
#define HYPRE_BOX_PRIVATE ZYPRE_BOX_PRIVATE
#define zypre_BoxLoopDeclare() \
HYPRE_Int hypre__tot, hypre__div, hypre__mod;\
HYPRE_Int hypre__block, hypre__num_blocks;\
HYPRE_Int hypre__d, hypre__ndim;\
HYPRE_Int hypre__I, hypre__J, hypre__IN, hypre__JN;\
HYPRE_Int hypre__i[HYPRE_MAXDIM+1], hypre__n[HYPRE_MAXDIM+1]
#define zypre_BoxLoopDeclareK(k) \
HYPRE_Int hypre__ikstart##k, hypre__i0inc##k;\
HYPRE_Int hypre__sk##k[HYPRE_MAXDIM], hypre__ikinc##k[HYPRE_MAXDIM+1]
#define zypre_BoxLoopInit(ndim, loop_size) \
hypre__ndim = ndim;\
hypre__n[0] = loop_size[0];\
hypre__tot = 1;\
for (hypre__d = 1; hypre__d < hypre__ndim; hypre__d++)\
{\
hypre__n[hypre__d] = loop_size[hypre__d];\
hypre__tot *= hypre__n[hypre__d];\
}\
hypre__n[hypre__ndim] = 2;\
hypre__num_blocks = hypre_NumThreads();\
if (hypre__tot < hypre__num_blocks)\
{\
hypre__num_blocks = hypre__tot;\
}\
if (hypre__num_blocks > 0)\
{\
hypre__div = hypre__tot / hypre__num_blocks;\
hypre__mod = hypre__tot % hypre__num_blocks;\
}
#define zypre_BoxLoopInitK(k, dboxk, startk, stridek, ik) \
hypre__sk##k[0] = stridek[0];\
hypre__ikinc##k[0] = 0;\
ik = hypre_BoxSizeD(dboxk, 0); /* temporarily use ik */\
for (hypre__d = 1; hypre__d < hypre__ndim; hypre__d++)\
{\
hypre__sk##k[hypre__d] = ik*stridek[hypre__d];\
hypre__ikinc##k[hypre__d] = hypre__ikinc##k[hypre__d-1] +\
hypre__sk##k[hypre__d] - hypre__n[hypre__d-1]*hypre__sk##k[hypre__d-1];\
ik *= hypre_BoxSizeD(dboxk, hypre__d);\
}\
hypre__i0inc##k = hypre__sk##k[0];\
hypre__ikinc##k[hypre__ndim] = 0;\
hypre__ikstart##k = hypre_BoxIndexRank(dboxk, startk)
#define zypre_BoxLoopSet() \
hypre__IN = hypre__n[0];\
if (hypre__num_blocks > 1)/* in case user sets num_blocks to 1 */\
{\
hypre__JN = hypre__div + ((hypre__mod > hypre__block) ? 1 : 0);\
hypre__J = hypre__block * hypre__div + hypre_min(hypre__mod, hypre__block);\
for (hypre__d = 1; hypre__d < hypre__ndim; hypre__d++)\
{\
hypre__i[hypre__d] = hypre__J % hypre__n[hypre__d];\
hypre__J /= hypre__n[hypre__d];\
}\
}\
else\
{\
hypre__JN = hypre__tot;\
for (hypre__d = 1; hypre__d < hypre__ndim; hypre__d++)\
{\
hypre__i[hypre__d] = 0;\
}\
}\
hypre__i[hypre__ndim] = 0
#define zypre_BoxLoopSetK(k, ik) \
ik = hypre__ikstart##k;\
for (hypre__d = 1; hypre__d < hypre__ndim; hypre__d++)\
{\
ik += hypre__i[hypre__d]*hypre__sk##k[hypre__d];\
}
#define zypre_BoxLoopInc1() \
hypre__d = 1;\
while ((hypre__i[hypre__d]+2) > hypre__n[hypre__d])\
{\
hypre__d++;\
}
#define zypre_BoxLoopInc2() \
hypre__i[hypre__d]++;\
while (hypre__d > 1)\
{\
hypre__d--;\
hypre__i[hypre__d] = 0;\
}
/* This returns the loop index (of type hypre_Index) for the current iteration,
* where the numbering starts at 0. It works even when threading is turned on,
* as long as 'index' is declared to be private. */
#define zypre_BoxLoopGetIndex(index) \
index[0] = hypre__I;\
for (hypre__d = 1; hypre__d < hypre__ndim; hypre__d++)\
{\
index[hypre__d] = hypre__i[hypre__d];\
}
/* Use this before the For macros below to force only one block */
#define zypre_BoxLoopSetOneBlock() hypre__num_blocks = 1
/* Use this to get the block iteration inside a BoxLoop */
#define zypre_BoxLoopBlock() hypre__block
#define zypre_BasicBoxLoopInitK(k, stridek) \
hypre__sk##k[0] = stridek[0];\
hypre__ikinc##k[0] = 0;\
for (hypre__d = 1; hypre__d < hypre__ndim; hypre__d++)\
{\
hypre__sk##k[hypre__d] = stridek[hypre__d];\
hypre__ikinc##k[hypre__d] = hypre__ikinc##k[hypre__d-1] +\
hypre__sk##k[hypre__d] - hypre__n[hypre__d-1]*hypre__sk##k[hypre__d-1];\
}\
hypre__i0inc##k = hypre__sk##k[0];\
hypre__ikinc##k[hypre__ndim] = 0;\
hypre__ikstart##k = 0
/*--------------------------------------------------------------------------
* NOTES - Keep these for reference here and elsewhere in the code
*--------------------------------------------------------------------------*/
#if 0
#define hypre_BoxLoop2Begin(loop_size,
dbox1, start1, stride1, i1,
dbox2, start2, stride2, i2)
{
/* init hypre__i1start */
HYPRE_Int hypre__i1start = hypre_BoxIndexRank(dbox1, start1);
HYPRE_Int hypre__i2start = hypre_BoxIndexRank(dbox2, start2);
/* declare and set hypre__s1 */
hypre_BoxLoopDeclareS(dbox1, stride1, hypre__sx1, hypre__sy1, hypre__sz1);
hypre_BoxLoopDeclareS(dbox2, stride2, hypre__sx2, hypre__sy2, hypre__sz2);
/* declare and set hypre__n, hypre__m, hypre__dir, hypre__max,
* hypre__div, hypre__mod, hypre__block, hypre__num_blocks */
hypre_BoxLoopDeclareN(loop_size);
#define hypre_BoxLoop2For(i, j, k, i1, i2)
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++)
{
/* set i and hypre__n */
hypre_BoxLoopSet(i, j, k);
/* set i1 */
i1 = hypre__i1start + i * hypre__sx1 + j * hypre__sy1 + k * hypre__sz1;
i2 = hypre__i2start + i * hypre__sx2 + j * hypre__sy2 + k * hypre__sz2;
for (k = 0; k < hypre__nz; k++)
{
for (j = 0; j < hypre__ny; j++)
{
for (i = 0; i < hypre__nx; i++)
{
#define hypre_BoxLoop2End(i1, i2)
i1 += hypre__sx1;
i2 += hypre__sx2;
}
i1 += hypre__sy1 - hypre__nx * hypre__sx1;
i2 += hypre__sy2 - hypre__nx * hypre__sx2;
}
i1 += hypre__sz1 - hypre__ny * hypre__sy1;
i2 += hypre__sz2 - hypre__ny * hypre__sy2;
}
}
}
/*----------------------------------------
* Idea 2: Simple version of Idea 3 below
*----------------------------------------*/
N = 1;
for (d = 0; d < ndim; d++)
{
N *= n[d];
i[d] = 0;
n[d] -= 2; /* this produces a simpler comparison below */
}
i[ndim] = 0;
n[ndim] = 0;
for (I = 0; I < N; I++)
{
/* loop body */
for (d = 0; i[d] > n[d]; d++)
{
i[d] = 0;
}
i[d]++;
i1 += s1[d]; /* NOTE: These are different from hypre__sx1, etc. above */
i2 += s2[d]; /* The lengths of i, n, and s must be (ndim+1) */
}
/*----------------------------------------
* Idea 3: Approach used in the box loops
*----------------------------------------*/
N = 1;
for (d = 1; d < ndim; d++)
{
N *= n[d];
i[d] = 0;
n[d] -= 2; /* this produces a simpler comparison below */
}
i[ndim] = 0;
n[ndim] = 0;
for (J = 0; J < N; J++)
{
for (I = 0; I < n[0]; I++)
{
/* loop body */
i1 += s1[0];
i2 += s2[0];
}
for (d = 1; i[d] > n[d]; d++)
{
i[d] = 0;
}
i[d]++;
i1 += s1[d]; /* NOTE: These are different from hypre__sx1, etc. above */
i2 += s2[d]; /* The lengths of i, n, and s must be (ndim+1) */
}
#endif
#endif /* #ifndef hypre_BOX_HEADER */
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/******************************************************************************
*
* Header info for the struct assumed partition
*
*****************************************************************************/
#ifndef hypre_ASSUMED_PART_HEADER
#define hypre_ASSUMED_PART_HEADER
typedef struct
{
/* the entries will be the same for all procs */
HYPRE_Int ndim; /* number of dimensions */
hypre_BoxArray *regions; /* areas of the grid with boxes */
HYPRE_Int num_regions; /* how many regions */
HYPRE_Int *proc_partitions; /* proc ids assigned to each region
(this is size num_regions +1) */
hypre_Index *divisions; /* number of proc divisions in each
direction for each region */
/* these entries are specific to each proc */
hypre_BoxArray *my_partition; /* my portion of grid (at most 2) */
hypre_BoxArray *my_partition_boxes; /* boxes in my portion */
HYPRE_Int *my_partition_proc_ids;
HYPRE_Int *my_partition_boxnums;
HYPRE_Int my_partition_ids_size;
HYPRE_Int my_partition_ids_alloc;
HYPRE_Int my_partition_num_distinct_procs;
} hypre_StructAssumedPart;
/*Accessor macros */
#define hypre_StructAssumedPartNDim(apart) ((apart)->ndim)
#define hypre_StructAssumedPartRegions(apart) ((apart)->regions)
#define hypre_StructAssumedPartNumRegions(apart) ((apart)->num_regions)
#define hypre_StructAssumedPartDivisions(apart) ((apart)->divisions)
#define hypre_StructAssumedPartDivision(apart, i) ((apart)->divisions[i])
#define hypre_StructAssumedPartProcPartitions(apart) ((apart)->proc_partitions)
#define hypre_StructAssumedPartProcPartition(apart, i) ((apart)->proc_partitions[i])
#define hypre_StructAssumedPartMyPartition(apart) ((apart)->my_partition)
#define hypre_StructAssumedPartMyPartitionBoxes(apart) ((apart)->my_partition_boxes)
#define hypre_StructAssumedPartMyPartitionProcIds(apart) ((apart)->my_partition_proc_ids)
#define hypre_StructAssumedPartMyPartitionIdsSize(apart) ((apart)->my_partition_ids_size)
#define hypre_StructAssumedPartMyPartitionIdsAlloc(apart) ((apart)->my_partition_ids_alloc)
#define hypre_StructAssumedPartMyPartitionNumDistinctProcs(apart) ((apart)->my_partition_num_distinct_procs)
#define hypre_StructAssumedPartMyPartitionBoxnums(apart) ((apart)->my_partition_boxnums)
#define hypre_StructAssumedPartMyPartitionProcId(apart, i) ((apart)->my_partition_proc_ids[i])
#define hypre_StructAssumedPartMyPartitionBoxnum(apart, i) ((apart)->my_partition_boxnums[i])
#endif
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
#ifndef hypre_BOX_MANAGER_HEADER
#define hypre_BOX_MANAGER_HEADER
/*--------------------------------------------------------------------------
* BoxManEntry
*--------------------------------------------------------------------------*/
typedef struct hypre_BoxManEntry_struct
{
hypre_Index imin; /* Extents of box */
hypre_Index imax;
HYPRE_Int ndim; /* Number of dimensions */
HYPRE_Int proc; /* This is a two-part unique id: (proc, id) */
HYPRE_Int id;
HYPRE_Int num_ghost[2 * HYPRE_MAXDIM];
HYPRE_Int position; /* This indicates the location of the entry in the the
* box manager entries array and is used for pairing with
* the info object (populated in addentry) */
void *boxman; /* The owning manager (populated in addentry) */
struct hypre_BoxManEntry_struct *next;
} hypre_BoxManEntry;
/*---------------------------------------------------------------------------
* Box Manager: organizes arbitrary information in a spatial way
*----------------------------------------------------------------------------*/
typedef struct
{
MPI_Comm comm;
HYPRE_Int max_nentries; /* storage allocated for entries */
HYPRE_Int is_gather_called; /* Boolean to indicate whether
GatherEntries function has been
called (prior to assemble) - may not
want this (can tell by the size of
gather_regions array) */
hypre_BoxArray *gather_regions; /* This is where we collect boxes input
by calls to BoxManGatherEntries - to
be gathered in the assemble. These
are then deleted after the assemble */
HYPRE_Int all_global_known; /* Boolean to say that every processor
already has all of the global data
for this manager (this could be
accessed by a coarsening routine,
for example) */
HYPRE_Int is_entries_sort; /* Boolean to say that entries were
added in sorted order (id, proc)
(this could be accessed by a
coarsening routine, for example) */
HYPRE_Int entry_info_size; /* In bytes, the (max) size of the info
object for the entries */
HYPRE_Int is_assembled; /* Flag to indicate if the box manager has
been assembled (used to control whether
or not functions can be used prior to
assemble) */
/* Storing the entries */
HYPRE_Int nentries; /* Number of entries stored */
hypre_BoxManEntry *entries; /* Actual box manager entries - sorted by
(proc, id) at the end of the assemble) */
HYPRE_Int *procs_sort; /* The sorted procs corresponding to entries */
HYPRE_Int *ids_sort; /* Sorted ids corresponding to the entries */
HYPRE_Int num_procs_sort; /* Number of distinct procs in entries */
HYPRE_Int *procs_sort_offsets; /* Offsets for procs into the
entry_sort array */
HYPRE_Int first_local; /* Position of local infomation in entries */
HYPRE_Int local_proc_offset; /* Position of local information in
offsets */
/* Here is the table that organizes the entries spatially (by index) */
hypre_BoxManEntry **index_table; /* This points into 'entries' array and
corresponds to the index arays */
HYPRE_Int *indexes[HYPRE_MAXDIM]; /* Indexes (ordered) for imin and
imax of each box in the entries
array */
HYPRE_Int size[HYPRE_MAXDIM]; /* How many indexes in each
direction */
HYPRE_Int last_index[HYPRE_MAXDIM]; /* Last index used in the
indexes map */
HYPRE_Int num_my_entries; /* Num entries with proc_id = myid */
HYPRE_Int *my_ids; /* Array of ids corresponding to my entries */
hypre_BoxManEntry **my_entries; /* Points into entries that are mine and
corresponds to my_ids array. This is
destroyed in the assemble. */
void *info_objects; /* Array of info objects (each of size
entry_info_size), managed byte-wise */
hypre_StructAssumedPart *assumed_partition; /* The assumed partition object.
For now this is only used
during the assemble (where it
is created). */
HYPRE_Int ndim; /* Problem dimension (known in the grid) */
hypre_Box *bounding_box; /* Bounding box from associated grid */
HYPRE_Int next_id; /* Counter to indicate the next id that would be
unique (regardless of proc id) */
/* Ghost stuff */
HYPRE_Int num_ghost[2 * HYPRE_MAXDIM];
} hypre_BoxManager;
/*--------------------------------------------------------------------------
* Accessor macros: hypre_BoxMan
*--------------------------------------------------------------------------*/
#define hypre_BoxManComm(manager) ((manager) -> comm)
#define hypre_BoxManMaxNEntries(manager) ((manager) -> max_nentries)
#define hypre_BoxManIsGatherCalled(manager) ((manager) -> is_gather_called)
#define hypre_BoxManIsEntriesSort(manager) ((manager) -> is_entries_sort)
#define hypre_BoxManGatherRegions(manager) ((manager) -> gather_regions)
#define hypre_BoxManAllGlobalKnown(manager) ((manager) -> all_global_known)
#define hypre_BoxManEntryInfoSize(manager) ((manager) -> entry_info_size)
#define hypre_BoxManNEntries(manager) ((manager) -> nentries)
#define hypre_BoxManEntries(manager) ((manager) -> entries)
#define hypre_BoxManInfoObjects(manager) ((manager) -> info_objects)
#define hypre_BoxManIsAssembled(manager) ((manager) -> is_assembled)
#define hypre_BoxManProcsSort(manager) ((manager) -> procs_sort)
#define hypre_BoxManIdsSort(manager) ((manager) -> ids_sort)
#define hypre_BoxManNumProcsSort(manager) ((manager) -> num_procs_sort)
#define hypre_BoxManProcsSortOffsets(manager) ((manager) -> procs_sort_offsets)
#define hypre_BoxManLocalProcOffset(manager) ((manager) -> local_proc_offset)
#define hypre_BoxManFirstLocal(manager) ((manager) -> first_local)
#define hypre_BoxManIndexTable(manager) ((manager) -> index_table)
#define hypre_BoxManIndexes(manager) ((manager) -> indexes)
#define hypre_BoxManSize(manager) ((manager) -> size)
#define hypre_BoxManLastIndex(manager) ((manager) -> last_index)
#define hypre_BoxManNumMyEntries(manager) ((manager) -> num_my_entries)
#define hypre_BoxManMyIds(manager) ((manager) -> my_ids)
#define hypre_BoxManMyEntries(manager) ((manager) -> my_entries)
#define hypre_BoxManAssumedPartition(manager) ((manager) -> assumed_partition)
#define hypre_BoxManNDim(manager) ((manager) -> ndim)
#define hypre_BoxManBoundingBox(manager) ((manager) -> bounding_box)
#define hypre_BoxManNextId(manager) ((manager) -> next_id)
#define hypre_BoxManNumGhost(manager) ((manager) -> num_ghost)
#define hypre_BoxManIndexesD(manager, d) hypre_BoxManIndexes(manager)[d]
#define hypre_BoxManSizeD(manager, d) hypre_BoxManSize(manager)[d]
#define hypre_BoxManLastIndexD(manager, d) hypre_BoxManLastIndex(manager)[d]
#define hypre_BoxManInfoObject(manager, i) \
(void *) ((char *)hypre_BoxManInfoObjects(manager) + i* hypre_BoxManEntryInfoSize(manager))
/*--------------------------------------------------------------------------
* Accessor macros: hypre_BoxManEntry
*--------------------------------------------------------------------------*/
#define hypre_BoxManEntryIMin(entry) ((entry) -> imin)
#define hypre_BoxManEntryIMax(entry) ((entry) -> imax)
#define hypre_BoxManEntryNDim(entry) ((entry) -> ndim)
#define hypre_BoxManEntryProc(entry) ((entry) -> proc)
#define hypre_BoxManEntryId(entry) ((entry) -> id)
#define hypre_BoxManEntryPosition(entry) ((entry) -> position)
#define hypre_BoxManEntryNumGhost(entry) ((entry) -> num_ghost)
#define hypre_BoxManEntryNext(entry) ((entry) -> next)
#define hypre_BoxManEntryBoxMan(entry) ((entry) -> boxman)
#endif
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/******************************************************************************
*
* Header info for the hypre_StructGrid structures
*
*****************************************************************************/
#ifndef hypre_STRUCT_GRID_HEADER
#define hypre_STRUCT_GRID_HEADER
/*--------------------------------------------------------------------------
* hypre_StructGrid:
*--------------------------------------------------------------------------*/
typedef struct hypre_StructGrid_struct
{
MPI_Comm comm;
HYPRE_Int ndim; /* Number of grid dimensions */
hypre_BoxArray *boxes; /* Array of boxes in this process */
HYPRE_Int *ids; /* Unique IDs for boxes */
hypre_Index max_distance; /* Neighborhood size - in each dimension*/
hypre_Box *bounding_box; /* Bounding box around grid */
HYPRE_Int local_size; /* Number of grid points locally */
HYPRE_BigInt global_size; /* Total number of grid points */
hypre_Index periodic; /* Indicates if grid is periodic */
HYPRE_Int num_periods; /* number of box set periods */
hypre_Index *pshifts; /* shifts of periodicity */
HYPRE_Int ref_count;
HYPRE_Int ghlocal_size; /* Number of vars in box including ghosts */
HYPRE_Int num_ghost[2 * HYPRE_MAXDIM]; /* ghost layer size */
hypre_BoxManager *boxman;
} hypre_StructGrid;
/*--------------------------------------------------------------------------
* Accessor macros: hypre_StructGrid
*--------------------------------------------------------------------------*/
#define hypre_StructGridComm(grid) ((grid) -> comm)
#define hypre_StructGridNDim(grid) ((grid) -> ndim)
#define hypre_StructGridBoxes(grid) ((grid) -> boxes)
#define hypre_StructGridIDs(grid) ((grid) -> ids)
#define hypre_StructGridMaxDistance(grid) ((grid) -> max_distance)
#define hypre_StructGridBoundingBox(grid) ((grid) -> bounding_box)
#define hypre_StructGridLocalSize(grid) ((grid) -> local_size)
#define hypre_StructGridGlobalSize(grid) ((grid) -> global_size)
#define hypre_StructGridPeriodic(grid) ((grid) -> periodic)
#define hypre_StructGridNumPeriods(grid) ((grid) -> num_periods)
#define hypre_StructGridPShifts(grid) ((grid) -> pshifts)
#define hypre_StructGridPShift(grid, i) ((grid) -> pshifts[i])
#define hypre_StructGridRefCount(grid) ((grid) -> ref_count)
#define hypre_StructGridGhlocalSize(grid) ((grid) -> ghlocal_size)
#define hypre_StructGridNumGhost(grid) ((grid) -> num_ghost)
#define hypre_StructGridBoxMan(grid) ((grid) -> boxman)
#define hypre_StructGridBox(grid, i) (hypre_BoxArrayBox(hypre_StructGridBoxes(grid), i))
#define hypre_StructGridNumBoxes(grid) (hypre_BoxArraySize(hypre_StructGridBoxes(grid)))
#define hypre_StructGridIDPeriod(grid) hypre_BoxNeighborsIDPeriod(hypre_StructGridNeighbors(grid))
#if 0 //defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP)
#define hypre_StructGridDataLocation(grid) ((grid) -> data_location)
#endif
/*--------------------------------------------------------------------------
* Looping macros:
*--------------------------------------------------------------------------*/
#define hypre_ForStructGridBoxI(i, grid) hypre_ForBoxI(i, hypre_StructGridBoxes(grid))
#if 0 //defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP)
#define HYPRE_MIN_GPU_SIZE (131072)
#define hypre_SetDeviceOn() hypre_HandleStructExecPolicy(hypre_handle()) = HYPRE_EXEC_DEVICE
#define hypre_SetDeviceOff() hypre_HandleStructExecPolicy(hypre_handle()) = HYPRE_EXEC_HOST
#endif
#endif
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/******************************************************************************
*
* Header info for hypre_StructStencil data structures
*
*****************************************************************************/
#ifndef hypre_STRUCT_STENCIL_HEADER
#define hypre_STRUCT_STENCIL_HEADER
/*--------------------------------------------------------------------------
* hypre_StructStencil
*--------------------------------------------------------------------------*/
typedef struct hypre_StructStencil_struct
{
hypre_Index *shape; /* Description of a stencil's shape */
HYPRE_Int size; /* Number of stencil coefficients */
HYPRE_Int ndim; /* Number of dimensions */
HYPRE_Int ref_count;
} hypre_StructStencil;
/*--------------------------------------------------------------------------
* Accessor functions for the hypre_StructStencil structure
*--------------------------------------------------------------------------*/
#define hypre_StructStencilShape(stencil) ((stencil) -> shape)
#define hypre_StructStencilSize(stencil) ((stencil) -> size)
#define hypre_StructStencilNDim(stencil) ((stencil) -> ndim)
#define hypre_StructStencilRefCount(stencil) ((stencil) -> ref_count)
#define hypre_StructStencilElement(stencil, i) hypre_StructStencilShape(stencil)[i]
#endif
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
#ifndef hypre_COMMUNICATION_HEADER
#define hypre_COMMUNICATION_HEADER
/*--------------------------------------------------------------------------
* hypre_CommInfo:
*
* For "reverse" communication, send_transforms is not needed (may be NULL).
* For "forward" communication, recv_transforms is not needed (may be NULL).
*--------------------------------------------------------------------------*/
typedef struct hypre_CommInfo_struct
{
HYPRE_Int ndim;
hypre_BoxArrayArray *send_boxes;
hypre_Index send_stride;
HYPRE_Int **send_processes;
HYPRE_Int **send_rboxnums;
hypre_BoxArrayArray *send_rboxes; /* send_boxes, some with periodic shift */
hypre_BoxArrayArray *recv_boxes;
hypre_Index recv_stride;
HYPRE_Int **recv_processes;
HYPRE_Int **recv_rboxnums;
hypre_BoxArrayArray *recv_rboxes; /* recv_boxes, some with periodic shift */
HYPRE_Int num_transforms; /* may be 0 = identity transform */
hypre_Index *coords; /* may be NULL = identity transform */
hypre_Index *dirs; /* may be NULL = identity transform */
HYPRE_Int **send_transforms; /* may be NULL = identity transform */
HYPRE_Int **recv_transforms; /* may be NULL = identity transform */
HYPRE_Int boxes_match; /* true (>0) if each send box has a
* matching box on the recv processor */
} hypre_CommInfo;
/*--------------------------------------------------------------------------
* hypre_CommEntryType:
*--------------------------------------------------------------------------*/
typedef struct hypre_CommEntryType_struct
{
HYPRE_Int offset; /* offset for the data */
HYPRE_Int dim; /* dimension of the communication */
HYPRE_Int length_array[HYPRE_MAXDIM]; /* last dim has length num_values */
HYPRE_Int stride_array[HYPRE_MAXDIM + 1];
HYPRE_Int *order; /* order of last dim values */
} hypre_CommEntryType;
/*--------------------------------------------------------------------------
* hypre_CommType:
*--------------------------------------------------------------------------*/
typedef struct hypre_CommType_struct
{
HYPRE_Int proc;
HYPRE_Int bufsize; /* message buffer size (in doubles) */
HYPRE_Int num_entries;
hypre_CommEntryType *entries;
/* this is only needed until first send buffer prefix is packed */
HYPRE_Int *rem_boxnums; /* entry remote box numbers */
hypre_Box *rem_boxes; /* entry remote boxes */
} hypre_CommType;
/*--------------------------------------------------------------------------
* hypre_CommPkg:
* Structure containing information for doing communications
*--------------------------------------------------------------------------*/
typedef struct hypre_CommPkg_struct
{
MPI_Comm comm;
/* is this the first communication? */
HYPRE_Int first_comm;
HYPRE_Int ndim;
HYPRE_Int num_values;
hypre_Index send_stride;
hypre_Index recv_stride;
/* total send buffer size (in doubles) */
HYPRE_Int send_bufsize;
/* total recv buffer size (in doubles) */
HYPRE_Int recv_bufsize;
/* total send buffer size (in doubles) at the first comm. */
HYPRE_Int send_bufsize_first_comm;
/* total recv buffer size (in doubles) at the first comm. */
HYPRE_Int recv_bufsize_first_comm;
HYPRE_Int num_sends;
HYPRE_Int num_recvs;
hypre_CommType *send_types;
hypre_CommType *recv_types;
hypre_CommType *copy_from_type;
hypre_CommType *copy_to_type;
/* these pointers are just to help free up memory for send/from types */
hypre_CommEntryType *entries;
HYPRE_Int *rem_boxnums;
hypre_Box *rem_boxes;
HYPRE_Int num_orders;
/* num_orders x num_values */
HYPRE_Int **orders;
/* offsets into recv data (by box) */
HYPRE_Int *recv_data_offsets;
/* recv data dimensions (by box) */
hypre_BoxArray *recv_data_space;
hypre_Index identity_coord;
hypre_Index identity_dir;
HYPRE_Int *identity_order;
} hypre_CommPkg;
/*--------------------------------------------------------------------------
* CommHandle:
*--------------------------------------------------------------------------*/
typedef struct hypre_CommHandle_struct
{
hypre_CommPkg *comm_pkg;
HYPRE_Complex *send_data;
HYPRE_Complex *recv_data;
HYPRE_Int num_requests;
hypre_MPI_Request *requests;
hypre_MPI_Status *status;
HYPRE_Complex **send_buffers;
HYPRE_Complex **recv_buffers;
HYPRE_Complex **send_buffers_mpi;
HYPRE_Complex **recv_buffers_mpi;
/* set = 0, add = 1 */
HYPRE_Int action;
} hypre_CommHandle;
/*--------------------------------------------------------------------------
* Accessor macros: hypre_CommInto
*--------------------------------------------------------------------------*/
#define hypre_CommInfoNDim(info) (info -> ndim)
#define hypre_CommInfoSendBoxes(info) (info -> send_boxes)
#define hypre_CommInfoSendStride(info) (info -> send_stride)
#define hypre_CommInfoSendProcesses(info) (info -> send_processes)
#define hypre_CommInfoSendRBoxnums(info) (info -> send_rboxnums)
#define hypre_CommInfoSendRBoxes(info) (info -> send_rboxes)
#define hypre_CommInfoRecvBoxes(info) (info -> recv_boxes)
#define hypre_CommInfoRecvStride(info) (info -> recv_stride)
#define hypre_CommInfoRecvProcesses(info) (info -> recv_processes)
#define hypre_CommInfoRecvRBoxnums(info) (info -> recv_rboxnums)
#define hypre_CommInfoRecvRBoxes(info) (info -> recv_rboxes)
#define hypre_CommInfoNumTransforms(info) (info -> num_transforms)
#define hypre_CommInfoCoords(info) (info -> coords)
#define hypre_CommInfoDirs(info) (info -> dirs)
#define hypre_CommInfoSendTransforms(info) (info -> send_transforms)
#define hypre_CommInfoRecvTransforms(info) (info -> recv_transforms)
#define hypre_CommInfoBoxesMatch(info) (info -> boxes_match)
/*--------------------------------------------------------------------------
* Accessor macros: hypre_CommEntryType
*--------------------------------------------------------------------------*/
#define hypre_CommEntryTypeOffset(entry) (entry -> offset)
#define hypre_CommEntryTypeDim(entry) (entry -> dim)
#define hypre_CommEntryTypeLengthArray(entry) (entry -> length_array)
#define hypre_CommEntryTypeStrideArray(entry) (entry -> stride_array)
#define hypre_CommEntryTypeOrder(entry) (entry -> order)
/*--------------------------------------------------------------------------
* Accessor macros: hypre_CommType
*--------------------------------------------------------------------------*/
#define hypre_CommTypeProc(type) (type -> proc)
#define hypre_CommTypeBufsize(type) (type -> bufsize)
#define hypre_CommTypeNumEntries(type) (type -> num_entries)
#define hypre_CommTypeEntries(type) (type -> entries)
#define hypre_CommTypeEntry(type, i) (&(type -> entries[i]))
#define hypre_CommTypeRemBoxnums(type) (type -> rem_boxnums)
#define hypre_CommTypeRemBoxnum(type, i) (type -> rem_boxnums[i])
#define hypre_CommTypeRemBoxes(type) (type -> rem_boxes)
#define hypre_CommTypeRemBox(type, i) (&(type -> rem_boxes[i]))
/*--------------------------------------------------------------------------
* Accessor macros: hypre_CommPkg
*--------------------------------------------------------------------------*/
#define hypre_CommPkgComm(comm_pkg) (comm_pkg -> comm)
#define hypre_CommPkgFirstComm(comm_pkg) (comm_pkg -> first_comm)
#define hypre_CommPkgNDim(comm_pkg) (comm_pkg -> ndim)
#define hypre_CommPkgNumValues(comm_pkg) (comm_pkg -> num_values)
#define hypre_CommPkgSendStride(comm_pkg) (comm_pkg -> send_stride)
#define hypre_CommPkgRecvStride(comm_pkg) (comm_pkg -> recv_stride)
#define hypre_CommPkgSendBufsize(comm_pkg) (comm_pkg -> send_bufsize)
#define hypre_CommPkgRecvBufsize(comm_pkg) (comm_pkg -> recv_bufsize)
#define hypre_CommPkgSendBufsizeFirstComm(comm_pkg) (comm_pkg -> send_bufsize_first_comm)
#define hypre_CommPkgRecvBufsizeFirstComm(comm_pkg) (comm_pkg -> recv_bufsize_first_comm)
#define hypre_CommPkgNumSends(comm_pkg) (comm_pkg -> num_sends)
#define hypre_CommPkgNumRecvs(comm_pkg) (comm_pkg -> num_recvs)
#define hypre_CommPkgSendTypes(comm_pkg) (comm_pkg -> send_types)
#define hypre_CommPkgSendType(comm_pkg, i) (&(comm_pkg -> send_types[i]))
#define hypre_CommPkgRecvTypes(comm_pkg) (comm_pkg -> recv_types)
#define hypre_CommPkgRecvType(comm_pkg, i) (&(comm_pkg -> recv_types[i]))
#define hypre_CommPkgCopyFromType(comm_pkg) (comm_pkg -> copy_from_type)
#define hypre_CommPkgCopyToType(comm_pkg) (comm_pkg -> copy_to_type)
#define hypre_CommPkgEntries(comm_pkg) (comm_pkg -> entries)
#define hypre_CommPkgRemBoxnums(comm_pkg) (comm_pkg -> rem_boxnums)
#define hypre_CommPkgRemBoxes(comm_pkg) (comm_pkg -> rem_boxes)
#define hypre_CommPkgNumOrders(comm_pkg) (comm_pkg -> num_orders)
#define hypre_CommPkgOrders(comm_pkg) (comm_pkg -> orders)
#define hypre_CommPkgRecvDataOffsets(comm_pkg) (comm_pkg -> recv_data_offsets)
#define hypre_CommPkgRecvDataSpace(comm_pkg) (comm_pkg -> recv_data_space)
#define hypre_CommPkgIdentityCoord(comm_pkg) (comm_pkg -> identity_coord)
#define hypre_CommPkgIdentityDir(comm_pkg) (comm_pkg -> identity_dir)
#define hypre_CommPkgIdentityOrder(comm_pkg) (comm_pkg -> identity_order)
/*--------------------------------------------------------------------------
* Accessor macros: hypre_CommHandle
*--------------------------------------------------------------------------*/
#define hypre_CommHandleCommPkg(comm_handle) (comm_handle -> comm_pkg)
#define hypre_CommHandleSendData(comm_handle) (comm_handle -> send_data)
#define hypre_CommHandleRecvData(comm_handle) (comm_handle -> recv_data)
#define hypre_CommHandleNumRequests(comm_handle) (comm_handle -> num_requests)
#define hypre_CommHandleRequests(comm_handle) (comm_handle -> requests)
#define hypre_CommHandleStatus(comm_handle) (comm_handle -> status)
#define hypre_CommHandleSendBuffers(comm_handle) (comm_handle -> send_buffers)
#define hypre_CommHandleRecvBuffers(comm_handle) (comm_handle -> recv_buffers)
#define hypre_CommHandleAction(comm_handle) (comm_handle -> action)
#define hypre_CommHandleSendBuffersMPI(comm_handle) (comm_handle -> send_buffers_mpi)
#define hypre_CommHandleRecvBuffersMPI(comm_handle) (comm_handle -> recv_buffers_mpi)
#endif
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/******************************************************************************
*
* Header info for computation
*
*****************************************************************************/
#ifndef hypre_COMPUTATION_HEADER
#define hypre_COMPUTATION_HEADER
/*--------------------------------------------------------------------------
* hypre_ComputeInfo:
*--------------------------------------------------------------------------*/
typedef struct hypre_ComputeInfo_struct
{
hypre_CommInfo *comm_info;
hypre_BoxArrayArray *indt_boxes;
hypre_BoxArrayArray *dept_boxes;
hypre_Index stride;
} hypre_ComputeInfo;
/*--------------------------------------------------------------------------
* hypre_ComputePkg:
* Structure containing information for doing computations.
*--------------------------------------------------------------------------*/
typedef struct hypre_ComputePkg_struct
{
hypre_CommPkg *comm_pkg;
hypre_BoxArrayArray *indt_boxes;
hypre_BoxArrayArray *dept_boxes;
hypre_Index stride;
hypre_StructGrid *grid;
hypre_BoxArray *data_space;
HYPRE_Int num_values;
} hypre_ComputePkg;
/*--------------------------------------------------------------------------
* Accessor macros: hypre_ComputeInfo
*--------------------------------------------------------------------------*/
#define hypre_ComputeInfoCommInfo(info) (info -> comm_info)
#define hypre_ComputeInfoIndtBoxes(info) (info -> indt_boxes)
#define hypre_ComputeInfoDeptBoxes(info) (info -> dept_boxes)
#define hypre_ComputeInfoStride(info) (info -> stride)
/*--------------------------------------------------------------------------
* Accessor macros: hypre_ComputePkg
*--------------------------------------------------------------------------*/
#define hypre_ComputePkgCommPkg(compute_pkg) (compute_pkg -> comm_pkg)
#define hypre_ComputePkgIndtBoxes(compute_pkg) (compute_pkg -> indt_boxes)
#define hypre_ComputePkgDeptBoxes(compute_pkg) (compute_pkg -> dept_boxes)
#define hypre_ComputePkgStride(compute_pkg) (compute_pkg -> stride)
#define hypre_ComputePkgGrid(compute_pkg) (compute_pkg -> grid)
#define hypre_ComputePkgDataSpace(compute_pkg) (compute_pkg -> data_space)
#define hypre_ComputePkgNumValues(compute_pkg) (compute_pkg -> num_values)
#endif
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/******************************************************************************
*
* Header info for the hypre_StructMatrix structures
*
*****************************************************************************/
#ifndef hypre_STRUCT_MATRIX_HEADER
#define hypre_STRUCT_MATRIX_HEADER
/*--------------------------------------------------------------------------
* hypre_StructMatrix:
*--------------------------------------------------------------------------*/
typedef struct hypre_StructMatrix_struct
{
MPI_Comm comm;
hypre_StructGrid *grid;
hypre_StructStencil *user_stencil;
hypre_StructStencil *stencil;
HYPRE_Int num_values; /* Number of "stored" coefficients */
hypre_BoxArray *data_space;
HYPRE_MemoryLocation memory_location; /* memory location of data */
HYPRE_Complex *data; /* Pointer to variable matrix data */
HYPRE_Complex *data_const; /* Pointer to constant matrix data */
HYPRE_Complex **stencil_data; /* Pointer for each stencil */
HYPRE_Int data_alloced; /* Boolean used for freeing data */
HYPRE_Int data_size; /* Size of variable matrix data */
HYPRE_Int data_const_size; /* Size of constant matrix data */
HYPRE_Int **data_indices; /* num-boxes by stencil-size array
of indices into the data array.
data_indices[b][s] is the starting
index of matrix data corresponding
to box b and stencil coefficient s */
HYPRE_Int constant_coefficient; /* normally 0; set to 1 for
constant coefficient matrices
or 2 for constant coefficient
with variable diagonal */
HYPRE_Int symmetric; /* Is the matrix symmetric */
HYPRE_Int *symm_elements; /* Which elements are "symmetric" */
HYPRE_Int num_ghost[2 * HYPRE_MAXDIM]; /* Num ghost layers in each direction */
HYPRE_BigInt global_size; /* Total number of nonzero coeffs */
hypre_CommPkg *comm_pkg; /* Info on how to update ghost data */
HYPRE_Int ref_count;
} hypre_StructMatrix;
/*--------------------------------------------------------------------------
* Accessor macros: hypre_StructMatrix
*--------------------------------------------------------------------------*/
#define hypre_StructMatrixComm(matrix) ((matrix) -> comm)
#define hypre_StructMatrixGrid(matrix) ((matrix) -> grid)
#define hypre_StructMatrixUserStencil(matrix) ((matrix) -> user_stencil)
#define hypre_StructMatrixStencil(matrix) ((matrix) -> stencil)
#define hypre_StructMatrixNumValues(matrix) ((matrix) -> num_values)
#define hypre_StructMatrixDataSpace(matrix) ((matrix) -> data_space)
#define hypre_StructMatrixMemoryLocation(matrix) ((matrix) -> memory_location)
#define hypre_StructMatrixData(matrix) ((matrix) -> data)
#define hypre_StructMatrixDataConst(matrix) ((matrix) -> data_const)
#define hypre_StructMatrixStencilData(matrix) ((matrix) -> stencil_data)
#define hypre_StructMatrixDataAlloced(matrix) ((matrix) -> data_alloced)
#define hypre_StructMatrixDataSize(matrix) ((matrix) -> data_size)
#define hypre_StructMatrixDataConstSize(matrix) ((matrix) -> data_const_size)
#define hypre_StructMatrixDataIndices(matrix) ((matrix) -> data_indices)
#define hypre_StructMatrixConstantCoefficient(matrix) ((matrix) -> constant_coefficient)
#define hypre_StructMatrixSymmetric(matrix) ((matrix) -> symmetric)
#define hypre_StructMatrixSymmElements(matrix) ((matrix) -> symm_elements)
#define hypre_StructMatrixNumGhost(matrix) ((matrix) -> num_ghost)
#define hypre_StructMatrixGlobalSize(matrix) ((matrix) -> global_size)
#define hypre_StructMatrixCommPkg(matrix) ((matrix) -> comm_pkg)
#define hypre_StructMatrixRefCount(matrix) ((matrix) -> ref_count)
#define hypre_StructMatrixNDim(matrix) \
hypre_StructGridNDim(hypre_StructMatrixGrid(matrix))
#define hypre_StructMatrixBox(matrix, b) \
hypre_BoxArrayBox(hypre_StructMatrixDataSpace(matrix), b)
#define hypre_StructMatrixBoxData(matrix, b, s) \
(hypre_StructMatrixStencilData(matrix)[s] + hypre_StructMatrixDataIndices(matrix)[b][s])
#define hypre_StructMatrixBoxDataValue(matrix, b, s, index) \
(hypre_StructMatrixBoxData(matrix, b, s) + \
hypre_BoxIndexRank(hypre_StructMatrixBox(matrix, b), index))
#define hypre_CCStructMatrixBoxDataValue(matrix, b, s, index) \
(hypre_StructMatrixBoxData(matrix, b, s) + \
hypre_CCBoxIndexRank(hypre_StructMatrixBox(matrix, b), index))
#endif
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/******************************************************************************
*
* Header info for the hypre_StructVector structures
*
*****************************************************************************/
#ifndef hypre_STRUCT_VECTOR_HEADER
#define hypre_STRUCT_VECTOR_HEADER
/*--------------------------------------------------------------------------
* hypre_StructVector:
*--------------------------------------------------------------------------*/
typedef struct hypre_StructVector_struct
{
MPI_Comm comm;
hypre_StructGrid *grid;
hypre_BoxArray *data_space;
HYPRE_MemoryLocation memory_location; /* memory location of data */
HYPRE_Complex *data; /* Pointer to vector data on device*/
HYPRE_Int data_alloced; /* Boolean used for freeing data */
HYPRE_Int data_size; /* Size of vector data */
HYPRE_Int *data_indices; /* num-boxes array of indices into
the data array. data_indices[b]
is the starting index of vector
data corresponding to box b. */
HYPRE_Int num_ghost[2 * HYPRE_MAXDIM]; /* Num ghost layers in each
* direction */
HYPRE_Int bghost_not_clear; /* Are boundary ghosts clear? */
HYPRE_BigInt global_size; /* Total number coefficients */
HYPRE_Int ref_count;
} hypre_StructVector;
/*--------------------------------------------------------------------------
* Accessor macros: hypre_StructVector
*--------------------------------------------------------------------------*/
#define hypre_StructVectorComm(vector) ((vector) -> comm)
#define hypre_StructVectorGrid(vector) ((vector) -> grid)
#define hypre_StructVectorDataSpace(vector) ((vector) -> data_space)
#define hypre_StructVectorMemoryLocation(vector) ((vector) -> memory_location)
#define hypre_StructVectorData(vector) ((vector) -> data)
#define hypre_StructVectorDataAlloced(vector) ((vector) -> data_alloced)
#define hypre_StructVectorDataSize(vector) ((vector) -> data_size)
#define hypre_StructVectorDataIndices(vector) ((vector) -> data_indices)
#define hypre_StructVectorNumGhost(vector) ((vector) -> num_ghost)
#define hypre_StructVectorBGhostNotClear(vector) ((vector) -> bghost_not_clear)
#define hypre_StructVectorGlobalSize(vector) ((vector) -> global_size)
#define hypre_StructVectorRefCount(vector) ((vector) -> ref_count)
#define hypre_StructVectorNDim(vector) \
hypre_StructGridNDim(hypre_StructVectorGrid(vector))
#define hypre_StructVectorBox(vector, b) \
hypre_BoxArrayBox(hypre_StructVectorDataSpace(vector), b)
#define hypre_StructVectorBoxData(vector, b) \
(hypre_StructVectorData(vector) + hypre_StructVectorDataIndices(vector)[b])
#define hypre_StructVectorBoxDataValue(vector, b, index) \
(hypre_StructVectorBoxData(vector, b) + \
hypre_BoxIndexRank(hypre_StructVectorBox(vector, b), index))
#endif
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/* assumed_part.c */
HYPRE_Int hypre_APSubdivideRegion ( hypre_Box *region, HYPRE_Int dim, HYPRE_Int level,
hypre_BoxArray *box_array, HYPRE_Int *num_new_boxes );
HYPRE_Int hypre_APFindMyBoxesInRegions ( hypre_BoxArray *region_array, hypre_BoxArray *my_box_array,
HYPRE_Int **p_count_array, HYPRE_Real **p_vol_array );
HYPRE_Int hypre_APGetAllBoxesInRegions ( hypre_BoxArray *region_array, hypre_BoxArray *my_box_array,
HYPRE_Int **p_count_array, HYPRE_Real **p_vol_array, MPI_Comm comm );
HYPRE_Int hypre_APShrinkRegions ( hypre_BoxArray *region_array, hypre_BoxArray *my_box_array,
MPI_Comm comm );
HYPRE_Int hypre_APPruneRegions ( hypre_BoxArray *region_array, HYPRE_Int **p_count_array,
HYPRE_Real **p_vol_array );
HYPRE_Int hypre_APRefineRegionsByVol ( hypre_BoxArray *region_array, HYPRE_Real *vol_array,
HYPRE_Int max_regions, HYPRE_Real gamma, HYPRE_Int dim, HYPRE_Int *return_code, MPI_Comm comm );
HYPRE_Int hypre_StructAssumedPartitionCreate ( HYPRE_Int dim, hypre_Box *bounding_box,
HYPRE_Real global_boxes_size, HYPRE_Int global_num_boxes, hypre_BoxArray *local_boxes,
HYPRE_Int *local_boxnums, HYPRE_Int max_regions, HYPRE_Int max_refinements, HYPRE_Real gamma,
MPI_Comm comm, hypre_StructAssumedPart **p_assumed_partition );
HYPRE_Int hypre_StructAssumedPartitionDestroy ( hypre_StructAssumedPart *assumed_part );
HYPRE_Int hypre_APFillResponseStructAssumedPart ( void *p_recv_contact_buf, HYPRE_Int contact_size,
HYPRE_Int contact_proc, void *ro, MPI_Comm comm, void **p_send_response_buf,
HYPRE_Int *response_message_size );
HYPRE_Int hypre_StructAssumedPartitionGetRegionsFromProc ( hypre_StructAssumedPart *assumed_part,
HYPRE_Int proc_id, hypre_BoxArray *assumed_regions );
HYPRE_Int hypre_StructAssumedPartitionGetProcsFromBox ( hypre_StructAssumedPart *assumed_part,
hypre_Box *box, HYPRE_Int *num_proc_array, HYPRE_Int *size_alloc_proc_array,
HYPRE_Int **p_proc_array );
/* box_algebra.c */
HYPRE_Int hypre_IntersectBoxes ( hypre_Box *box1, hypre_Box *box2, hypre_Box *ibox );
HYPRE_Int hypre_SubtractBoxes ( hypre_Box *box1, hypre_Box *box2, hypre_BoxArray *box_array );
HYPRE_Int hypre_SubtractBoxArrays ( hypre_BoxArray *box_array1, hypre_BoxArray *box_array2,
hypre_BoxArray *tmp_box_array );
HYPRE_Int hypre_UnionBoxes ( hypre_BoxArray *boxes );
HYPRE_Int hypre_MinUnionBoxes ( hypre_BoxArray *boxes );
/* box_boundary.c */
HYPRE_Int hypre_BoxBoundaryIntersect ( hypre_Box *box, hypre_StructGrid *grid, HYPRE_Int d,
HYPRE_Int dir, hypre_BoxArray *boundary );
HYPRE_Int hypre_BoxBoundaryG ( hypre_Box *box, hypre_StructGrid *g, hypre_BoxArray *boundary );
HYPRE_Int hypre_BoxBoundaryDG ( hypre_Box *box, hypre_StructGrid *g, hypre_BoxArray *boundarym,
hypre_BoxArray *boundaryp, HYPRE_Int d );
HYPRE_Int hypre_GeneralBoxBoundaryIntersect( hypre_Box *box, hypre_StructGrid *grid,
hypre_Index stencil_element, hypre_BoxArray *boundary );
/* box.c */
HYPRE_Int hypre_SetIndex ( hypre_Index index, HYPRE_Int val );
HYPRE_Int hypre_CopyIndex( hypre_Index in_index, hypre_Index out_index );
HYPRE_Int hypre_CopyToCleanIndex( hypre_Index in_index, HYPRE_Int ndim, hypre_Index out_index );
HYPRE_Int hypre_IndexEqual ( hypre_Index index, HYPRE_Int val, HYPRE_Int ndim );
HYPRE_Int hypre_IndexMin( hypre_Index index, HYPRE_Int ndim );
HYPRE_Int hypre_IndexMax( hypre_Index index, HYPRE_Int ndim );
HYPRE_Int hypre_AddIndexes ( hypre_Index index1, hypre_Index index2, HYPRE_Int ndim,
hypre_Index result );
HYPRE_Int hypre_SubtractIndexes ( hypre_Index index1, hypre_Index index2, HYPRE_Int ndim,
hypre_Index result );
HYPRE_Int hypre_IndexesEqual ( hypre_Index index1, hypre_Index index2, HYPRE_Int ndim );
HYPRE_Int hypre_IndexPrint ( FILE *file, HYPRE_Int ndim, hypre_Index index );
HYPRE_Int hypre_IndexRead ( FILE *file, HYPRE_Int ndim, hypre_Index index );
hypre_Box *hypre_BoxCreate ( HYPRE_Int ndim );
HYPRE_Int hypre_BoxDestroy ( hypre_Box *box );
HYPRE_Int hypre_BoxInit( hypre_Box *box, HYPRE_Int ndim );
HYPRE_Int hypre_BoxSetExtents ( hypre_Box *box, hypre_Index imin, hypre_Index imax );
HYPRE_Int hypre_CopyBox( hypre_Box *box1, hypre_Box *box2 );
hypre_Box *hypre_BoxDuplicate ( hypre_Box *box );
HYPRE_Int hypre_BoxVolume( hypre_Box *box );
HYPRE_Real hypre_doubleBoxVolume( hypre_Box *box );
HYPRE_Int hypre_IndexInBox ( hypre_Index index, hypre_Box *box );
HYPRE_Int hypre_BoxGetSize ( hypre_Box *box, hypre_Index size );
HYPRE_Int hypre_BoxGetStrideSize ( hypre_Box *box, hypre_Index stride, hypre_Index size );
HYPRE_Int hypre_BoxGetStrideVolume ( hypre_Box *box, hypre_Index stride, HYPRE_Int *volume_ptr );
HYPRE_Int hypre_BoxIndexRank( hypre_Box *box, hypre_Index index );
HYPRE_Int hypre_BoxRankIndex( hypre_Box *box, HYPRE_Int rank, hypre_Index index );
HYPRE_Int hypre_BoxOffsetDistance( hypre_Box *box, hypre_Index index );
HYPRE_Int hypre_BoxShiftPos( hypre_Box *box, hypre_Index shift );
HYPRE_Int hypre_BoxShiftNeg( hypre_Box *box, hypre_Index shift );
HYPRE_Int hypre_BoxGrowByIndex( hypre_Box *box, hypre_Index index );
HYPRE_Int hypre_BoxGrowByValue( hypre_Box *box, HYPRE_Int val );
HYPRE_Int hypre_BoxGrowByArray ( hypre_Box *box, HYPRE_Int *array );
HYPRE_Int hypre_BoxPrint ( FILE *file, hypre_Box *box );
HYPRE_Int hypre_BoxRead ( FILE *file, HYPRE_Int ndim, hypre_Box **box_ptr );
hypre_BoxArray *hypre_BoxArrayCreate ( HYPRE_Int size, HYPRE_Int ndim );
HYPRE_Int hypre_BoxArrayDestroy ( hypre_BoxArray *box_array );
HYPRE_Int hypre_BoxArraySetSize ( hypre_BoxArray *box_array, HYPRE_Int size );
hypre_BoxArray *hypre_BoxArrayDuplicate ( hypre_BoxArray *box_array );
HYPRE_Int hypre_AppendBox ( hypre_Box *box, hypre_BoxArray *box_array );
HYPRE_Int hypre_DeleteBox ( hypre_BoxArray *box_array, HYPRE_Int index );
HYPRE_Int hypre_DeleteMultipleBoxes ( hypre_BoxArray *box_array, HYPRE_Int *indices,
HYPRE_Int num );
HYPRE_Int hypre_AppendBoxArray ( hypre_BoxArray *box_array_0, hypre_BoxArray *box_array_1 );
hypre_BoxArrayArray *hypre_BoxArrayArrayCreate ( HYPRE_Int size, HYPRE_Int ndim );
HYPRE_Int hypre_BoxArrayArrayDestroy ( hypre_BoxArrayArray *box_array_array );
hypre_BoxArrayArray *hypre_BoxArrayArrayDuplicate ( hypre_BoxArrayArray *box_array_array );
/* box_manager.c */
HYPRE_Int hypre_BoxManEntryGetInfo ( hypre_BoxManEntry *entry, void **info_ptr );
HYPRE_Int hypre_BoxManEntryGetExtents ( hypre_BoxManEntry *entry, hypre_Index imin,
hypre_Index imax );
HYPRE_Int hypre_BoxManEntryCopy ( hypre_BoxManEntry *fromentry, hypre_BoxManEntry *toentry );
HYPRE_Int hypre_BoxManSetAllGlobalKnown ( hypre_BoxManager *manager, HYPRE_Int known );
HYPRE_Int hypre_BoxManGetAllGlobalKnown ( hypre_BoxManager *manager, HYPRE_Int *known );
HYPRE_Int hypre_BoxManSetIsEntriesSort ( hypre_BoxManager *manager, HYPRE_Int is_sort );
HYPRE_Int hypre_BoxManGetIsEntriesSort ( hypre_BoxManager *manager, HYPRE_Int *is_sort );
HYPRE_Int hypre_BoxManGetGlobalIsGatherCalled ( hypre_BoxManager *manager, MPI_Comm comm,
HYPRE_Int *is_gather );
HYPRE_Int hypre_BoxManGetAssumedPartition ( hypre_BoxManager *manager,
hypre_StructAssumedPart **assumed_partition );
HYPRE_Int hypre_BoxManSetAssumedPartition ( hypre_BoxManager *manager,
hypre_StructAssumedPart *assumed_partition );
HYPRE_Int hypre_BoxManSetBoundingBox ( hypre_BoxManager *manager, hypre_Box *bounding_box );
HYPRE_Int hypre_BoxManSetNumGhost ( hypre_BoxManager *manager, HYPRE_Int *num_ghost );
HYPRE_Int hypre_BoxManDeleteMultipleEntriesAndInfo ( hypre_BoxManager *manager, HYPRE_Int *indices,
HYPRE_Int num );
HYPRE_Int hypre_BoxManCreate ( HYPRE_Int max_nentries, HYPRE_Int info_size, HYPRE_Int dim,
hypre_Box *bounding_box, MPI_Comm comm, hypre_BoxManager **manager_ptr );
HYPRE_Int hypre_BoxManIncSize ( hypre_BoxManager *manager, HYPRE_Int inc_size );
HYPRE_Int hypre_BoxManDestroy ( hypre_BoxManager *manager );
HYPRE_Int hypre_BoxManAddEntry ( hypre_BoxManager *manager, hypre_Index imin, hypre_Index imax,
HYPRE_Int proc_id, HYPRE_Int box_id, void *info );
HYPRE_Int hypre_BoxManGetEntry ( hypre_BoxManager *manager, HYPRE_Int proc, HYPRE_Int id,
hypre_BoxManEntry **entry_ptr );
HYPRE_Int hypre_BoxManGetAllEntries ( hypre_BoxManager *manager, HYPRE_Int *num_entries,
hypre_BoxManEntry **entries );
HYPRE_Int hypre_BoxManGetAllEntriesBoxes ( hypre_BoxManager *manager, hypre_BoxArray *boxes );
HYPRE_Int hypre_BoxManGetLocalEntriesBoxes ( hypre_BoxManager *manager, hypre_BoxArray *boxes );
HYPRE_Int hypre_BoxManGetAllEntriesBoxesProc ( hypre_BoxManager *manager, hypre_BoxArray *boxes,
HYPRE_Int **procs_ptr );
HYPRE_Int hypre_BoxManGatherEntries ( hypre_BoxManager *manager, hypre_Index imin,
hypre_Index imax );
HYPRE_Int hypre_BoxManAssemble ( hypre_BoxManager *manager );
HYPRE_Int hypre_BoxManIntersect ( hypre_BoxManager *manager, hypre_Index ilower, hypre_Index iupper,
hypre_BoxManEntry ***entries_ptr, HYPRE_Int *nentries_ptr );
HYPRE_Int hypre_FillResponseBoxManAssemble1 ( void *p_recv_contact_buf, HYPRE_Int contact_size,
HYPRE_Int contact_proc, void *ro, MPI_Comm comm, void **p_send_response_buf,
HYPRE_Int *response_message_size );
HYPRE_Int hypre_FillResponseBoxManAssemble2 ( void *p_recv_contact_buf, HYPRE_Int contact_size,
HYPRE_Int contact_proc, void *ro, MPI_Comm comm, void **p_send_response_buf,
HYPRE_Int *response_message_size );
/* communication_info.c */
HYPRE_Int hypre_CommInfoCreate ( hypre_BoxArrayArray *send_boxes, hypre_BoxArrayArray *recv_boxes,
HYPRE_Int **send_procs, HYPRE_Int **recv_procs, HYPRE_Int **send_rboxnums,
HYPRE_Int **recv_rboxnums, hypre_BoxArrayArray *send_rboxes, hypre_BoxArrayArray *recv_rboxes,
HYPRE_Int boxes_match, hypre_CommInfo **comm_info_ptr );
HYPRE_Int hypre_CommInfoSetTransforms ( hypre_CommInfo *comm_info, HYPRE_Int num_transforms,
hypre_Index *coords, hypre_Index *dirs, HYPRE_Int **send_transforms, HYPRE_Int **recv_transforms );
HYPRE_Int hypre_CommInfoGetTransforms ( hypre_CommInfo *comm_info, HYPRE_Int *num_transforms,
hypre_Index **coords, hypre_Index **dirs );
HYPRE_Int hypre_CommInfoProjectSend ( hypre_CommInfo *comm_info, hypre_Index index,
hypre_Index stride );
HYPRE_Int hypre_CommInfoProjectRecv ( hypre_CommInfo *comm_info, hypre_Index index,
hypre_Index stride );
HYPRE_Int hypre_CommInfoDestroy ( hypre_CommInfo *comm_info );
HYPRE_Int hypre_CreateCommInfoFromStencil ( hypre_StructGrid *grid, hypre_StructStencil *stencil,
hypre_CommInfo **comm_info_ptr );
HYPRE_Int hypre_CreateCommInfoFromNumGhost ( hypre_StructGrid *grid, HYPRE_Int *num_ghost,
hypre_CommInfo **comm_info_ptr );
HYPRE_Int hypre_CreateCommInfoFromGrids ( hypre_StructGrid *from_grid, hypre_StructGrid *to_grid,
hypre_CommInfo **comm_info_ptr );
/* computation.c */
HYPRE_Int hypre_ComputeInfoCreate ( hypre_CommInfo *comm_info, hypre_BoxArrayArray *indt_boxes,
hypre_BoxArrayArray *dept_boxes, hypre_ComputeInfo **compute_info_ptr );
HYPRE_Int hypre_ComputeInfoProjectSend ( hypre_ComputeInfo *compute_info, hypre_Index index,
hypre_Index stride );
HYPRE_Int hypre_ComputeInfoProjectRecv ( hypre_ComputeInfo *compute_info, hypre_Index index,
hypre_Index stride );
HYPRE_Int hypre_ComputeInfoProjectComp ( hypre_ComputeInfo *compute_info, hypre_Index index,
hypre_Index stride );
HYPRE_Int hypre_ComputeInfoDestroy ( hypre_ComputeInfo *compute_info );
HYPRE_Int hypre_CreateComputeInfo ( hypre_StructGrid *grid, hypre_StructStencil *stencil,
hypre_ComputeInfo **compute_info_ptr );
HYPRE_Int hypre_ComputePkgCreate ( hypre_ComputeInfo *compute_info, hypre_BoxArray *data_space,
HYPRE_Int num_values, hypre_StructGrid *grid, hypre_ComputePkg **compute_pkg_ptr );
HYPRE_Int hypre_ComputePkgDestroy ( hypre_ComputePkg *compute_pkg );
HYPRE_Int hypre_InitializeIndtComputations ( hypre_ComputePkg *compute_pkg, HYPRE_Complex *data,
hypre_CommHandle **comm_handle_ptr );
HYPRE_Int hypre_FinalizeIndtComputations ( hypre_CommHandle *comm_handle );
/* HYPRE_struct_grid.c */
HYPRE_Int HYPRE_StructGridCreate ( MPI_Comm comm, HYPRE_Int dim, HYPRE_StructGrid *grid );
HYPRE_Int HYPRE_StructGridDestroy ( HYPRE_StructGrid grid );
HYPRE_Int HYPRE_StructGridSetExtents ( HYPRE_StructGrid grid, HYPRE_Int *ilower,
HYPRE_Int *iupper );
HYPRE_Int HYPRE_StructGridSetPeriodic ( HYPRE_StructGrid grid, HYPRE_Int *periodic );
HYPRE_Int HYPRE_StructGridAssemble ( HYPRE_StructGrid grid );
HYPRE_Int HYPRE_StructGridSetNumGhost ( HYPRE_StructGrid grid, HYPRE_Int *num_ghost );
/* HYPRE_struct_matrix.c */
HYPRE_Int HYPRE_StructMatrixCreate ( MPI_Comm comm, HYPRE_StructGrid grid,
HYPRE_StructStencil stencil, HYPRE_StructMatrix *matrix );
HYPRE_Int HYPRE_StructMatrixDestroy ( HYPRE_StructMatrix matrix );
HYPRE_Int HYPRE_StructMatrixInitialize ( HYPRE_StructMatrix matrix );
HYPRE_Int HYPRE_StructMatrixSetValues ( HYPRE_StructMatrix matrix, HYPRE_Int *grid_index,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices, HYPRE_Complex *values );
HYPRE_Int HYPRE_StructMatrixGetValues ( HYPRE_StructMatrix matrix, HYPRE_Int *grid_index,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices, HYPRE_Complex *values );
HYPRE_Int HYPRE_StructMatrixSetBoxValues ( HYPRE_StructMatrix matrix, HYPRE_Int *ilower,
HYPRE_Int *iupper, HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices,
HYPRE_Complex *values );
HYPRE_Int HYPRE_StructMatrixGetBoxValues ( HYPRE_StructMatrix matrix, HYPRE_Int *ilower,
HYPRE_Int *iupper, HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices,
HYPRE_Complex *values );
HYPRE_Int HYPRE_StructMatrixSetConstantValues ( HYPRE_StructMatrix matrix,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices, HYPRE_Complex *values );
HYPRE_Int HYPRE_StructMatrixAddToValues ( HYPRE_StructMatrix matrix, HYPRE_Int *grid_index,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices, HYPRE_Complex *values );
HYPRE_Int HYPRE_StructMatrixAddToBoxValues ( HYPRE_StructMatrix matrix, HYPRE_Int *ilower,
HYPRE_Int *iupper, HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices,
HYPRE_Complex *values );
HYPRE_Int HYPRE_StructMatrixAddToConstantValues ( HYPRE_StructMatrix matrix,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices, HYPRE_Complex *values );
HYPRE_Int HYPRE_StructMatrixAssemble ( HYPRE_StructMatrix matrix );
HYPRE_Int HYPRE_StructMatrixSetNumGhost ( HYPRE_StructMatrix matrix, HYPRE_Int *num_ghost );
HYPRE_Int HYPRE_StructMatrixGetGrid ( HYPRE_StructMatrix matrix, HYPRE_StructGrid *grid );
HYPRE_Int HYPRE_StructMatrixSetSymmetric ( HYPRE_StructMatrix matrix, HYPRE_Int symmetric );
HYPRE_Int HYPRE_StructMatrixSetConstantEntries ( HYPRE_StructMatrix matrix, HYPRE_Int nentries,
HYPRE_Int *entries );
HYPRE_Int HYPRE_StructMatrixPrint ( const char *filename, HYPRE_StructMatrix matrix,
HYPRE_Int all );
HYPRE_Int HYPRE_StructMatrixMatvec ( HYPRE_Complex alpha, HYPRE_StructMatrix A,
HYPRE_StructVector x, HYPRE_Complex beta, HYPRE_StructVector y );
HYPRE_Int HYPRE_StructMatrixClearBoundary( HYPRE_StructMatrix matrix );
/* HYPRE_struct_stencil.c */
HYPRE_Int HYPRE_StructStencilCreate ( HYPRE_Int dim, HYPRE_Int size, HYPRE_StructStencil *stencil );
HYPRE_Int HYPRE_StructStencilSetElement ( HYPRE_StructStencil stencil, HYPRE_Int element_index,
HYPRE_Int *offset );
HYPRE_Int HYPRE_StructStencilDestroy ( HYPRE_StructStencil stencil );
/* HYPRE_struct_vector.c */
HYPRE_Int HYPRE_StructVectorCreate ( MPI_Comm comm, HYPRE_StructGrid grid,
HYPRE_StructVector *vector );
HYPRE_Int HYPRE_StructVectorDestroy ( HYPRE_StructVector struct_vector );
HYPRE_Int HYPRE_StructVectorInitialize ( HYPRE_StructVector vector );
HYPRE_Int HYPRE_StructVectorSetValues ( HYPRE_StructVector vector, HYPRE_Int *grid_index,
HYPRE_Complex values );
HYPRE_Int HYPRE_StructVectorSetBoxValues ( HYPRE_StructVector vector, HYPRE_Int *ilower,
HYPRE_Int *iupper, HYPRE_Complex *values );
HYPRE_Int HYPRE_StructVectorAddToValues ( HYPRE_StructVector vector, HYPRE_Int *grid_index,
HYPRE_Complex values );
HYPRE_Int HYPRE_StructVectorAddToBoxValues ( HYPRE_StructVector vector, HYPRE_Int *ilower,
HYPRE_Int *iupper, HYPRE_Complex *values );
HYPRE_Int HYPRE_StructVectorScaleValues ( HYPRE_StructVector vector, HYPRE_Complex factor );
HYPRE_Int HYPRE_StructVectorGetValues ( HYPRE_StructVector vector, HYPRE_Int *grid_index,
HYPRE_Complex *values );
HYPRE_Int HYPRE_StructVectorGetBoxValues ( HYPRE_StructVector vector, HYPRE_Int *ilower,
HYPRE_Int *iupper, HYPRE_Complex *values );
HYPRE_Int HYPRE_StructVectorAssemble ( HYPRE_StructVector vector );
HYPRE_Int hypre_StructVectorPrintData ( FILE *file, hypre_StructVector *vector, HYPRE_Int all );
HYPRE_Int hypre_StructVectorReadData ( FILE *file, hypre_StructVector *vector );
HYPRE_Int HYPRE_StructVectorPrint ( const char *filename, HYPRE_StructVector vector,
HYPRE_Int all );
HYPRE_Int HYPRE_StructVectorRead ( MPI_Comm comm, const char *filename,
HYPRE_Int *num_ghost, HYPRE_StructVector *vector );
HYPRE_Int HYPRE_StructVectorSetNumGhost ( HYPRE_StructVector vector, HYPRE_Int *num_ghost );
HYPRE_Int HYPRE_StructVectorCopy ( HYPRE_StructVector x, HYPRE_StructVector y );
HYPRE_Int HYPRE_StructVectorSetConstantValues ( HYPRE_StructVector vector, HYPRE_Complex values );
HYPRE_Int HYPRE_StructVectorGetMigrateCommPkg ( HYPRE_StructVector from_vector,
HYPRE_StructVector to_vector, HYPRE_CommPkg *comm_pkg );
HYPRE_Int HYPRE_StructVectorMigrate ( HYPRE_CommPkg comm_pkg, HYPRE_StructVector from_vector,
HYPRE_StructVector to_vector );
HYPRE_Int HYPRE_CommPkgDestroy ( HYPRE_CommPkg comm_pkg );
/* project.c */
HYPRE_Int hypre_ProjectBox ( hypre_Box *box, hypre_Index index, hypre_Index stride );
HYPRE_Int hypre_ProjectBoxArray ( hypre_BoxArray *box_array, hypre_Index index,
hypre_Index stride );
HYPRE_Int hypre_ProjectBoxArrayArray ( hypre_BoxArrayArray *box_array_array, hypre_Index index,
hypre_Index stride );
/* struct_axpy.c */
HYPRE_Int hypre_StructAxpy ( HYPRE_Complex alpha, hypre_StructVector *x, hypre_StructVector *y );
/* struct_communication.c */
HYPRE_Int hypre_CommPkgCreate ( hypre_CommInfo *comm_info, hypre_BoxArray *send_data_space,
hypre_BoxArray *recv_data_space, HYPRE_Int num_values, HYPRE_Int **orders, HYPRE_Int reverse,
MPI_Comm comm, hypre_CommPkg **comm_pkg_ptr );
HYPRE_Int hypre_CommTypeSetEntries ( hypre_CommType *comm_type, HYPRE_Int *boxnums,
hypre_Box *boxes, hypre_Index stride, hypre_Index coord, hypre_Index dir, HYPRE_Int *order,
hypre_BoxArray *data_space, HYPRE_Int *data_offsets );
HYPRE_Int hypre_CommTypeSetEntry ( hypre_Box *box, hypre_Index stride, hypre_Index coord,
hypre_Index dir, HYPRE_Int *order, hypre_Box *data_box, HYPRE_Int data_box_offset,
hypre_CommEntryType *comm_entry );
HYPRE_Int hypre_InitializeCommunication ( hypre_CommPkg *comm_pkg, HYPRE_Complex *send_data,
HYPRE_Complex *recv_data, HYPRE_Int action, HYPRE_Int tag, hypre_CommHandle **comm_handle_ptr );
HYPRE_Int hypre_FinalizeCommunication ( hypre_CommHandle *comm_handle );
HYPRE_Int hypre_ExchangeLocalData ( hypre_CommPkg *comm_pkg, HYPRE_Complex *send_data,
HYPRE_Complex *recv_data, HYPRE_Int action );
HYPRE_Int hypre_CommPkgDestroy ( hypre_CommPkg *comm_pkg );
/* struct_copy.c */
HYPRE_Int hypre_StructCopy ( hypre_StructVector *x, hypre_StructVector *y );
HYPRE_Int hypre_StructPartialCopy ( hypre_StructVector *x, hypre_StructVector *y,
hypre_BoxArrayArray *array_boxes );
/* struct_grid.c */
HYPRE_Int hypre_StructGridCreate ( MPI_Comm comm, HYPRE_Int dim, hypre_StructGrid **grid_ptr );
HYPRE_Int hypre_StructGridRef ( hypre_StructGrid *grid, hypre_StructGrid **grid_ref );
HYPRE_Int hypre_StructGridDestroy ( hypre_StructGrid *grid );
HYPRE_Int hypre_StructGridSetPeriodic ( hypre_StructGrid *grid, hypre_Index periodic );
HYPRE_Int hypre_StructGridSetExtents ( hypre_StructGrid *grid, hypre_Index ilower,
hypre_Index iupper );
HYPRE_Int hypre_StructGridSetBoxes ( hypre_StructGrid *grid, hypre_BoxArray *boxes );
HYPRE_Int hypre_StructGridSetBoundingBox ( hypre_StructGrid *grid, hypre_Box *new_bb );
HYPRE_Int hypre_StructGridSetIDs ( hypre_StructGrid *grid, HYPRE_Int *ids );
HYPRE_Int hypre_StructGridSetBoxManager ( hypre_StructGrid *grid, hypre_BoxManager *boxman );
HYPRE_Int hypre_StructGridSetMaxDistance ( hypre_StructGrid *grid, hypre_Index dist );
HYPRE_Int hypre_StructGridAssemble ( hypre_StructGrid *grid );
HYPRE_Int hypre_GatherAllBoxes ( MPI_Comm comm, hypre_BoxArray *boxes, HYPRE_Int dim,
hypre_BoxArray **all_boxes_ptr, HYPRE_Int **all_procs_ptr, HYPRE_Int *first_local_ptr );
HYPRE_Int hypre_ComputeBoxnums ( hypre_BoxArray *boxes, HYPRE_Int *procs, HYPRE_Int **boxnums_ptr );
HYPRE_Int hypre_StructGridPrint ( FILE *file, hypre_StructGrid *grid );
HYPRE_Int hypre_StructGridRead ( MPI_Comm comm, FILE *file, hypre_StructGrid **grid_ptr );
HYPRE_Int hypre_StructGridSetNumGhost ( hypre_StructGrid *grid, HYPRE_Int *num_ghost );
HYPRE_Int hypre_StructGridGetMaxBoxSize ( hypre_StructGrid *grid );
#if defined(HYPRE_USING_CUDA) || defined(HYPRE_USING_HIP)
HYPRE_Int hypre_StructGridSetDataLocation( HYPRE_StructGrid grid,
HYPRE_MemoryLocation data_location );
#endif
/* struct_innerprod.c */
HYPRE_Real hypre_StructInnerProd ( hypre_StructVector *x, hypre_StructVector *y );
/* struct_io.c */
HYPRE_Int hypre_PrintBoxArrayData ( FILE *file, hypre_BoxArray *box_array,
hypre_BoxArray *data_space, HYPRE_Int num_values, HYPRE_Int dim, HYPRE_Complex *data );
HYPRE_Int hypre_PrintCCVDBoxArrayData ( FILE *file, hypre_BoxArray *box_array,
hypre_BoxArray *data_space, HYPRE_Int num_values, HYPRE_Int center_rank, HYPRE_Int stencil_size,
HYPRE_Int *symm_elements, HYPRE_Int dim, HYPRE_Complex *data );
HYPRE_Int hypre_PrintCCBoxArrayData ( FILE *file, hypre_BoxArray *box_array,
hypre_BoxArray *data_space, HYPRE_Int num_values, HYPRE_Complex *data );
HYPRE_Int hypre_ReadBoxArrayData ( FILE *file, hypre_BoxArray *box_array,
hypre_BoxArray *data_space, HYPRE_Int num_values, HYPRE_Int dim, HYPRE_Complex *data );
HYPRE_Int hypre_ReadBoxArrayData_CC ( FILE *file, hypre_BoxArray *box_array,
hypre_BoxArray *data_space, HYPRE_Int stencil_size, HYPRE_Int real_stencil_size,
HYPRE_Int constant_coefficient, HYPRE_Int dim, HYPRE_Complex *data );
/* struct_matrix.c */
HYPRE_Complex *hypre_StructMatrixExtractPointerByIndex ( hypre_StructMatrix *matrix, HYPRE_Int b,
hypre_Index index );
hypre_StructMatrix *hypre_StructMatrixCreate ( MPI_Comm comm, hypre_StructGrid *grid,
hypre_StructStencil *user_stencil );
hypre_StructMatrix *hypre_StructMatrixRef ( hypre_StructMatrix *matrix );
HYPRE_Int hypre_StructMatrixDestroy ( hypre_StructMatrix *matrix );
HYPRE_Int hypre_StructMatrixInitializeShell ( hypre_StructMatrix *matrix );
HYPRE_Int hypre_StructMatrixInitializeData ( hypre_StructMatrix *matrix, HYPRE_Complex *data,
HYPRE_Complex *data_const);
HYPRE_Int hypre_StructMatrixInitialize ( hypre_StructMatrix *matrix );
HYPRE_Int hypre_StructMatrixSetValues ( hypre_StructMatrix *matrix, hypre_Index grid_index,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices, HYPRE_Complex *values, HYPRE_Int action,
HYPRE_Int boxnum, HYPRE_Int outside );
HYPRE_Int hypre_StructMatrixSetBoxValues ( hypre_StructMatrix *matrix, hypre_Box *set_box,
hypre_Box *value_box, HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices,
HYPRE_Complex *values, HYPRE_Int action, HYPRE_Int boxnum, HYPRE_Int outside );
HYPRE_Int hypre_StructMatrixSetConstantValues ( hypre_StructMatrix *matrix,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices, HYPRE_Complex *values,
HYPRE_Int action );
HYPRE_Int hypre_StructMatrixClearValues ( hypre_StructMatrix *matrix, hypre_Index grid_index,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices, HYPRE_Int boxnum, HYPRE_Int outside );
HYPRE_Int hypre_StructMatrixClearBoxValues ( hypre_StructMatrix *matrix, hypre_Box *clear_box,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices, HYPRE_Int boxnum, HYPRE_Int outside );
HYPRE_Int hypre_StructMatrixAssemble ( hypre_StructMatrix *matrix );
HYPRE_Int hypre_StructMatrixSetNumGhost ( hypre_StructMatrix *matrix, HYPRE_Int *num_ghost );
HYPRE_Int hypre_StructMatrixSetConstantCoefficient ( hypre_StructMatrix *matrix,
HYPRE_Int constant_coefficient );
HYPRE_Int hypre_StructMatrixSetConstantEntries ( hypre_StructMatrix *matrix, HYPRE_Int nentries,
HYPRE_Int *entries );
HYPRE_Int hypre_StructMatrixClearGhostValues ( hypre_StructMatrix *matrix );
HYPRE_Int hypre_StructMatrixPrintData ( FILE *file, hypre_StructMatrix *matrix, HYPRE_Int all );
HYPRE_Int hypre_StructMatrixReadData ( FILE *file, hypre_StructMatrix *matrix );
HYPRE_Int hypre_StructMatrixPrint ( const char *filename, hypre_StructMatrix *matrix,
HYPRE_Int all );
hypre_StructMatrix *hypre_StructMatrixRead ( MPI_Comm comm, const char *filename,
HYPRE_Int *num_ghost );
HYPRE_Int hypre_StructMatrixMigrate ( hypre_StructMatrix *from_matrix,
hypre_StructMatrix *to_matrix );
HYPRE_Int hypre_StructMatrixClearBoundary( hypre_StructMatrix *matrix);
/* struct_matrix_mask.c */
hypre_StructMatrix *hypre_StructMatrixCreateMask ( hypre_StructMatrix *matrix,
HYPRE_Int num_stencil_indices, HYPRE_Int *stencil_indices );
/* struct_matvec.c */
void *hypre_StructMatvecCreate ( void );
HYPRE_Int hypre_StructMatvecSetup ( void *matvec_vdata, hypre_StructMatrix *A,
hypre_StructVector *x );
HYPRE_Int hypre_StructMatvecCompute ( void *matvec_vdata, HYPRE_Complex alpha,
hypre_StructMatrix *A, hypre_StructVector *x, HYPRE_Complex beta, hypre_StructVector *y );
HYPRE_Int hypre_StructMatvecCC0 ( HYPRE_Complex alpha, hypre_StructMatrix *A, hypre_StructVector *x,
hypre_StructVector *y, hypre_BoxArrayArray *compute_box_aa, hypre_IndexRef stride );
HYPRE_Int hypre_StructMatvecCC1 ( HYPRE_Complex alpha, hypre_StructMatrix *A, hypre_StructVector *x,
hypre_StructVector *y, hypre_BoxArrayArray *compute_box_aa, hypre_IndexRef stride );
HYPRE_Int hypre_StructMatvecCC2 ( HYPRE_Complex alpha, hypre_StructMatrix *A, hypre_StructVector *x,
hypre_StructVector *y, hypre_BoxArrayArray *compute_box_aa, hypre_IndexRef stride );
HYPRE_Int hypre_StructMatvecDestroy ( void *matvec_vdata );
HYPRE_Int hypre_StructMatvec ( HYPRE_Complex alpha, hypre_StructMatrix *A, hypre_StructVector *x,
HYPRE_Complex beta, hypre_StructVector *y );
/* struct_scale.c */
HYPRE_Int hypre_StructScale ( HYPRE_Complex alpha, hypre_StructVector *y );
/* struct_stencil.c */
hypre_StructStencil *hypre_StructStencilCreate ( HYPRE_Int dim, HYPRE_Int size,
hypre_Index *shape );
hypre_StructStencil *hypre_StructStencilRef ( hypre_StructStencil *stencil );
HYPRE_Int hypre_StructStencilDestroy ( hypre_StructStencil *stencil );
HYPRE_Int hypre_StructStencilElementRank ( hypre_StructStencil *stencil,
hypre_Index stencil_element );
HYPRE_Int hypre_StructStencilSymmetrize ( hypre_StructStencil *stencil,
hypre_StructStencil **symm_stencil_ptr, HYPRE_Int **symm_elements_ptr );
/* struct_vector.c */
hypre_StructVector *hypre_StructVectorCreate ( MPI_Comm comm, hypre_StructGrid *grid );
hypre_StructVector *hypre_StructVectorRef ( hypre_StructVector *vector );
HYPRE_Int hypre_StructVectorDestroy ( hypre_StructVector *vector );
HYPRE_Int hypre_StructVectorInitializeShell ( hypre_StructVector *vector );
HYPRE_Int hypre_StructVectorInitializeData ( hypre_StructVector *vector, HYPRE_Complex *data);
HYPRE_Int hypre_StructVectorInitialize ( hypre_StructVector *vector );
HYPRE_Int hypre_StructVectorSetValues ( hypre_StructVector *vector, hypre_Index grid_index,
HYPRE_Complex *values, HYPRE_Int action, HYPRE_Int boxnum, HYPRE_Int outside );
HYPRE_Int hypre_StructVectorSetBoxValues ( hypre_StructVector *vector, hypre_Box *set_box,
hypre_Box *value_box, HYPRE_Complex *values, HYPRE_Int action, HYPRE_Int boxnum,
HYPRE_Int outside );
HYPRE_Int hypre_StructVectorClearValues ( hypre_StructVector *vector, hypre_Index grid_index,
HYPRE_Int boxnum, HYPRE_Int outside );
HYPRE_Int hypre_StructVectorClearBoxValues ( hypre_StructVector *vector, hypre_Box *clear_box,
HYPRE_Int boxnum, HYPRE_Int outside );
HYPRE_Int hypre_StructVectorClearAllValues ( hypre_StructVector *vector );
HYPRE_Int hypre_StructVectorSetNumGhost ( hypre_StructVector *vector, HYPRE_Int *num_ghost );
HYPRE_Int hypre_StructVectorSetDataSize(hypre_StructVector *vector, HYPRE_Int *data_size,
HYPRE_Int *data_host_size);
HYPRE_Int hypre_StructVectorAssemble ( hypre_StructVector *vector );
HYPRE_Int hypre_StructVectorCopy ( hypre_StructVector *x, hypre_StructVector *y );
HYPRE_Int hypre_StructVectorSetConstantValues ( hypre_StructVector *vector, HYPRE_Complex values );
HYPRE_Int hypre_StructVectorSetFunctionValues ( hypre_StructVector *vector,
HYPRE_Complex (*fcn )( HYPRE_Int, HYPRE_Int, HYPRE_Int ));
HYPRE_Int hypre_StructVectorClearGhostValues ( hypre_StructVector *vector );
HYPRE_Int hypre_StructVectorClearBoundGhostValues ( hypre_StructVector *vector, HYPRE_Int force );
HYPRE_Int hypre_StructVectorScaleValues ( hypre_StructVector *vector, HYPRE_Complex factor );
hypre_CommPkg *hypre_StructVectorGetMigrateCommPkg ( hypre_StructVector *from_vector,
hypre_StructVector *to_vector );
HYPRE_Int hypre_StructVectorMigrate ( hypre_CommPkg *comm_pkg, hypre_StructVector *from_vector,
hypre_StructVector *to_vector );
HYPRE_Int hypre_StructVectorPrint ( const char *filename, hypre_StructVector *vector,
HYPRE_Int all );
hypre_StructVector *hypre_StructVectorRead ( MPI_Comm comm, const char *filename,
HYPRE_Int *num_ghost );
hypre_StructVector *hypre_StructVectorClone ( hypre_StructVector *vector );
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/******************************************************************************
*
* Header info for the BoxLoop
*
*****************************************************************************/
/*--------------------------------------------------------------------------
* BoxLoop macros:
*--------------------------------------------------------------------------*/
#ifndef HYPRE_BOXLOOP_DEVICEOMP_HEADER
#define HYPRE_BOXLOOP_DEVICEOMP_HEADER
#if defined(HYPRE_USING_DEVICE_OPENMP) && !defined(HYPRE_USING_RAJA) && !defined(HYPRE_USING_KOKKOS)
#include "omp.h"
/* concatenation:
*/
#define HYPRE_CONCAT2(x, y) x ## _ ## y
#define HYPRE_XCONCAT2(x, y) HYPRE_CONCAT2(x, y)
#define HYPRE_CONCAT3(x, y, z) x ## _ ## y ## _ ## z
#define HYPRE_XCONCAT3(x, y, z) HYPRE_CONCAT3(x, y, z)
/* if use OMP 4.5 default team size and number of teams */
#define AUTO_OMP_TEAM
#ifndef AUTO_OMP_TEAM
/* omp team size (aka. gpu block size) */
#define hypre_gpu_block_size HYPRE_1D_BLOCK_SIZE
/* the max number of omp teams */
#define hypre_max_num_blocks 1000000
#endif
//#define HYPRE_BOXLOOP_ENTRY_PRINT hypre_printf("%s %s %d\n", __FILE__, __func__, __LINE__);
#define HYPRE_BOXLOOP_ENTRY_PRINT
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
BOX LOOPS [TEAM DISTRIBUTE VERSION]
!!! NOTE: THIS CODE ONLY WORKS FOR DIM <= 3 !!!
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
/*
#define hypre_BoxLoop0For()
#define hypre_BoxLoop1For(i1)
#define hypre_BoxLoop2For(i1, i2)
#define hypre_BoxLoop3For(i1, i2, i3)
#define hypre_BoxLoop4For(i1, i2, i3, i4)
*/
#define hypre_BoxLoopBlock() 0
#define hypre_BoxLoop0Begin zypre_omp4_dist_BoxLoop0Begin
#define hypre_BoxLoop0End zypre_omp4_dist_BoxLoopEnd
#define hypre_BoxLoop1Begin zypre_omp4_dist_BoxLoop1Begin
#define hypre_BoxLoop1End zypre_omp4_dist_BoxLoopEnd
#define hypre_BasicBoxLoop2Begin zypre_omp4_dist_BoxLoop2_v2_Begin
#define hypre_BoxLoop2Begin zypre_omp4_dist_BoxLoop2Begin
#define hypre_BoxLoop2End zypre_omp4_dist_BoxLoopEnd
#define hypre_BoxLoop3Begin zypre_omp4_dist_BoxLoop3Begin
#if 0
#define hypre_BoxLoop3_SAME_STRIDE_Begin zypre_omp4_dist_BoxLoop3_SAME_STRIDE_Begin
#endif
#define hypre_BoxLoop3End zypre_omp4_dist_BoxLoopEnd
#define hypre_BoxLoop4Begin zypre_omp4_dist_BoxLoop4Begin
#define hypre_BoxLoop4End zypre_omp4_dist_BoxLoopEnd
#define hypre_LoopBegin zypre_omp4_dist_LoopBegin
#define hypre_LoopEnd zypre_omp4_dist_BoxLoopEnd
/* Look for more in struct_ls/red_black_gs.h" */
#define zypre_omp4_dist_BoxLoopEnd(...) \
}\
/*cudaDeviceSynchronize();*/ \
}
#define HYPRE_BOX_REDUCTION
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* host code: declare variables used in the box loop
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_BoxLoopDeclareInit_0(ndim, loop_size) \
HYPRE_Int hypre__ndim = ndim, hypre__tot = 1; \
/* HYPRE_Int hypre__thread; */ \
/* loop size */ \
HYPRE_Int hypre__loop_size_0, hypre__loop_size_1, hypre__loop_size_2; \
if (hypre__ndim > 0) { hypre__loop_size_0 = loop_size[0]; hypre__tot *= hypre__loop_size_0; } \
if (hypre__ndim > 1) { hypre__loop_size_1 = loop_size[1]; hypre__tot *= hypre__loop_size_1; } \
if (hypre__ndim > 2) { hypre__loop_size_2 = loop_size[2]; hypre__tot *= hypre__loop_size_2; }
#ifdef AUTO_OMP_TEAM
#define TEAM_CLAUSE
#define zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) zypre_omp4_BoxLoopDeclareInit_0(ndim, loop_size)
#else
#define TEAM_CLAUSE num_teams(num_blocks) thread_limit(block_size)
#define zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) zypre_omp4_BoxLoopDeclareInit_0(ndim, loop_size) \
/* GPU block numbers and dimensions */ \
HYPRE_Int block_size = hypre_gpu_block_size; \
HYPRE_Int num_blocks = hypre_min(hypre_max_num_blocks, (hypre__tot + hypre_gpu_block_size - 1) / hypre_gpu_block_size);
#endif
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* host code: declare and initialize variables for box k
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_BoxKDeclareInitBody(j, k, startk, dboxk, stridek) \
HYPRE_XCONCAT3(hypre__stride,j,k) = stridek[j]; \
/* precompute some entities used in the parallel for loop */ \
HYPRE_XCONCAT3(hypre__box_start_imin,j,k) = startk[j] - dboxk->imin[j]; \
HYPRE_XCONCAT3(hypre__box_imax_imin,j,k) = dboxk->imax[j] - dboxk->imin[j] + 1;
#define zypre_omp4_BoxKDeclareInit(k, startk, dboxk, stridek)\
/* start - imin */ \
HYPRE_Int HYPRE_XCONCAT3(hypre__box_start_imin,0,k), HYPRE_XCONCAT3(hypre__box_start_imin,1,k), HYPRE_XCONCAT3(hypre__box_start_imin,2,k); \
/* imax - imin + 1 */ \
HYPRE_Int HYPRE_XCONCAT3(hypre__box_imax_imin,0,k), HYPRE_XCONCAT3(hypre__box_imax_imin,1,k), HYPRE_XCONCAT3(hypre__box_imax_imin,2,k); \
/* stride */ \
HYPRE_Int HYPRE_XCONCAT3(hypre__stride,0,k), HYPRE_XCONCAT3(hypre__stride,1,k), HYPRE_XCONCAT3(hypre__stride,2,k); \
/*if (hypre__ndim > 0)*/ { zypre_omp4_BoxKDeclareInitBody(0, k, startk, dboxk, stridek) } \
if (hypre__ndim > 1) { zypre_omp4_BoxKDeclareInitBody(1, k, startk, dboxk, stridek) } \
if (hypre__ndim > 2) { zypre_omp4_BoxKDeclareInitBody(2, k, startk, dboxk, stridek) } \
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* map clause
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define MAP_CLAUSE0
#define MAP_CLAUSE1
#define MAP_CLAUSE2
#define MAP_CLAUSE3
#define MAP_CLAUSE4
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* if clause
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define IF_CLAUSE if (hypre__global_offload && hypre__tot > 0)
//#define IF_CLAUSE
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* is_device_ptr clause
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#if defined(HYPRE_DEVICE_OPENMP_ALLOC)
#define IS_DEVICE_CLAUSE DEVICE_VAR
#else
#define IS_DEVICE_CLAUSE
#endif
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* device code for BoxLoop 1, set i1
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_BoxLoopSet1Body(j, i1) \
/* coord in dimension j */ \
hypre__i = hypre__J % HYPRE_XCONCAT2(hypre__loop_size,j); \
/* once */ \
hypre__i_1 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,1) + HYPRE_XCONCAT3(hypre__box_start_imin,j,1);\
/* once */ \
i1 += hypre__i_1 * hypre__I_1; \
/* once */ \
hypre__I_1 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,1); \
/* */ \
hypre__J /= HYPRE_XCONCAT2(hypre__loop_size,j); \
/* save the 3-D id */ \
HYPRE_XCONCAT2(hypre__id,j) = hypre__i;
#define zypre_omp4_BoxLoopSet1(i1) \
HYPRE_Int hypre__I_1, hypre__i, hypre__i_1, hypre__J, i1, idx; \
HYPRE_Int hypre__id_0, hypre__id_1, hypre__id_2; \
hypre__I_1 = 1; idx = hypre__J = hypre__thread; i1 = 0; \
/*if (hypre__ndim > 0)*/ { zypre_omp4_BoxLoopSet1Body(0, i1) } \
if (hypre__ndim > 1) { zypre_omp4_BoxLoopSet1Body(1, i1) } \
if (hypre__ndim > 2) { zypre_omp4_BoxLoopSet1Body(2, i1) }
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* device code for BoxLoop 2, set i1, i2
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_BoxLoopSet2Body(j, i1, i2) \
/* */ \
hypre__i = hypre__J % HYPRE_XCONCAT2(hypre__loop_size,j); \
/* twice */ \
hypre__i_1 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,1) + HYPRE_XCONCAT3(hypre__box_start_imin,j,1);\
hypre__i_2 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,2) + HYPRE_XCONCAT3(hypre__box_start_imin,j,2);\
/* twice */ \
i1 += hypre__i_1 * hypre__I_1; \
i2 += hypre__i_2 * hypre__I_2; \
/* twice */ \
hypre__I_1 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,1); \
hypre__I_2 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,2); \
/* */ \
hypre__J /= HYPRE_XCONCAT2(hypre__loop_size,j); \
/* save the 3-D id */ \
HYPRE_XCONCAT2(hypre__id,j) = hypre__i;
#define zypre_omp4_BoxLoopSet2(i1, i2) \
HYPRE_Int hypre__I_1, hypre__I_2, hypre__i, hypre__i_1, hypre__i_2, hypre__J, i1, i2; \
HYPRE_Int hypre__id_0, hypre__id_1, hypre__id_2; \
hypre__I_1 = hypre__I_2 = 1; hypre__J = hypre__thread; i1 = i2 = 0; \
/*if (hypre__ndim > 0)*/ { zypre_omp4_BoxLoopSet2Body(0, i1, i2) } \
if (hypre__ndim > 1) { zypre_omp4_BoxLoopSet2Body(1, i1, i2) } \
if (hypre__ndim > 2) { zypre_omp4_BoxLoopSet2Body(2, i1, i2) }
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* device code for BoxLoop 3, set i1, i2, i3
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_BoxLoopSet3Body(j, i1, i2, i3) \
/* */ \
hypre__i = hypre__J % HYPRE_XCONCAT2(hypre__loop_size,j); \
/* 3 times */ \
hypre__i_1 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,1) + HYPRE_XCONCAT3(hypre__box_start_imin,j,1);\
hypre__i_2 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,2) + HYPRE_XCONCAT3(hypre__box_start_imin,j,2);\
hypre__i_3 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,3) + HYPRE_XCONCAT3(hypre__box_start_imin,j,3);\
/* 3 times */ \
i1 += hypre__i_1 * hypre__I_1; \
i2 += hypre__i_2 * hypre__I_2; \
i3 += hypre__i_3 * hypre__I_3; \
/* 3 times */ \
hypre__I_1 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,1); \
hypre__I_2 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,2); \
hypre__I_3 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,3); \
/* */ \
hypre__J /= HYPRE_XCONCAT2(hypre__loop_size,j); \
/* save the 3-D id */ \
HYPRE_XCONCAT2(hypre__id,j) = hypre__i;
#define zypre_omp4_BoxLoopSet3(i1, i2, i3) \
HYPRE_Int hypre__I_1, hypre__I_2, hypre__I_3, hypre__i, hypre__i_1, hypre__i_2, hypre__i_3, hypre__J, i1, i2, i3; \
HYPRE_Int hypre__id_0, hypre__id_1, hypre__id_2; \
hypre__I_1 = hypre__I_2 = hypre__I_3 = 1; hypre__J = hypre__thread; i1 = i2 = i3 = 0; \
/*if (hypre__ndim > 0)*/ { zypre_omp4_BoxLoopSet3Body(0, i1, i2, i3) } \
if (hypre__ndim > 1) { zypre_omp4_BoxLoopSet3Body(1, i1, i2, i3) } \
if (hypre__ndim > 2) { zypre_omp4_BoxLoopSet3Body(2, i1, i2, i3) }
#if 0
/* - - - - - special Box 3: XXX */
#define zypre_omp4_BoxLoopSet3_SAME_STRIDE_Body(j, i1, i2, i3) \
/* */ \
hypre__i = (hypre__J % HYPRE_XCONCAT2(hypre__loop_size,j)) * HYPRE_XCONCAT3(hypre__stride,j,1); \
/* 3 times */ \
hypre__i_1 = hypre__i + HYPRE_XCONCAT3(hypre__box_start_imin,j,1);\
hypre__i_2 = hypre__i + HYPRE_XCONCAT3(hypre__box_start_imin,j,2);\
hypre__i_3 = hypre__i + HYPRE_XCONCAT3(hypre__box_start_imin,j,3);\
/* 3 times */ \
i1 += hypre__i_1 * hypre__I_1; \
i2 += hypre__i_2 * hypre__I_2; \
i3 += hypre__i_3 * hypre__I_3; \
/* 3 times */ \
hypre__I_1 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,1); \
hypre__I_2 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,2); \
hypre__I_3 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,3); \
/* */ \
hypre__J /= HYPRE_XCONCAT2(hypre__loop_size,j);
#define zypre_omp4_BoxLoopSet3_SAME_STRIDE(i1, i2, o2, i3) \
HYPRE_Int hypre__I_1, hypre__I_2, hypre__I_3, hypre__i, hypre__i_1, hypre__i_2, hypre__i_3, hypre__J; \
hypre__I_1 = hypre__I_2 = hypre__I_3 = 1; hypre__J = hypre__thread; i1 = i3 = 0; i2 = o2;\
/*if (hypre__ndim > 0)*/ { zypre_omp4_BoxLoopSet3_SAME_STRIDE_Body(0, i1, i2, i3) } \
if (hypre__ndim > 1) { zypre_omp4_BoxLoopSet3_SAME_STRIDE_Body(1, i1, i2, i3) } \
if (hypre__ndim > 2) { zypre_omp4_BoxLoopSet3_SAME_STRIDE_Body(2, i1, i2, i3) }
#endif
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* device code for BoxLoop 4, set i1, i2, i3, i4
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_BoxLoopSet4Body(j, i1, i2, i3, i4) \
/* */ \
hypre__i = hypre__J % HYPRE_XCONCAT2(hypre__loop_size,j); \
/* 4 times */ \
hypre__i_1 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,1) + HYPRE_XCONCAT3(hypre__box_start_imin,j,1);\
hypre__i_2 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,2) + HYPRE_XCONCAT3(hypre__box_start_imin,j,2);\
hypre__i_3 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,3) + HYPRE_XCONCAT3(hypre__box_start_imin,j,3);\
hypre__i_4 = hypre__i * HYPRE_XCONCAT3(hypre__stride,j,4) + HYPRE_XCONCAT3(hypre__box_start_imin,j,4);\
/* 4 times */ \
i1 += hypre__i_1 * hypre__I_1; \
i2 += hypre__i_2 * hypre__I_2; \
i3 += hypre__i_3 * hypre__I_3; \
i4 += hypre__i_4 * hypre__I_4; \
/* 4 times */ \
hypre__I_1 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,1); \
hypre__I_2 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,2); \
hypre__I_3 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,3); \
hypre__I_4 *= HYPRE_XCONCAT3(hypre__box_imax_imin,j,4); \
/* */ \
hypre__J /= HYPRE_XCONCAT2(hypre__loop_size,j); \
/* save the 3-D id */ \
HYPRE_XCONCAT2(hypre__id,j) = hypre__i;
#define zypre_omp4_BoxLoopSet4(i1, i2, i3, i4) \
HYPRE_Int hypre__I_1, hypre__I_2, hypre__I_3, hypre__I_4, hypre__i, hypre__i_1, hypre__i_2, hypre__i_3, hypre__i_4, hypre__J, i1, i2, i3, i4; \
HYPRE_Int hypre__id_0, hypre__id_1, hypre__id_2; \
hypre__I_1 = hypre__I_2 = hypre__I_3 = hypre__I_4 = 1; hypre__J = hypre__thread; i1 = i2 = i3 = i4 = 0; \
/*if (hypre__ndim > 0)*/ { zypre_omp4_BoxLoopSet4Body(0, i1, i2, i3, i4) } \
if (hypre__ndim > 1) { zypre_omp4_BoxLoopSet4Body(1, i1, i2, i3, i4) } \
if (hypre__ndim > 2) { zypre_omp4_BoxLoopSet4Body(2, i1, i2, i3, i4) }
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* BoxLoop 0
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_BoxLoop0Begin(ndim, loop_size) \
{\
/* host code: */ \
HYPRE_BOXLOOP_ENTRY_PRINT \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE0 IS_DEVICE_CLAUSE TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{\
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* BoxLoop 1
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_BoxLoop1Begin(ndim, loop_size, dbox1, start1, stride1, i1) \
{\
/* host code: */ \
HYPRE_BOXLOOP_ENTRY_PRINT \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
zypre_omp4_BoxKDeclareInit(1, start1, dbox1, stride1) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE1 IS_DEVICE_CLAUSE HYPRE_BOX_REDUCTION TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{\
zypre_omp4_BoxLoopSet1(i1)
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* BoxLoop 2
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_BoxLoop2Begin(ndim, loop_size, dbox1, start1, stride1, i1, dbox2, start2, stride2, i2) \
{\
/* host code: */ \
HYPRE_BOXLOOP_ENTRY_PRINT \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
zypre_omp4_BoxKDeclareInit(1, start1, dbox1, stride1) \
zypre_omp4_BoxKDeclareInit(2, start2, dbox2, stride2) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE2 IS_DEVICE_CLAUSE HYPRE_BOX_REDUCTION TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{\
zypre_omp4_BoxLoopSet2(i1, i2)
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* BoxLoop 3
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_BoxLoop3Begin(ndim, loop_size, \
dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2, \
dbox3, start3, stride3, i3) \
{\
/* host code: */ \
HYPRE_BOXLOOP_ENTRY_PRINT \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
zypre_omp4_BoxKDeclareInit(1, start1, dbox1, stride1) \
zypre_omp4_BoxKDeclareInit(2, start2, dbox2, stride2) \
zypre_omp4_BoxKDeclareInit(3, start3, dbox3, stride3) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE3 IS_DEVICE_CLAUSE TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{\
zypre_omp4_BoxLoopSet3(i1, i2, i3)
#if 0
#define zypre_omp4_dist_BoxLoop3_SAME_STRIDE_Begin(ndim, loop_size, \
dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2, o2, \
dbox3, start3, stride3, i3) \
{\
/* host code: */ \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
zypre_omp4_BoxKDeclareInit(1, start1, dbox1, stride1) \
zypre_omp4_BoxKDeclareInit(2, start2, dbox2, stride2) \
zypre_omp4_BoxKDeclareInit(3, start3, dbox3, stride3) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE3 IS_DEVICE_CLAUSE TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{\
zypre_omp4_BoxLoopSet3_SAME_STRIDE(i1, i2, o2, i3)
#endif
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* BoxLoop 4
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_BoxLoop4Begin(ndim, loop_size, \
dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2, \
dbox3, start3, stride3, i3, \
dbox4, start4, stride4, i4) \
{\
/* host code: */ \
HYPRE_BOXLOOP_ENTRY_PRINT \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
zypre_omp4_BoxKDeclareInit(1, start1, dbox1, stride1) \
zypre_omp4_BoxKDeclareInit(2, start2, dbox2, stride2) \
zypre_omp4_BoxKDeclareInit(3, start3, dbox3, stride3) \
zypre_omp4_BoxKDeclareInit(4, start4, dbox4, stride4) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE4 IS_DEVICE_CLAUSE TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{\
zypre_omp4_BoxLoopSet4(i1, i2, i3, i4)
#if 0
/* no longer needed, use the above BoxLoop's for reductions */
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* BoxLoop 1 reduction
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_Red_BoxLoop1Begin(ndim, loop_size, dbox1, start1, stride1, i1, xsum) \
{\
/* host code: */ \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
zypre_omp4_BoxKDeclareInit(1, start1, dbox1, stride1) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE1 map(tofrom: xsum) reduction(+:xsum) TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{\
zypre_omp4_BoxLoopSet1(i1)
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* BoxLoop 2 reduction
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_Red_BoxLoop2Begin(ndim, loop_size, dbox1, start1, stride1, i1, dbox2, start2, stride2, i2, xsum) \
{\
/* host code: */ \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
zypre_omp4_BoxKDeclareInit(1, start1, dbox1, stride1) \
zypre_omp4_BoxKDeclareInit(2, start2, dbox2, stride2) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE2 map(tofrom: xsum) reduction(+:xsum) TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{\
zypre_omp4_BoxLoopSet2(i1, i2)
#endif
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* v2
* host code: declare and initialize variables for box k
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_BoxKDeclareInit_v2(k, stridek)\
/* stridek[0,1,2] */ \
HYPRE_Int HYPRE_XCONCAT3(hypre__stride,0,k), HYPRE_XCONCAT3(hypre__stride,1,k), HYPRE_XCONCAT3(hypre__stride,2,k); \
/*if (hypre__ndim > 0)*/ { HYPRE_XCONCAT3(hypre__stride,0,k) = stridek[0]; } \
if (hypre__ndim > 1) { HYPRE_XCONCAT3(hypre__stride,1,k) = stridek[1]; } \
if (hypre__ndim > 2) { HYPRE_XCONCAT3(hypre__stride,2,k) = stridek[2]; } \
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* v2
* device code for BoxLoop 1, set i1
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_BoxLoopSet1Body_v2(j, i1) \
i1 += ( hypre__J % HYPRE_XCONCAT2(hypre__loop_size,j) ) * HYPRE_XCONCAT3(hypre__stride,j,1);\
hypre__J /= HYPRE_XCONCAT2(hypre__loop_size,j);
#define zypre_omp4_BoxLoopSet1_v2(i1, idx) \
HYPRE_Int hypre__J, i1, idx; \
idx = hypre__J = hypre__thread; i1 = 0; \
/*if (hypre__ndim > 0)*/ { zypre_omp4_BoxLoopSet1Body_v2(0, i1) } \
if (hypre__ndim > 1) { zypre_omp4_BoxLoopSet1Body_v2(1, i1) } \
if (hypre__ndim > 2) { zypre_omp4_BoxLoopSet1Body_v2(2, i1) }
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* v2: Basic
* BoxLoop 1
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_BoxLoop1_v2_Begin(ndim, loop_size, stride1, i1, idx) \
{\
/* host code: */ \
HYPRE_BOXLOOP_ENTRY_PRINT \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
zypre_omp4_BoxKDeclareInit_v2(1, stride1) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE1 IS_DEVICE_CLAUSE TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{\
zypre_omp4_BoxLoopSet1_v2(i1, idx)
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* v2
* device code for BoxLoop 2, set i1, i2
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_BoxLoopSet2Body_v2(j, i1, i2) \
hypre__i = hypre__J % HYPRE_XCONCAT2(hypre__loop_size,j); \
/* twice */ \
i1 += hypre__i * HYPRE_XCONCAT3(hypre__stride,j,1); \
i2 += hypre__i * HYPRE_XCONCAT3(hypre__stride,j,2); \
hypre__J /= HYPRE_XCONCAT2(hypre__loop_size,j);
#define zypre_omp4_BoxLoopSet2_v2(i1, i2) \
HYPRE_Int hypre__i, hypre__J, i1, i2; \
hypre__J = hypre__thread; i1 = i2 = 0; \
/*if (hypre__ndim > 0)*/ { zypre_omp4_BoxLoopSet2Body_v2(0, i1, i2) } \
if (hypre__ndim > 1) { zypre_omp4_BoxLoopSet2Body_v2(1, i1, i2) } \
if (hypre__ndim > 2) { zypre_omp4_BoxLoopSet2Body_v2(2, i1, i2) }
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* v2: Basic
* BoxLoop 2
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_BoxLoop2_v2_Begin(ndim, loop_size, stride1, i1, stride2, i2) \
{ \
/* host code: */ \
HYPRE_BOXLOOP_ENTRY_PRINT \
zypre_omp4_BoxLoopDeclareInit(ndim, loop_size) \
zypre_omp4_BoxKDeclareInit_v2(1, stride1) \
zypre_omp4_BoxKDeclareInit_v2(2, stride2) \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE2 IS_DEVICE_CLAUSE TEAM_CLAUSE)) \
for (HYPRE_Int hypre__thread = 0; hypre__thread < hypre__tot; hypre__thread++) \
{ \
zypre_omp4_BoxLoopSet2_v2(i1, i2)
/* - - - - - - - - - - - - - - - - - - - - - - - - - - - -
* Basic Loop
* - - - - - - - - - - - - - - - - - - - - - - - - - - - -*/
#define zypre_omp4_dist_LoopBegin(size, idx) \
{ \
/* host code: */ \
/* HYPRE_Int idx = 0; */\
HYPRE_Int hypre__tot = size; \
HYPRE_BOXLOOP_ENTRY_PRINT \
/* device code: */ \
_Pragma (HYPRE_XSTR(omp target teams distribute parallel for IF_CLAUSE MAP_CLAUSE2 IS_DEVICE_CLAUSE TEAM_CLAUSE)) \
for (HYPRE_Int idx = 0; idx < hypre__tot; idx++) \
{
#define hypre_BoxLoopGetIndex(index) \
index[0] = hypre__id_0; \
index[1] = hypre__id_1; \
index[2] = hypre__id_2;
/* Reduction */
#define hypre_BoxLoop1ReductionBegin(ndim, loop_size, dbox1, start1, stride1, i1, reducesum) \
hypre_BoxLoop1Begin(ndim, loop_size, dbox1, start1, stride1, i1)
#define hypre_BoxLoop1ReductionEnd(i1, reducesum) \
hypre_BoxLoop1End(i1)
#define hypre_BoxLoop2ReductionBegin(ndim, loop_size, dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2, reducesum) \
hypre_BoxLoop2Begin(ndim, loop_size, dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2)
#define hypre_BoxLoop2ReductionEnd(i1, i2, reducesum) \
hypre_BoxLoop2End(i1, i2)
#endif
#endif /* #ifndef HYPRE_BOXLOOP_DEVICEOMP_HEADER */
/******************************************************************************
* Copyright (c) 1998 Lawrence Livermore National Security, LLC and other
* HYPRE Project Developers. See the top-level COPYRIGHT file for details.
*
* SPDX-License-Identifier: (Apache-2.0 OR MIT)
******************************************************************************/
/******************************************************************************
*
* Header info for the BoxLoop
*
*****************************************************************************/
/*--------------------------------------------------------------------------
* BoxLoop macros:
*--------------------------------------------------------------------------*/
#ifndef HYPRE_BOXLOOP_HOST_HEADER
#define HYPRE_BOXLOOP_HOST_HEADER
#if defined(HYPRE_USING_OPENMP)
#define HYPRE_BOX_REDUCTION
#define HYPRE_OMP_CLAUSE
#if defined(WIN32) && defined(_MSC_VER)
#define Pragma(x) __pragma(x)
#else
#define Pragma(x) _Pragma(HYPRE_XSTR(x))
#endif
#define OMP0 Pragma(omp parallel for HYPRE_OMP_CLAUSE HYPRE_BOX_REDUCTION HYPRE_SMP_SCHEDULE)
#define OMP1 Pragma(omp parallel for private(HYPRE_BOX_PRIVATE) HYPRE_OMP_CLAUSE HYPRE_BOX_REDUCTION HYPRE_SMP_SCHEDULE)
#else /* #if defined(HYPRE_USING_OPENMP) */
#define OMP0
#define OMP1
#endif /* #if defined(HYPRE_USING_OPENMP) */
#define zypre_BoxLoop0Begin(ndim, loop_size) \
{ \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopInit(ndim, loop_size); \
OMP1 \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
zypre_BoxLoopSet(); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define zypre_BoxLoop0End() \
} \
zypre_BoxLoopInc1(); \
zypre_BoxLoopInc2(); \
} \
} \
}
#define zypre_BoxLoop1Begin(ndim, loop_size, \
dbox1, start1, stride1, i1) \
{ \
HYPRE_Int i1; \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopDeclareK(1); \
zypre_BoxLoopInit(ndim, loop_size); \
zypre_BoxLoopInitK(1, dbox1, start1, stride1, i1); \
OMP1 \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
HYPRE_Int i1; \
zypre_BoxLoopSet(); \
zypre_BoxLoopSetK(1, i1); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define zypre_BoxLoop1End(i1) \
i1 += hypre__i0inc1; \
} \
zypre_BoxLoopInc1(); \
i1 += hypre__ikinc1[hypre__d]; \
zypre_BoxLoopInc2(); \
} \
} \
}
#define zypre_BoxLoop2Begin(ndim, loop_size, \
dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2) \
{ \
HYPRE_Int i1, i2; \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopDeclareK(1); \
zypre_BoxLoopDeclareK(2); \
zypre_BoxLoopInit(ndim, loop_size); \
zypre_BoxLoopInitK(1, dbox1, start1, stride1, i1); \
zypre_BoxLoopInitK(2, dbox2, start2, stride2, i2); \
OMP1 \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
HYPRE_Int i1, i2; \
zypre_BoxLoopSet(); \
zypre_BoxLoopSetK(1, i1); \
zypre_BoxLoopSetK(2, i2); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define zypre_BoxLoop2End(i1, i2) \
i1 += hypre__i0inc1; \
i2 += hypre__i0inc2; \
} \
zypre_BoxLoopInc1(); \
i1 += hypre__ikinc1[hypre__d]; \
i2 += hypre__ikinc2[hypre__d]; \
zypre_BoxLoopInc2(); \
} \
} \
}
#define zypre_BoxLoop3Begin(ndim, loop_size, \
dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2, \
dbox3, start3, stride3, i3) \
{ \
HYPRE_Int i1, i2, i3; \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopDeclareK(1); \
zypre_BoxLoopDeclareK(2); \
zypre_BoxLoopDeclareK(3); \
zypre_BoxLoopInit(ndim, loop_size); \
zypre_BoxLoopInitK(1, dbox1, start1, stride1, i1); \
zypre_BoxLoopInitK(2, dbox2, start2, stride2, i2); \
zypre_BoxLoopInitK(3, dbox3, start3, stride3, i3); \
OMP1 \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
HYPRE_Int i1, i2, i3; \
zypre_BoxLoopSet(); \
zypre_BoxLoopSetK(1, i1); \
zypre_BoxLoopSetK(2, i2); \
zypre_BoxLoopSetK(3, i3); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define zypre_BoxLoop3End(i1, i2, i3) \
i1 += hypre__i0inc1; \
i2 += hypre__i0inc2; \
i3 += hypre__i0inc3; \
} \
zypre_BoxLoopInc1(); \
i1 += hypre__ikinc1[hypre__d]; \
i2 += hypre__ikinc2[hypre__d]; \
i3 += hypre__ikinc3[hypre__d]; \
zypre_BoxLoopInc2(); \
} \
} \
}
#define zypre_BoxLoop4Begin(ndim, loop_size, \
dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2, \
dbox3, start3, stride3, i3, \
dbox4, start4, stride4, i4) \
{ \
HYPRE_Int i1, i2, i3, i4; \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopDeclareK(1); \
zypre_BoxLoopDeclareK(2); \
zypre_BoxLoopDeclareK(3); \
zypre_BoxLoopDeclareK(4); \
zypre_BoxLoopInit(ndim, loop_size); \
zypre_BoxLoopInitK(1, dbox1, start1, stride1, i1); \
zypre_BoxLoopInitK(2, dbox2, start2, stride2, i2); \
zypre_BoxLoopInitK(3, dbox3, start3, stride3, i3); \
zypre_BoxLoopInitK(4, dbox4, start4, stride4, i4); \
OMP1 \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
HYPRE_Int i1, i2, i3, i4; \
zypre_BoxLoopSet(); \
zypre_BoxLoopSetK(1, i1); \
zypre_BoxLoopSetK(2, i2); \
zypre_BoxLoopSetK(3, i3); \
zypre_BoxLoopSetK(4, i4); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define zypre_BoxLoop4End(i1, i2, i3, i4) \
i1 += hypre__i0inc1; \
i2 += hypre__i0inc2; \
i3 += hypre__i0inc3; \
i4 += hypre__i0inc4; \
} \
zypre_BoxLoopInc1(); \
i1 += hypre__ikinc1[hypre__d]; \
i2 += hypre__ikinc2[hypre__d]; \
i3 += hypre__ikinc3[hypre__d]; \
i4 += hypre__ikinc4[hypre__d]; \
zypre_BoxLoopInc2(); \
} \
} \
}
#define zypre_BasicBoxLoop1Begin(ndim, loop_size, \
stride1, i1) \
{ \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopDeclareK(1); \
zypre_BoxLoopInit(ndim, loop_size); \
zypre_BasicBoxLoopInitK(1, stride1); \
OMP1 \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
HYPRE_Int i1; \
zypre_BoxLoopSet(); \
zypre_BoxLoopSetK(1, i1); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define zypre_BasicBoxLoop2Begin(ndim, loop_size, \
stride1, i1, \
stride2, i2) \
{ \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopDeclareK(1); \
zypre_BoxLoopDeclareK(2); \
zypre_BoxLoopInit(ndim, loop_size); \
zypre_BasicBoxLoopInitK(1, stride1); \
zypre_BasicBoxLoopInitK(2, stride2); \
OMP1 \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
HYPRE_Int i1, i2; \
zypre_BoxLoopSet(); \
zypre_BoxLoopSetK(1, i1); \
zypre_BoxLoopSetK(2, i2); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define zypre_LoopBegin(size, idx) \
{ \
HYPRE_Int idx; \
OMP0 \
for (idx = 0; idx < size; idx ++) \
{
#define zypre_LoopEnd() \
} \
}
/*--------------------------------------------------------------------------
* Serial BoxLoop macros:
* [same as the ones above (without OMP and with SetOneBlock)]
* TODO: combine them
*--------------------------------------------------------------------------*/
#define hypre_SerialBoxLoop0Begin(ndim, loop_size) \
{ \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopInit(ndim, loop_size); \
zypre_BoxLoopSetOneBlock(); \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
zypre_BoxLoopSet(); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define hypre_SerialBoxLoop0End() \
} \
zypre_BoxLoopInc1(); \
zypre_BoxLoopInc2(); \
} \
} \
}
#define hypre_SerialBoxLoop1Begin(ndim, loop_size, \
dbox1, start1, stride1, i1) \
{ \
HYPRE_Int i1; \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopDeclareK(1); \
zypre_BoxLoopInit(ndim, loop_size); \
zypre_BoxLoopInitK(1, dbox1, start1, stride1, i1); \
zypre_BoxLoopSetOneBlock(); \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
zypre_BoxLoopSet(); \
zypre_BoxLoopSetK(1, i1); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define hypre_SerialBoxLoop1End(i1) zypre_BoxLoop1End(i1)
#define hypre_SerialBoxLoop2Begin(ndim, loop_size, \
dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2) \
{ \
HYPRE_Int i1,i2; \
zypre_BoxLoopDeclare(); \
zypre_BoxLoopDeclareK(1); \
zypre_BoxLoopDeclareK(2); \
zypre_BoxLoopInit(ndim, loop_size); \
zypre_BoxLoopInitK(1, dbox1, start1, stride1, i1); \
zypre_BoxLoopInitK(2, dbox2, start2, stride2, i2); \
zypre_BoxLoopSetOneBlock(); \
for (hypre__block = 0; hypre__block < hypre__num_blocks; hypre__block++) \
{ \
zypre_BoxLoopSet(); \
zypre_BoxLoopSetK(1, i1); \
zypre_BoxLoopSetK(2, i2); \
for (hypre__J = 0; hypre__J < hypre__JN; hypre__J++) \
{ \
for (hypre__I = 0; hypre__I < hypre__IN; hypre__I++) \
{
#define hypre_SerialBoxLoop2End(i1, i2) zypre_BoxLoop2End(i1, i2)
/* Reduction BoxLoop1 */
#define zypre_BoxLoop1ReductionBegin(ndim, loop_size, dbox1, start1, stride1, i1, reducesum) \
zypre_BoxLoop1Begin(ndim, loop_size, dbox1, start1, stride1, i1)
#define zypre_BoxLoop1ReductionEnd(i1, reducesum) zypre_BoxLoop1End(i1)
/* Reduction BoxLoop2 */
#define zypre_BoxLoop2ReductionBegin(ndim, loop_size, dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2, reducesum) \
zypre_BoxLoop2Begin(ndim, loop_size, dbox1, start1, stride1, i1, \
dbox2, start2, stride2, i2)
#define zypre_BoxLoop2ReductionEnd(i1, i2, reducesum) zypre_BoxLoop2End(i1, i2)
/* Renaming */
#define hypre_BoxLoopGetIndexHost zypre_BoxLoopGetIndex
#define hypre_BoxLoopBlockHost zypre_BoxLoopBlock
#define hypre_BoxLoop0BeginHost zypre_BoxLoop0Begin
#define hypre_BoxLoop0EndHost zypre_BoxLoop0End
#define hypre_BoxLoop1BeginHost zypre_BoxLoop1Begin
#define hypre_BoxLoop1EndHost zypre_BoxLoop1End
#define hypre_BoxLoop2BeginHost zypre_BoxLoop2Begin
#define hypre_BoxLoop2EndHost zypre_BoxLoop2End
#define hypre_BoxLoop3BeginHost zypre_BoxLoop3Begin
#define hypre_BoxLoop3EndHost zypre_BoxLoop3End
#define hypre_BoxLoop4BeginHost zypre_BoxLoop4Begin
#define hypre_BoxLoop4EndHost zypre_BoxLoop4End
#define hypre_BasicBoxLoop1BeginHost zypre_BasicBoxLoop1Begin
#define hypre_BasicBoxLoop2BeginHost zypre_BasicBoxLoop2Begin
#define hypre_LoopBeginHost zypre_LoopBegin
#define hypre_LoopEndHost zypre_LoopEnd
#define hypre_BoxLoop1ReductionBeginHost zypre_BoxLoop1ReductionBegin
#define hypre_BoxLoop1ReductionEndHost zypre_BoxLoop1ReductionEnd
#define hypre_BoxLoop2ReductionBeginHost zypre_BoxLoop2ReductionBegin
#define hypre_BoxLoop2ReductionEndHost zypre_BoxLoop2ReductionEnd
//TODO TEMP FIX
#if !defined(HYPRE_USING_RAJA) && !defined(HYPRE_USING_KOKKOS) && !defined(HYPRE_USING_CUDA) && !defined(HYPRE_USING_HIP) && !defined(HYPRE_USING_DEVICE_OPENMP) && !defined(HYPRE_USING_SYCL)
#define hypre_BoxLoopGetIndex hypre_BoxLoopGetIndexHost
#define hypre_BoxLoopBlock() 0
#define hypre_BoxLoop0Begin hypre_BoxLoop0BeginHost
#define hypre_BoxLoop0End hypre_BoxLoop0EndHost
#define hypre_BoxLoop1Begin hypre_BoxLoop1BeginHost
#define hypre_BoxLoop1End hypre_BoxLoop1EndHost
#define hypre_BoxLoop2Begin hypre_BoxLoop2BeginHost
#define hypre_BoxLoop2End hypre_BoxLoop2EndHost
#define hypre_BoxLoop3Begin hypre_BoxLoop3BeginHost
#define hypre_BoxLoop3End hypre_BoxLoop3EndHost
#define hypre_BoxLoop4Begin hypre_BoxLoop4BeginHost
#define hypre_BoxLoop4End hypre_BoxLoop4EndHost
#define hypre_BasicBoxLoop1Begin hypre_BasicBoxLoop1BeginHost
#define hypre_BasicBoxLoop2Begin hypre_BasicBoxLoop2BeginHost
#define hypre_LoopBegin hypre_LoopBeginHost
#define hypre_LoopEnd hypre_LoopEndHost
#define hypre_BoxLoop1ReductionBegin hypre_BoxLoop1ReductionBeginHost
#define hypre_BoxLoop1ReductionEnd hypre_BoxLoop1ReductionEndHost
#define hypre_BoxLoop2ReductionBegin hypre_BoxLoop2ReductionBeginHost
#define hypre_BoxLoop2ReductionEnd hypre_BoxLoop2ReductionEndHost
#endif
#endif /* #ifndef HYPRE_BOXLOOP_HOST_HEADER */
#ifdef __cplusplus
}
#endif
#endif
Reference in New Issue View Git Blame Copy Permalink