From 51475c7c4e99ed8faf0a644c51f7b001cf768463 Mon Sep 17 00:00:00 2001 From: Jeremy L Thompson Date: Thu, 20 Apr 2023 10:22:57 -0600 Subject: [PATCH 1/2] minor - clean up backend headers for const and argument names --- backends/cuda-ref/ceed-cuda-ref-basis.c | 2 +- backends/cuda-ref/ceed-cuda-ref.h | 14 +++++++------- backends/cuda/ceed-cuda-compile.h | 9 ++++----- backends/hip-gen/ceed-hip-gen-operator-build.h | 2 +- backends/hip-ref/ceed-hip-ref-basis.c | 2 +- backends/hip-ref/ceed-hip-ref.h | 11 +++++------ backends/hip/ceed-hip-compile.h | 10 +++++----- backends/magma/ceed-magma.h | 2 +- 8 files changed, 25 insertions(+), 27 deletions(-) diff --git a/backends/cuda-ref/ceed-cuda-ref-basis.c b/backends/cuda-ref/ceed-cuda-ref-basis.c index 103c424939..976eda2511 100644 --- a/backends/cuda-ref/ceed-cuda-ref-basis.c +++ b/backends/cuda-ref/ceed-cuda-ref-basis.c @@ -251,7 +251,7 @@ int CeedBasisCreateTensorH1_Cuda(CeedInt dim, CeedInt P_1d, CeedInt Q_1d, const // Create non-tensor //------------------------------------------------------------------------------ int CeedBasisCreateH1_Cuda(CeedElemTopology topo, CeedInt dim, CeedInt num_nodes, CeedInt num_qpts, const CeedScalar *interp, const CeedScalar *grad, - const CeedScalar *qref, const CeedScalar *q_weight, CeedBasis basis) { + const CeedScalar *q_ref, const CeedScalar *q_weight, CeedBasis basis) { Ceed ceed; CeedCallBackend(CeedBasisGetCeed(basis, &ceed)); CeedBasisNonTensor_Cuda *data; diff --git a/backends/cuda-ref/ceed-cuda-ref.h b/backends/cuda-ref/ceed-cuda-ref.h index 308ee1724b..7a1a1abae1 100644 --- a/backends/cuda-ref/ceed-cuda-ref.h +++ b/backends/cuda-ref/ceed-cuda-ref.h @@ -115,19 +115,19 @@ CEED_INTERN int CeedVectorCreate_Cuda(CeedSize n, CeedVector vec); CEED_INTERN int CeedElemRestrictionCreate_Cuda(CeedMemType mem_type, CeedCopyMode copy_mode, const CeedInt *indices, CeedElemRestriction r); -CEED_INTERN int CeedElemRestrictionCreateBlocked_Cuda(const CeedMemType mem_type, const CeedCopyMode copy_mode, const CeedInt *indices, - const CeedElemRestriction res); +CEED_INTERN int CeedElemRestrictionCreateBlocked_Cuda(const CeedMemType mem_type, CeedCopyMode copy_mode, const CeedInt *indices, + CeedElemRestriction res); -CEED_INTERN int CeedBasisApplyElems_Cuda(CeedBasis basis, const CeedInt num_elem, CeedTransposeMode t_mode, CeedEvalMode eval_mode, - const CeedVector u, CeedVector v); +CEED_INTERN int CeedBasisApplyElems_Cuda(CeedBasis basis, CeedInt num_elem, CeedTransposeMode t_mode, CeedEvalMode eval_mode, const CeedVector u, + CeedVector v); -CEED_INTERN int CeedQFunctionApplyElems_Cuda(CeedQFunction qf, const CeedInt Q, const CeedVector *const u, const CeedVector *v); +CEED_INTERN int CeedQFunctionApplyElems_Cuda(CeedQFunction qf, CeedInt Q, const CeedVector *const u, const CeedVector *v); CEED_INTERN int CeedBasisCreateTensorH1_Cuda(CeedInt dim, CeedInt P_1d, CeedInt Q_1d, const CeedScalar *interp_1d, const CeedScalar *grad_1d, const CeedScalar *q_ref_1d, const CeedScalar *q_weight_1d, CeedBasis basis); -CEED_INTERN int CeedBasisCreateH1_Cuda(CeedElemTopology, CeedInt, CeedInt, CeedInt, const CeedScalar *, const CeedScalar *, const CeedScalar *, - const CeedScalar *, CeedBasis); +CEED_INTERN int CeedBasisCreateH1_Cuda(CeedElemTopology topo, CeedInt dim, CeedInt num_nodes, CeedInt num_qpts, const CeedScalar *interp, + const CeedScalar *grad, const CeedScalar *q_ref, const CeedScalar *q_weight, CeedBasis basis); CEED_INTERN int CeedQFunctionCreate_Cuda(CeedQFunction qf); diff --git a/backends/cuda/ceed-cuda-compile.h b/backends/cuda/ceed-cuda-compile.h index a4bdfc7de3..72e898e9e8 100644 --- a/backends/cuda/ceed-cuda-compile.h +++ b/backends/cuda/ceed-cuda-compile.h @@ -19,14 +19,13 @@ CEED_INTERN int CeedCompileCuda(Ceed ceed, const char *source, CUmodule *module, CEED_INTERN int CeedGetKernelCuda(Ceed ceed, CUmodule module, const char *name, CUfunction *kernel); -CEED_INTERN int CeedRunKernelCuda(Ceed ceed, CUfunction kernel, const int grid_size, const int block_size, void **args); +CEED_INTERN int CeedRunKernelCuda(Ceed ceed, CUfunction kernel, int grid_size, int block_size, void **args); CEED_INTERN int CeedRunKernelAutoblockCuda(Ceed ceed, CUfunction kernel, size_t points, void **args); -CEED_INTERN int CeedRunKernelDimCuda(Ceed ceed, CUfunction kernel, const int grid_size, const int block_size_x, const int block_size_y, - const int block_size_z, void **args); +CEED_INTERN int CeedRunKernelDimCuda(Ceed ceed, CUfunction kernel, int grid_size, int block_size_x, int block_size_y, int block_size_z, void **args); -CEED_INTERN int CeedRunKernelDimSharedCuda(Ceed ceed, CUfunction kernel, const int grid_size, const int block_size_x, const int block_size_y, - const int block_size_z, const int shared_mem_size, void **args); +CEED_INTERN int CeedRunKernelDimSharedCuda(Ceed ceed, CUfunction kernel, int grid_size, int block_size_x, int block_size_y, int block_size_z, + int shared_mem_size, void **args); #endif // _ceed_cuda_compile_h diff --git a/backends/hip-gen/ceed-hip-gen-operator-build.h b/backends/hip-gen/ceed-hip-gen-operator-build.h index 5e72045fc2..c41efcc947 100644 --- a/backends/hip-gen/ceed-hip-gen-operator-build.h +++ b/backends/hip-gen/ceed-hip-gen-operator-build.h @@ -8,7 +8,7 @@ #ifndef _ceed_hip_gen_operator_build_h #define _ceed_hip_gen_operator_build_h -CEED_INTERN int BlockGridCalculate_Hip_gen(const CeedInt dim, const CeedInt num_elem, const CeedInt P_1d, const CeedInt Q_1d, CeedInt *block_sizes); +CEED_INTERN int BlockGridCalculate_Hip_gen(const CeedInt dim, CeedInt num_elem, CeedInt P_1d, CeedInt Q_1d, CeedInt *block_sizes); CEED_INTERN int CeedHipGenOperatorBuild(CeedOperator op); #endif // _ceed_hip_gen_operator_build_h diff --git a/backends/hip-ref/ceed-hip-ref-basis.c b/backends/hip-ref/ceed-hip-ref-basis.c index 26c1f239d1..95a39824c8 100644 --- a/backends/hip-ref/ceed-hip-ref-basis.c +++ b/backends/hip-ref/ceed-hip-ref-basis.c @@ -247,7 +247,7 @@ int CeedBasisCreateTensorH1_Hip(CeedInt dim, CeedInt P_1d, CeedInt Q_1d, const C // Create non-tensor //------------------------------------------------------------------------------ int CeedBasisCreateH1_Hip(CeedElemTopology topo, CeedInt dim, CeedInt num_nodes, CeedInt num_qpts, const CeedScalar *interp, const CeedScalar *grad, - const CeedScalar *qref, const CeedScalar *q_weight, CeedBasis basis) { + const CeedScalar *q_ref, const CeedScalar *q_weight, CeedBasis basis) { Ceed ceed; CeedCallBackend(CeedBasisGetCeed(basis, &ceed)); CeedBasisNonTensor_Hip *data; diff --git a/backends/hip-ref/ceed-hip-ref.h b/backends/hip-ref/ceed-hip-ref.h index 7b05f3bbc1..ebcb0a6791 100644 --- a/backends/hip-ref/ceed-hip-ref.h +++ b/backends/hip-ref/ceed-hip-ref.h @@ -116,19 +116,18 @@ CEED_INTERN int CeedVectorCreate_Hip(CeedSize n, CeedVector vec); CEED_INTERN int CeedElemRestrictionCreate_Hip(CeedMemType mtype, CeedCopyMode cmode, const CeedInt *indices, CeedElemRestriction r); -CEED_INTERN int CeedElemRestrictionCreateBlocked_Hip(const CeedMemType mtype, const CeedCopyMode cmode, const CeedInt *indices, - const CeedElemRestriction res); +CEED_INTERN int CeedElemRestrictionCreateBlocked_Hip(CeedMemType mtype, CeedCopyMode cmode, const CeedInt *indices, CeedElemRestriction res); -CEED_INTERN int CeedBasisApplyElems_Hip(CeedBasis basis, const CeedInt nelem, CeedTransposeMode tmode, CeedEvalMode emode, const CeedVector u, +CEED_INTERN int CeedBasisApplyElems_Hip(CeedBasis basis, CeedInt nelem, CeedTransposeMode tmode, CeedEvalMode emode, const CeedVector u, CeedVector v); -CEED_INTERN int CeedQFunctionApplyElems_Hip(CeedQFunction qf, const CeedInt Q, const CeedVector *const u, const CeedVector *v); +CEED_INTERN int CeedQFunctionApplyElems_Hip(CeedQFunction qf, CeedInt Q, const CeedVector *const u, const CeedVector *v); CEED_INTERN int CeedBasisCreateTensorH1_Hip(CeedInt dim, CeedInt P_1d, CeedInt Q_1d, const CeedScalar *interp_1d, const CeedScalar *grad_1d, const CeedScalar *q_ref_1d, const CeedScalar *q_weight_1d, CeedBasis basis); -CEED_INTERN int CeedBasisCreateH1_Hip(CeedElemTopology, CeedInt, CeedInt, CeedInt, const CeedScalar *, const CeedScalar *, const CeedScalar *, - const CeedScalar *, CeedBasis); +CEED_INTERN int CeedBasisCreateH1_Hip(CeedElemTopology topo, CeedInt dim, CeedInt num_nodes, CeedInt num_qpts, const CeedScalar *interp, + const CeedScalar *grad, const CeedScalar *q_ref, const CeedScalar *q_weight, CeedBasis basis); CEED_INTERN int CeedQFunctionCreate_Hip(CeedQFunction qf); diff --git a/backends/hip/ceed-hip-compile.h b/backends/hip/ceed-hip-compile.h index 41c7978e9f..3cda9a960b 100644 --- a/backends/hip/ceed-hip-compile.h +++ b/backends/hip/ceed-hip-compile.h @@ -18,12 +18,12 @@ CEED_INTERN int CeedCompileHip(Ceed ceed, const char *source, hipModule_t *modul CEED_INTERN int CeedGetKernelHip(Ceed ceed, hipModule_t module, const char *name, hipFunction_t *kernel); -CEED_INTERN int CeedRunKernelHip(Ceed ceed, hipFunction_t kernel, const int grid_size, const int block_size, void **args); +CEED_INTERN int CeedRunKernelHip(Ceed ceed, hipFunction_t kernel, int grid_size, int block_size, void **args); -CEED_INTERN int CeedRunKernelDimHip(Ceed ceed, hipFunction_t kernel, const int grid_size, const int block_size_x, const int block_size_y, - const int block_size_z, void **args); +CEED_INTERN int CeedRunKernelDimHip(Ceed ceed, hipFunction_t kernel, int grid_size, int block_size_x, int block_size_y, int block_size_z, + void **args); -CEED_INTERN int CeedRunKernelDimSharedHip(Ceed ceed, hipFunction_t kernel, const int grid_size, const int block_size_x, const int block_size_y, - const int block_size_z, const int shared_mem_size, void **args); +CEED_INTERN int CeedRunKernelDimSharedHip(Ceed ceed, hipFunction_t kernel, int grid_size, int block_size_x, int block_size_y, int block_size_z, + int shared_mem_size, void **args); #endif // _ceed_hip_compile_h diff --git a/backends/magma/ceed-magma.h b/backends/magma/ceed-magma.h index 1406c97cb9..f971627098 100644 --- a/backends/magma/ceed-magma.h +++ b/backends/magma/ceed-magma.h @@ -137,7 +137,7 @@ CEED_INTERN int CeedBasisCreateH1_Magma(CeedElemTopology topo, CeedInt dim, Ceed CEED_INTERN int CeedElemRestrictionCreate_Magma(CeedMemType mtype, CeedCopyMode cmode, const CeedInt *offsets, CeedElemRestriction r); CEED_INTERN int CeedElemRestrictionCreateBlocked_Magma(const CeedMemType mtype, const CeedCopyMode cmode, const CeedInt *offsets, - const CeedElemRestriction res); + CeedElemRestriction res); // comment the line below to use the default magma_is_devptr function #define magma_is_devptr magma_isdevptr From 472941f0c10177d86dfda5d1965729671a5c15be Mon Sep 17 00:00:00 2001 From: Jeremy L Thompson Date: Fri, 21 Apr 2023 10:02:53 -0600 Subject: [PATCH 2/2] minor - fix static vs CEED_INTERN in backend file --- backends/blocked/ceed-blocked.c | 2 +- backends/cuda-ref/ceed-cuda-ref.h | 8 -------- backends/cuda-ref/ceed-cuda-restriction.c | 8 ++++---- backends/hip-gen/ceed-hip-gen-operator-build.h | 2 +- backends/hip-ref/ceed-hip-ref-restriction.c | 12 ++++++------ backends/hip-ref/ceed-hip-ref.h | 9 +-------- backends/magma/ceed-magma.h | 3 --- backends/ref/ceed-ref.h | 3 +-- 8 files changed, 14 insertions(+), 33 deletions(-) diff --git a/backends/blocked/ceed-blocked.c b/backends/blocked/ceed-blocked.c index 27a0978e38..c1718a1040 100644 --- a/backends/blocked/ceed-blocked.c +++ b/backends/blocked/ceed-blocked.c @@ -15,7 +15,7 @@ //------------------------------------------------------------------------------ // Backend Init //------------------------------------------------------------------------------ -CEED_INTERN int CeedInit_Blocked(const char *resource, Ceed ceed) { +static int CeedInit_Blocked(const char *resource, Ceed ceed) { CeedCheck(!strcmp(resource, "/cpu/self") || !strcmp(resource, "/cpu/self/ref/blocked"), ceed, CEED_ERROR_BACKEND, "Blocked backend cannot use resource: %s", resource); CeedCallBackend(CeedSetDeterministic(ceed, true)); diff --git a/backends/cuda-ref/ceed-cuda-ref.h b/backends/cuda-ref/ceed-cuda-ref.h index 7a1a1abae1..77af061ff3 100644 --- a/backends/cuda-ref/ceed-cuda-ref.h +++ b/backends/cuda-ref/ceed-cuda-ref.h @@ -115,14 +115,6 @@ CEED_INTERN int CeedVectorCreate_Cuda(CeedSize n, CeedVector vec); CEED_INTERN int CeedElemRestrictionCreate_Cuda(CeedMemType mem_type, CeedCopyMode copy_mode, const CeedInt *indices, CeedElemRestriction r); -CEED_INTERN int CeedElemRestrictionCreateBlocked_Cuda(const CeedMemType mem_type, CeedCopyMode copy_mode, const CeedInt *indices, - CeedElemRestriction res); - -CEED_INTERN int CeedBasisApplyElems_Cuda(CeedBasis basis, CeedInt num_elem, CeedTransposeMode t_mode, CeedEvalMode eval_mode, const CeedVector u, - CeedVector v); - -CEED_INTERN int CeedQFunctionApplyElems_Cuda(CeedQFunction qf, CeedInt Q, const CeedVector *const u, const CeedVector *v); - CEED_INTERN int CeedBasisCreateTensorH1_Cuda(CeedInt dim, CeedInt P_1d, CeedInt Q_1d, const CeedScalar *interp_1d, const CeedScalar *grad_1d, const CeedScalar *q_ref_1d, const CeedScalar *q_weight_1d, CeedBasis basis); diff --git a/backends/cuda-ref/ceed-cuda-restriction.c b/backends/cuda-ref/ceed-cuda-restriction.c index 92f1ab0d21..6104b10be4 100644 --- a/backends/cuda-ref/ceed-cuda-restriction.c +++ b/backends/cuda-ref/ceed-cuda-restriction.c @@ -90,11 +90,11 @@ static int CeedElemRestrictionApply_Cuda(CeedElemRestriction r, CeedTransposeMod //------------------------------------------------------------------------------ // Get offsets //------------------------------------------------------------------------------ -static int CeedElemRestrictionGetOffsets_Cuda(CeedElemRestriction rstr, CeedMemType m_type, const CeedInt **offsets) { +static int CeedElemRestrictionGetOffsets_Cuda(CeedElemRestriction rstr, CeedMemType mem_type, const CeedInt **offsets) { CeedElemRestriction_Cuda *impl; CeedCallBackend(CeedElemRestrictionGetData(rstr, &impl)); - switch (m_type) { + switch (mem_type) { case CEED_MEM_HOST: *offsets = impl->h_ind; break; @@ -209,7 +209,7 @@ static int CeedElemRestrictionOffset_Cuda(const CeedElemRestriction r, const Cee //------------------------------------------------------------------------------ // Create restriction //------------------------------------------------------------------------------ -int CeedElemRestrictionCreate_Cuda(CeedMemType m_type, CeedCopyMode copy_mode, const CeedInt *indices, CeedElemRestriction r) { +int CeedElemRestrictionCreate_Cuda(CeedMemType mem_type, CeedCopyMode copy_mode, const CeedInt *indices, CeedElemRestriction r) { Ceed ceed; CeedCallBackend(CeedElemRestrictionGetCeed(r, &ceed)); CeedElemRestriction_Cuda *impl; @@ -247,7 +247,7 @@ int CeedElemRestrictionCreate_Cuda(CeedMemType m_type, CeedCopyMode copy_mode, c CeedCallBackend(CeedElemRestrictionSetELayout(r, layout)); // Set up device indices/offset arrays - switch (m_type) { + switch (mem_type) { case CEED_MEM_HOST: { switch (copy_mode) { case CEED_OWN_POINTER: diff --git a/backends/hip-gen/ceed-hip-gen-operator-build.h b/backends/hip-gen/ceed-hip-gen-operator-build.h index c41efcc947..fae58e2614 100644 --- a/backends/hip-gen/ceed-hip-gen-operator-build.h +++ b/backends/hip-gen/ceed-hip-gen-operator-build.h @@ -8,7 +8,7 @@ #ifndef _ceed_hip_gen_operator_build_h #define _ceed_hip_gen_operator_build_h -CEED_INTERN int BlockGridCalculate_Hip_gen(const CeedInt dim, CeedInt num_elem, CeedInt P_1d, CeedInt Q_1d, CeedInt *block_sizes); +CEED_INTERN int BlockGridCalculate_Hip_gen(CeedInt dim, CeedInt num_elem, CeedInt P_1d, CeedInt Q_1d, CeedInt *block_sizes); CEED_INTERN int CeedHipGenOperatorBuild(CeedOperator op); #endif // _ceed_hip_gen_operator_build_h diff --git a/backends/hip-ref/ceed-hip-ref-restriction.c b/backends/hip-ref/ceed-hip-ref-restriction.c index 1c5a2116e1..cf7f05d6a3 100644 --- a/backends/hip-ref/ceed-hip-ref-restriction.c +++ b/backends/hip-ref/ceed-hip-ref-restriction.c @@ -88,11 +88,11 @@ static int CeedElemRestrictionApply_Hip(CeedElemRestriction r, CeedTransposeMode //------------------------------------------------------------------------------ // Get offsets //------------------------------------------------------------------------------ -static int CeedElemRestrictionGetOffsets_Hip(CeedElemRestriction rstr, CeedMemType mtype, const CeedInt **offsets) { +static int CeedElemRestrictionGetOffsets_Hip(CeedElemRestriction rstr, CeedMemType mem_type, const CeedInt **offsets) { CeedElemRestriction_Hip *impl; CeedCallBackend(CeedElemRestrictionGetData(rstr, &impl)); - switch (mtype) { + switch (mem_type) { case CEED_MEM_HOST: *offsets = impl->h_ind; break; @@ -207,7 +207,7 @@ static int CeedElemRestrictionOffset_Hip(const CeedElemRestriction r, const Ceed //------------------------------------------------------------------------------ // Create restriction //------------------------------------------------------------------------------ -int CeedElemRestrictionCreate_Hip(CeedMemType mtype, CeedCopyMode cmode, const CeedInt *indices, CeedElemRestriction r) { +int CeedElemRestrictionCreate_Hip(CeedMemType mem_type, CeedCopyMode copy_mode, const CeedInt *indices, CeedElemRestriction r) { Ceed ceed; CeedCallBackend(CeedElemRestrictionGetCeed(r, &ceed)); CeedElemRestriction_Hip *impl; @@ -245,9 +245,9 @@ int CeedElemRestrictionCreate_Hip(CeedMemType mtype, CeedCopyMode cmode, const C CeedCallBackend(CeedElemRestrictionSetELayout(r, layout)); // Set up device indices/offset arrays - switch (mtype) { + switch (mem_type) { case CEED_MEM_HOST: { - switch (cmode) { + switch (copy_mode) { case CEED_OWN_POINTER: impl->h_ind_allocated = (CeedInt *)indices; impl->h_ind = (CeedInt *)indices; @@ -272,7 +272,7 @@ int CeedElemRestrictionCreate_Hip(CeedMemType mtype, CeedCopyMode cmode, const C break; } case CEED_MEM_DEVICE: { - switch (cmode) { + switch (copy_mode) { case CEED_COPY_VALUES: if (indices != NULL) { CeedCallHip(ceed, hipMalloc((void **)&impl->d_ind, size * sizeof(CeedInt))); diff --git a/backends/hip-ref/ceed-hip-ref.h b/backends/hip-ref/ceed-hip-ref.h index ebcb0a6791..807666761a 100644 --- a/backends/hip-ref/ceed-hip-ref.h +++ b/backends/hip-ref/ceed-hip-ref.h @@ -114,14 +114,7 @@ CEED_INTERN int CeedHipGetHipblasHandle(Ceed ceed, hipblasHandle_t *handle); CEED_INTERN int CeedVectorCreate_Hip(CeedSize n, CeedVector vec); -CEED_INTERN int CeedElemRestrictionCreate_Hip(CeedMemType mtype, CeedCopyMode cmode, const CeedInt *indices, CeedElemRestriction r); - -CEED_INTERN int CeedElemRestrictionCreateBlocked_Hip(CeedMemType mtype, CeedCopyMode cmode, const CeedInt *indices, CeedElemRestriction res); - -CEED_INTERN int CeedBasisApplyElems_Hip(CeedBasis basis, CeedInt nelem, CeedTransposeMode tmode, CeedEvalMode emode, const CeedVector u, - CeedVector v); - -CEED_INTERN int CeedQFunctionApplyElems_Hip(CeedQFunction qf, CeedInt Q, const CeedVector *const u, const CeedVector *v); +CEED_INTERN int CeedElemRestrictionCreate_Hip(CeedMemType mem_type, CeedCopyMode copy_mode, const CeedInt *indices, CeedElemRestriction r); CEED_INTERN int CeedBasisCreateTensorH1_Hip(CeedInt dim, CeedInt P_1d, CeedInt Q_1d, const CeedScalar *interp_1d, const CeedScalar *grad_1d, const CeedScalar *q_ref_1d, const CeedScalar *q_weight_1d, CeedBasis basis); diff --git a/backends/magma/ceed-magma.h b/backends/magma/ceed-magma.h index f971627098..1996cc10ba 100644 --- a/backends/magma/ceed-magma.h +++ b/backends/magma/ceed-magma.h @@ -136,9 +136,6 @@ CEED_INTERN int CeedBasisCreateH1_Magma(CeedElemTopology topo, CeedInt dim, Ceed CEED_INTERN int CeedElemRestrictionCreate_Magma(CeedMemType mtype, CeedCopyMode cmode, const CeedInt *offsets, CeedElemRestriction r); -CEED_INTERN int CeedElemRestrictionCreateBlocked_Magma(const CeedMemType mtype, const CeedCopyMode cmode, const CeedInt *offsets, - CeedElemRestriction res); - // comment the line below to use the default magma_is_devptr function #define magma_is_devptr magma_isdevptr diff --git a/backends/ref/ceed-ref.h b/backends/ref/ceed-ref.h index 325accfb13..168ba273be 100644 --- a/backends/ref/ceed-ref.h +++ b/backends/ref/ceed-ref.h @@ -30,8 +30,7 @@ typedef struct { // Orientation, if it exists, is true when the face must be flipped (multiplies by -1.). const bool *orient; bool *orient_allocated; - int (*Apply)(CeedElemRestriction, const CeedInt, const CeedInt, const CeedInt, CeedInt, CeedInt, CeedTransposeMode, CeedVector, CeedVector, - CeedRequest *); + int (*Apply)(CeedElemRestriction, CeedInt, CeedInt, CeedInt, CeedInt, CeedInt, CeedTransposeMode, CeedVector, CeedVector, CeedRequest *); } CeedElemRestriction_Ref; typedef struct {