From: Daniel Arndt <arndtd@ornl.gov>
Date: Thu, 2 Mar 2023 08:52:14 +0000 (+0100)
Subject: Introduce DEAL_II_HOST
X-Git-Tag: v9.5.0-rc1~507^2
X-Git-Url: https://gitweb.dealii.org/cgi-bin/gitweb.cgi?a=commitdiff_plain;h=529695f7e3569830ab0c3ca252537b6d0f18bbd3;p=dealii.git

Introduce DEAL_II_HOST
---

diff --git a/include/deal.II/base/numbers.h b/include/deal.II/base/numbers.h
index 0e13bb2a41..5a16e72385 100644
--- a/include/deal.II/base/numbers.h
+++ b/include/deal.II/base/numbers.h
@@ -36,6 +36,17 @@
 #define DEAL_II_CUDA_HOST_DEV DEAL_II_HOST_DEVICE
 #define DEAL_II_HOST_DEVICE_ALWAYS_INLINE KOKKOS_FORCEINLINE_FUNCTION
 
+// clang++ assumes that all constexpr functions are __host__ __device__ when
+// compiling CUDA code, i.e, when Kokkos was configured with CUDA support.
+// This is problematic when calling non-constexpr functions in constexpr
+// functions. Hence, we need a way to annotate functions explicitly as
+// host-only.
+#if defined(__clang__) && defined(__CUDA__)
+#  define DEAL_II_HOST __host__
+#else
+#  define DEAL_II_HOST
+#endif
+
 // Forward-declare the automatic differentiation types so we can add prototypes
 // for our own wrappers.
 #ifdef DEAL_II_WITH_ADOLC
diff --git a/include/deal.II/base/symmetric_tensor.h b/include/deal.II/base/symmetric_tensor.h
index 90b2899d3f..4195faee00 100644
--- a/include/deal.II/base/symmetric_tensor.h
+++ b/include/deal.II/base/symmetric_tensor.h
@@ -45,11 +45,9 @@ class SymmetricTensor;
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number = double>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE SymmetricTensor<2, dim, Number>
-                                                 unit_symmetric_tensor();
+DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<2, dim, Number>
+  unit_symmetric_tensor();
 
 /**
  * Return the tensor of rank 4 that, when multiplied by a symmetric rank 2
@@ -80,11 +78,9 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number = double>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE SymmetricTensor<4, dim, Number>
-                                                 deviator_tensor();
+DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<4, dim, Number>
+  deviator_tensor();
 
 /**
  * Return the fourth-order symmetric identity tensor $\mathbb S$ which maps
@@ -124,25 +120,17 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number = double>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE SymmetricTensor<4, dim, Number>
-                                                 identity_tensor();
+DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<4, dim, Number>
+  identity_tensor();
 
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<2, dim, Number>
-  invert(const SymmetricTensor<2, dim, Number> &);
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<2, dim, Number>
+invert(const SymmetricTensor<2, dim, Number> &);
 
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<4, dim, Number>
-  invert(const SymmetricTensor<4, dim, Number> &);
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<4, dim, Number>
+invert(const SymmetricTensor<4, dim, Number> &);
 
 /**
  * Compute and return the trace of a tensor of rank 2, i.e. the sum of its
@@ -154,11 +142,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim2, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE Number
-  trace(const SymmetricTensor<2, dim2, Number> &);
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE Number
+trace(const SymmetricTensor<2, dim2, Number> &);
 
 /**
  * Compute the deviator of a symmetric tensor, which is defined as
@@ -171,10 +156,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<2, dim, Number>
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<2, dim, Number>
   deviator(const SymmetricTensor<2, dim, Number> &);
 
 /**
@@ -191,11 +174,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE Number
-  determinant(const SymmetricTensor<2, dim, Number> &);
+DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE Number
+determinant(const SymmetricTensor<2, dim, Number> &);
 
 
 
@@ -264,19 +244,17 @@ namespace internal
    */
   namespace SymmetricTensorAccessors
   {
-/**
- * Create a TableIndices<2> object where the first entries up to
- * <tt>position-1</tt> are taken from previous_indices, and new_index is
- * put at position <tt>position</tt>. The remaining indices remain in
- * invalid state.
- */
-#if defined(__clang__) && defined(__CUDA__)
-    __host__
-#endif
-      DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE TableIndices<2>
-      merge(const TableIndices<2> &previous_indices,
-            const unsigned int     new_index,
-            const unsigned int     position)
+    /**
+     * Create a TableIndices<2> object where the first entries up to
+     * <tt>position-1</tt> are taken from previous_indices, and new_index is
+     * put at position <tt>position</tt>. The remaining indices remain in
+     * invalid state.
+     */
+    DEAL_II_HOST
+    DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE TableIndices<2>
+    merge(const TableIndices<2> &previous_indices,
+          const unsigned int     new_index,
+          const unsigned int     position)
     {
       AssertIndexRange(position, 2);
 
@@ -288,19 +266,17 @@ namespace internal
 
 
 
-/**
- * Create a TableIndices<4> object where the first entries up to
- * <tt>position-1</tt> are taken from previous_indices, and new_index is
- * put at position <tt>position</tt>. The remaining indices remain in
- * invalid state.
- */
-#if defined(__clang__) && defined(__CUDA__)
-    __host__
-#endif
-      DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE TableIndices<4>
-      merge(const TableIndices<4> &previous_indices,
-            const unsigned int     new_index,
-            const unsigned int     position)
+    /**
+     * Create a TableIndices<4> object where the first entries up to
+     * <tt>position-1</tt> are taken from previous_indices, and new_index is
+     * put at position <tt>position</tt>. The remaining indices remain in
+     * invalid state.
+     */
+    DEAL_II_HOST
+    DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE TableIndices<4>
+    merge(const TableIndices<4> &previous_indices,
+          const unsigned int     new_index,
+          const unsigned int     position)
     {
       AssertIndexRange(position, 4);
 
@@ -514,57 +490,49 @@ namespace internal
         typename AccessorTypes<rank, dim, constness, Number>::tensor_type;
 
     private:
-/**
- * Constructor. Take a reference to the tensor object which we will
- * access.
- *
- * The second argument denotes the values of previous indices into the
- * tensor. For example, for a rank-4 tensor, if P=2, then we will
- * already have had two successive element selections (e.g. through
- * <tt>tensor[1][2]</tt>), and the two index values have to be stored
- * somewhere. This class therefore only makes use of the first rank-P
- * elements of this array, but passes it on to the next level with P-1
- * which fills the next entry, and so on.
- *
- * The constructor is made private in order to prevent you having such
- * objects around. The only way to create such objects is via the
- * <tt>Table</tt> class, which only generates them as temporary objects.
- * This guarantees that the accessor objects go out of scope earlier
- * than the mother object, avoid problems with data consistency.
- */
-#if defined(__clang__) && defined(__CUDA__)
-      __host__
-#endif
-        constexpr Accessor(tensor_type &             tensor,
-                           const TableIndices<rank> &previous_indices);
+      /**
+       * Constructor. Take a reference to the tensor object which we will
+       * access.
+       *
+       * The second argument denotes the values of previous indices into the
+       * tensor. For example, for a rank-4 tensor, if P=2, then we will
+       * already have had two successive element selections (e.g. through
+       * <tt>tensor[1][2]</tt>), and the two index values have to be stored
+       * somewhere. This class therefore only makes use of the first rank-P
+       * elements of this array, but passes it on to the next level with P-1
+       * which fills the next entry, and so on.
+       *
+       * The constructor is made private in order to prevent you having such
+       * objects around. The only way to create such objects is via the
+       * <tt>Table</tt> class, which only generates them as temporary objects.
+       * This guarantees that the accessor objects go out of scope earlier
+       * than the mother object, avoid problems with data consistency.
+       */
+      DEAL_II_HOST
+      constexpr Accessor(tensor_type &             tensor,
+                         const TableIndices<rank> &previous_indices);
 
-/**
- * Copy constructor.
- */
-#if defined(__clang__) && defined(__CUDA__)
-      __host__
-#endif
-        constexpr DEAL_II_ALWAYS_INLINE
-        Accessor(const Accessor &) = default;
+      /**
+       * Copy constructor.
+       */
+      DEAL_II_HOST
+      constexpr DEAL_II_ALWAYS_INLINE
+      Accessor(const Accessor &) = default;
 
     public:
-/**
- * Index operator.
- */
-#if defined(__clang__) && defined(__CUDA__)
-      __host__
-#endif
-        constexpr Accessor<rank, dim, constness, P - 1, Number>
-        operator[](const unsigned int i);
+      /**
+       * Index operator.
+       */
+      DEAL_II_HOST
+      constexpr Accessor<rank, dim, constness, P - 1, Number>
+      operator[](const unsigned int i);
 
-/**
- * Index operator.
- */
-#if defined(__clang__) && defined(__CUDA__)
-      __host__
-#endif
-        constexpr Accessor<rank, dim, constness, P - 1, Number>
-        operator[](const unsigned int i) const;
+      /**
+       * Index operator.
+       */
+      DEAL_II_HOST
+      constexpr Accessor<rank, dim, constness, P - 1, Number>
+      operator[](const unsigned int i) const;
 
     private:
       /**
@@ -605,60 +573,52 @@ namespace internal
         typename AccessorTypes<rank, dim, constness, Number>::tensor_type;
 
     private:
-/**
- * Constructor. Take a reference to the tensor object which we will
- * access.
- *
- * The second argument denotes the values of previous indices into the
- * tensor. For example, for a rank-4 tensor, if P=2, then we will
- * already have had two successive element selections (e.g. through
- * <tt>tensor[1][2]</tt>), and the two index values have to be stored
- * somewhere. This class therefore only makes use of the first rank-P
- * elements of this array, but passes it on to the next level with P-1
- * which fills the next entry, and so on.
- *
- * For this particular specialization, i.e. for P==1, all but the last
- * index are already filled.
- *
- * The constructor is made private in order to prevent you having such
- * objects around. The only way to create such objects is via the
- * <tt>Table</tt> class, which only generates them as temporary objects.
- * This guarantees that the accessor objects go out of scope earlier
- * than the mother object, avoid problems with data consistency.
- */
-#if defined(__clang__) && defined(__CUDA__)
-      __host__
-#endif
-        constexpr Accessor(tensor_type &             tensor,
-                           const TableIndices<rank> &previous_indices);
+      /**
+       * Constructor. Take a reference to the tensor object which we will
+       * access.
+       *
+       * The second argument denotes the values of previous indices into the
+       * tensor. For example, for a rank-4 tensor, if P=2, then we will
+       * already have had two successive element selections (e.g. through
+       * <tt>tensor[1][2]</tt>), and the two index values have to be stored
+       * somewhere. This class therefore only makes use of the first rank-P
+       * elements of this array, but passes it on to the next level with P-1
+       * which fills the next entry, and so on.
+       *
+       * For this particular specialization, i.e. for P==1, all but the last
+       * index are already filled.
+       *
+       * The constructor is made private in order to prevent you having such
+       * objects around. The only way to create such objects is via the
+       * <tt>Table</tt> class, which only generates them as temporary objects.
+       * This guarantees that the accessor objects go out of scope earlier
+       * than the mother object, avoid problems with data consistency.
+       */
+      DEAL_II_HOST
+      constexpr Accessor(tensor_type &             tensor,
+                         const TableIndices<rank> &previous_indices);
 
-/**
- * Copy constructor.
- */
-#if defined(__clang__) && defined(__CUDA__)
-      __host__
-#endif
-        constexpr DEAL_II_ALWAYS_INLINE
-        Accessor(const Accessor &) = default;
+      /**
+       * Copy constructor.
+       */
+      DEAL_II_HOST
+      constexpr DEAL_II_ALWAYS_INLINE
+      Accessor(const Accessor &) = default;
 
     public:
-/**
- * Index operator.
- */
-#if defined(__clang__) && defined(__CUDA__)
-      __host__
-#endif
-        constexpr reference
-        operator[](const unsigned int);
+      /**
+       * Index operator.
+       */
+      DEAL_II_HOST
+      constexpr reference
+      operator[](const unsigned int);
 
-/**
- * Index operator.
- */
-#if defined(__clang__) && defined(__CUDA__)
-      __host__
-#endif
-        constexpr reference
-        operator[](const unsigned int) const;
+      /**
+       * Index operator.
+       */
+      DEAL_II_HOST
+      constexpr reference
+      operator[](const unsigned int) const;
 
     private:
       /**
@@ -784,14 +744,12 @@ public:
     internal::SymmetricTensorAccessors::StorageType<rank_, dim, Number>::
       n_independent_components;
 
-/**
- * Default constructor. Creates a tensor with all entries equal to zero.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr DEAL_II_ALWAYS_INLINE
-    SymmetricTensor() = default;
+  /**
+   * Default constructor. Creates a tensor with all entries equal to zero.
+   */
+  DEAL_II_HOST
+  constexpr DEAL_II_ALWAYS_INLINE
+  SymmetricTensor() = default;
 
   /**
    * Constructor. Generate a symmetric tensor from a general one. Assumes that
@@ -809,25 +767,23 @@ public:
   template <typename OtherNumber>
   explicit SymmetricTensor(const Tensor<2, dim, OtherNumber> &t);
 
-/**
- * A constructor that creates a symmetric tensor from an array holding its
- * independent elements. Using this constructor assumes that the caller
- * knows the order in which elements are stored in symmetric tensors; its
- * use is therefore discouraged, but if you think you want to use it anyway
- * you can query the order of elements using the unrolled_index() function.
- *
- * This constructor is currently only implemented for symmetric tensors of
- * rank 2.
- *
- * The size of the array passed is equal to
- * SymmetricTensor<rank_,dim>::n_independent_components; the reason for using
- * the object from the internal namespace is to work around bugs in some
- * older compilers.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr SymmetricTensor(const Number (&array)[n_independent_components]);
+  /**
+   * A constructor that creates a symmetric tensor from an array holding its
+   * independent elements. Using this constructor assumes that the caller
+   * knows the order in which elements are stored in symmetric tensors; its
+   * use is therefore discouraged, but if you think you want to use it anyway
+   * you can query the order of elements using the unrolled_index() function.
+   *
+   * This constructor is currently only implemented for symmetric tensors of
+   * rank 2.
+   *
+   * The size of the array passed is equal to
+   * SymmetricTensor<rank_,dim>::n_independent_components; the reason for using
+   * the object from the internal namespace is to work around bugs in some
+   * older compilers.
+   */
+  DEAL_II_HOST
+  constexpr SymmetricTensor(const Number (&array)[n_independent_components]);
 
   /**
    * Copy constructor from tensors with different underlying scalar type. This
@@ -835,11 +791,8 @@ public:
    * Number.
    */
   template <typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr explicit SymmetricTensor(
-      const SymmetricTensor<rank_, dim, OtherNumber> &initializer);
+  DEAL_II_HOST constexpr explicit SymmetricTensor(
+    const SymmetricTensor<rank_, dim, OtherNumber> &initializer);
 
   /**
    * Return a pointer to the first element of the underlying storage.
@@ -897,101 +850,75 @@ public:
    * @p Number.
    */
   template <typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr SymmetricTensor &
-    operator=(const SymmetricTensor<rank_, dim, OtherNumber> &rhs);
+  DEAL_II_HOST constexpr SymmetricTensor &
+  operator=(const SymmetricTensor<rank_, dim, OtherNumber> &rhs);
 
-/**
- * This operator assigns a scalar to a tensor. To avoid confusion with what
- * exactly it means to assign a scalar value to a tensor, zero is the only
- * value allowed for <tt>d</tt>, allowing the intuitive notation
- * $\mathbf A = 0$ to reset all elements of the tensor to zero.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr SymmetricTensor &
-    operator=(const Number &d);
+  /**
+   * This operator assigns a scalar to a tensor. To avoid confusion with what
+   * exactly it means to assign a scalar value to a tensor, zero is the only
+   * value allowed for <tt>d</tt>, allowing the intuitive notation
+   * $\mathbf A = 0$ to reset all elements of the tensor to zero.
+   */
+  DEAL_II_HOST
+  constexpr SymmetricTensor &
+  operator=(const Number &d);
 
-/**
- * Convert the present symmetric tensor into a full tensor with the same
- * elements, but using the different storage scheme of full tensors.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr
-    operator Tensor<rank_, dim, Number>() const;
+  /**
+   * Convert the present symmetric tensor into a full tensor with the same
+   * elements, but using the different storage scheme of full tensors.
+   */
+  DEAL_II_HOST
+  constexpr operator Tensor<rank_, dim, Number>() const;
 
-/**
- * Test for equality of two tensors.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr bool
-    operator==(const SymmetricTensor &) const;
+  /**
+   * Test for equality of two tensors.
+   */
+  DEAL_II_HOST
+  constexpr bool
+  operator==(const SymmetricTensor &) const;
 
-/**
- * Test for inequality of two tensors.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr bool
-    operator!=(const SymmetricTensor &) const;
+  /**
+   * Test for inequality of two tensors.
+   */
+  DEAL_II_HOST
+  constexpr bool
+  operator!=(const SymmetricTensor &) const;
 
   /**
    * Add another tensor.
    */
   template <typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr SymmetricTensor &
-    operator+=(const SymmetricTensor<rank_, dim, OtherNumber> &);
+  DEAL_II_HOST constexpr SymmetricTensor &
+  operator+=(const SymmetricTensor<rank_, dim, OtherNumber> &);
 
   /**
    * Subtract another tensor.
    */
   template <typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr SymmetricTensor &
-    operator-=(const SymmetricTensor<rank_, dim, OtherNumber> &);
+  DEAL_II_HOST constexpr SymmetricTensor &
+  operator-=(const SymmetricTensor<rank_, dim, OtherNumber> &);
 
   /**
    * Scale the tensor by <tt>factor</tt>, i.e. multiply all components by
    * <tt>factor</tt>.
    */
   template <typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr SymmetricTensor &
-    operator*=(const OtherNumber &factor);
+  DEAL_II_HOST constexpr SymmetricTensor &
+  operator*=(const OtherNumber &factor);
 
   /**
    * Scale the tensor by <tt>1/factor</tt>.
    */
   template <typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr SymmetricTensor &
-    operator/=(const OtherNumber &factor);
+  DEAL_II_HOST constexpr SymmetricTensor &
+  operator/=(const OtherNumber &factor);
 
-/**
- * Unary minus operator. Negate all entries of a tensor.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr SymmetricTensor
-    operator-() const;
+  /**
+   * Unary minus operator. Negate all entries of a tensor.
+   */
+  DEAL_II_HOST
+  constexpr SymmetricTensor
+  operator-() const;
 
   /**
    * Double contraction product between the present symmetric tensor and a
@@ -1046,125 +973,101 @@ public:
    *   symmetric tensors.
    */
   template <typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    DEAL_II_CONSTEXPR typename internal::SymmetricTensorAccessors::
-      double_contraction_result<rank_, 2, dim, Number, OtherNumber>::type
-      operator*(const SymmetricTensor<2, dim, OtherNumber> &s) const;
+  DEAL_II_HOST DEAL_II_CONSTEXPR typename internal::SymmetricTensorAccessors::
+    double_contraction_result<rank_, 2, dim, Number, OtherNumber>::type
+    operator*(const SymmetricTensor<2, dim, OtherNumber> &s) const;
 
   /**
    * Contraction over the last two indices of the present object with the first
    * two indices of the rank-4 symmetric tensor given as argument.
    */
   template <typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    DEAL_II_CONSTEXPR typename internal::SymmetricTensorAccessors::
-      double_contraction_result<rank_, 4, dim, Number, OtherNumber>::type
-      operator*(const SymmetricTensor<4, dim, OtherNumber> &s) const;
+  DEAL_II_HOST DEAL_II_CONSTEXPR typename internal::SymmetricTensorAccessors::
+    double_contraction_result<rank_, 4, dim, Number, OtherNumber>::type
+    operator*(const SymmetricTensor<4, dim, OtherNumber> &s) const;
 
-/**
- * Return a read-write reference to the indicated element.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr Number &
-    operator()(const TableIndices<rank_> &indices);
+  /**
+   * Return a read-write reference to the indicated element.
+   */
+  DEAL_II_HOST
+  constexpr Number &
+  operator()(const TableIndices<rank_> &indices);
 
-/**
- * Return a @p const reference to the value referred to by the argument.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr const Number &
-    operator()(const TableIndices<rank_> &indices) const;
+  /**
+   * Return a @p const reference to the value referred to by the argument.
+   */
+  DEAL_II_HOST
+  constexpr const Number &
+  operator()(const TableIndices<rank_> &indices) const;
 
-/**
- * Access the elements of a row of this symmetric tensor. This function is
- * called for constant tensors.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr internal::SymmetricTensorAccessors::
-      Accessor<rank_, dim, true, rank_ - 1, Number>
-      operator[](const unsigned int row) const;
+  /**
+   * Access the elements of a row of this symmetric tensor. This function is
+   * called for constant tensors.
+   */
+  DEAL_II_HOST
+  constexpr internal::SymmetricTensorAccessors::
+    Accessor<rank_, dim, true, rank_ - 1, Number>
+    operator[](const unsigned int row) const;
 
-/**
- * Access the elements of a row of this symmetric tensor. This function is
- * called for non-constant tensors.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr internal::SymmetricTensorAccessors::
-      Accessor<rank_, dim, false, rank_ - 1, Number>
-      operator[](const unsigned int row);
+  /**
+   * Access the elements of a row of this symmetric tensor. This function is
+   * called for non-constant tensors.
+   */
+  DEAL_II_HOST
+  constexpr internal::SymmetricTensorAccessors::
+    Accessor<rank_, dim, false, rank_ - 1, Number>
+    operator[](const unsigned int row);
 
-/**
- * Return a @p const reference to the value referred to by the argument.
- *
- * Exactly the same as operator().
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr const Number &
-    operator[](const TableIndices<rank_> &indices) const;
+  /**
+   * Return a @p const reference to the value referred to by the argument.
+   *
+   * Exactly the same as operator().
+   */
+  DEAL_II_HOST
+  constexpr const Number &
+  operator[](const TableIndices<rank_> &indices) const;
 
-/**
- * Return a read-write reference to the indicated element.
- *
- * Exactly the same as operator().
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr Number &
-    operator[](const TableIndices<rank_> &indices);
+  /**
+   * Return a read-write reference to the indicated element.
+   *
+   * Exactly the same as operator().
+   */
+  DEAL_II_HOST
+  constexpr Number &
+  operator[](const TableIndices<rank_> &indices);
 
-/**
- * Access to an element according to unrolled index. The function
- * <tt>s.access_raw_entry(unrolled_index)</tt> does the same as
- * <tt>s[s.unrolled_to_component_indices(unrolled_index)]</tt>, but more
- * efficiently.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr const Number &
-    access_raw_entry(const unsigned int unrolled_index) const;
+  /**
+   * Access to an element according to unrolled index. The function
+   * <tt>s.access_raw_entry(unrolled_index)</tt> does the same as
+   * <tt>s[s.unrolled_to_component_indices(unrolled_index)]</tt>, but more
+   * efficiently.
+   */
+  DEAL_II_HOST
+  constexpr const Number &
+  access_raw_entry(const unsigned int unrolled_index) const;
 
-/**
- * Access to an element according to unrolled index. The function
- * <tt>s.access_raw_entry(unrolled_index)</tt> does the same as
- * <tt>s[s.unrolled_to_component_indices(unrolled_index)]</tt>, but more
- * efficiently.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr Number &
-    access_raw_entry(const unsigned int unrolled_index);
+  /**
+   * Access to an element according to unrolled index. The function
+   * <tt>s.access_raw_entry(unrolled_index)</tt> does the same as
+   * <tt>s[s.unrolled_to_component_indices(unrolled_index)]</tt>, but more
+   * efficiently.
+   */
+  DEAL_II_HOST
+  constexpr Number &
+  access_raw_entry(const unsigned int unrolled_index);
 
-/**
- * Return the Frobenius-norm of a tensor, i.e. the square root of the sum of
- * squares of all entries. This norm is induced by the scalar product
- * defined above for two symmetric tensors. Note that it includes <i>all</i>
- * entries of the tensor, counting symmetry, not only the unique ones (for
- * example, for rank-2 tensors, this norm includes adding up the squares of
- * upper right as well as lower left entries, not just one of them, although
- * they are equal for symmetric tensors).
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr typename numbers::NumberTraits<Number>::real_type
-    norm() const;
+  /**
+   * Return the Frobenius-norm of a tensor, i.e. the square root of the sum of
+   * squares of all entries. This norm is induced by the scalar product
+   * defined above for two symmetric tensors. Note that it includes <i>all</i>
+   * entries of the tensor, counting symmetry, not only the unique ones (for
+   * example, for rank-2 tensors, this norm includes adding up the squares of
+   * upper right as well as lower left entries, not just one of them, although
+   * they are equal for symmetric tensors).
+   */
+  DEAL_II_HOST
+  constexpr typename numbers::NumberTraits<Number>::real_type
+  norm() const;
 
   /**
    * Tensor objects can be unrolled by simply pasting all elements into one
@@ -1173,53 +1076,39 @@ public:
    * <code>[0,n_independent_components)</code> the given entry in a symmetric
    * tensor has.
    */
-  static
-#if defined(__clang__) && defined(__CUDA__)
-    __host__
-#endif
-    constexpr unsigned int
-    component_to_unrolled_index(const TableIndices<rank_> &indices);
+  static DEAL_II_HOST constexpr unsigned int
+  component_to_unrolled_index(const TableIndices<rank_> &indices);
 
   /**
    * The opposite of the previous function: given an index $i$ in the unrolled
    * form of the tensor, return what set of indices $(k,l)$ (for rank-2
    * tensors) or $(k,l,m,n)$ (for rank-4 tensors) corresponds to it.
    */
-  static
-#if defined(__clang__) && defined(__CUDA__)
-    __host__
-#endif
-    constexpr TableIndices<rank_>
-    unrolled_to_component_indices(const unsigned int i);
+  static DEAL_II_HOST constexpr TableIndices<rank_>
+  unrolled_to_component_indices(const unsigned int i);
 
-/**
- * Reset all values to zero.
- *
- * Note that this is partly inconsistent with the semantics of the @p
- * clear() member functions of the standard library containers and of
- * several other classes within deal.II, which not only reset the values of
- * stored elements to zero, but release all memory and return the object
- * into a virginial state. However, since the size of objects of the present
- * type is determined by its template parameters, resizing is not an option,
- * and indeed the state where all elements have a zero value is the state
- * right after construction of such an object.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    constexpr void
-    clear();
+  /**
+   * Reset all values to zero.
+   *
+   * Note that this is partly inconsistent with the semantics of the @p
+   * clear() member functions of the standard library containers and of
+   * several other classes within deal.II, which not only reset the values of
+   * stored elements to zero, but release all memory and return the object
+   * into a virginial state. However, since the size of objects of the present
+   * type is determined by its template parameters, resizing is not an option,
+   * and indeed the state where all elements have a zero value is the state
+   * right after construction of such an object.
+   */
+  DEAL_II_HOST
+  constexpr void
+  clear();
 
   /**
    * Determine an estimate for the memory consumption (in bytes) of this
    * object.
    */
-  static
-#if defined(__clang__) && defined(__CUDA__)
-    __host__
-#endif
-    constexpr std::size_t
-    memory_consumption();
+  static DEAL_II_HOST constexpr std::size_t
+  memory_consumption();
 
   /**
    * Read or write the data of this object to or from a stream for the purpose
@@ -1255,52 +1144,28 @@ private:
 
   // Make a few more functions friends.
   template <int dim2, typename Number2>
-  friend
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-    constexpr Number2
-    trace(const SymmetricTensor<2, dim2, Number2> &d);
+  friend DEAL_II_HOST constexpr Number2
+  trace(const SymmetricTensor<2, dim2, Number2> &d);
 
   template <int dim2, typename Number2>
-  friend
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      DEAL_II_CONSTEXPR Number2
-      determinant(const SymmetricTensor<2, dim2, Number2> &t);
+  friend DEAL_II_HOST DEAL_II_CONSTEXPR Number2
+  determinant(const SymmetricTensor<2, dim2, Number2> &t);
 
   template <int dim2, typename Number2>
-  friend
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-    constexpr SymmetricTensor<2, dim2, Number2>
-    deviator(const SymmetricTensor<2, dim2, Number2> &t);
+  friend DEAL_II_HOST constexpr SymmetricTensor<2, dim2, Number2>
+  deviator(const SymmetricTensor<2, dim2, Number2> &t);
 
   template <int dim2, typename Number2>
-  friend
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      DEAL_II_CONSTEXPR SymmetricTensor<2, dim2, Number2>
-                        unit_symmetric_tensor();
+  friend DEAL_II_HOST DEAL_II_CONSTEXPR SymmetricTensor<2, dim2, Number2>
+                                        unit_symmetric_tensor();
 
   template <int dim2, typename Number2>
-  friend
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      DEAL_II_CONSTEXPR SymmetricTensor<4, dim2, Number2>
-                        deviator_tensor();
+  friend DEAL_II_HOST DEAL_II_CONSTEXPR SymmetricTensor<4, dim2, Number2>
+                                        deviator_tensor();
 
   template <int dim2, typename Number2>
-  friend
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      DEAL_II_CONSTEXPR SymmetricTensor<4, dim2, Number2>
-                        identity_tensor();
+  friend DEAL_II_HOST DEAL_II_CONSTEXPR SymmetricTensor<4, dim2, Number2>
+                                        identity_tensor();
 
 
   // Make a few helper classes friends as well.
@@ -1331,13 +1196,10 @@ namespace internal
   namespace SymmetricTensorAccessors
   {
     template <int rank_, int dim, bool constness, int P, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      constexpr DEAL_II_ALWAYS_INLINE
-      Accessor<rank_, dim, constness, P, Number>::Accessor(
-        tensor_type &              tensor,
-        const TableIndices<rank_> &previous_indices)
+    DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
+    Accessor<rank_, dim, constness, P, Number>::Accessor(
+      tensor_type &              tensor,
+      const TableIndices<rank_> &previous_indices)
       : tensor(tensor)
       , previous_indices(previous_indices)
     {}
@@ -1345,13 +1207,10 @@ namespace internal
 
 
     template <int rank_, int dim, bool constness, int P, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      constexpr inline DEAL_II_ALWAYS_INLINE
-        Accessor<rank_, dim, constness, P - 1, Number>
-        Accessor<rank_, dim, constness, P, Number>::operator[](
-          const unsigned int i)
+    DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+      Accessor<rank_, dim, constness, P - 1, Number>
+      Accessor<rank_, dim, constness, P, Number>::operator[](
+        const unsigned int i)
     {
       return Accessor<rank_, dim, constness, P - 1, Number>(
         tensor, merge(previous_indices, i, rank_ - P));
@@ -1360,13 +1219,10 @@ namespace internal
 
 
     template <int rank_, int dim, bool constness, int P, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      constexpr DEAL_II_ALWAYS_INLINE
-        Accessor<rank_, dim, constness, P - 1, Number>
-        Accessor<rank_, dim, constness, P, Number>::operator[](
-          const unsigned int i) const
+    DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
+      Accessor<rank_, dim, constness, P - 1, Number>
+      Accessor<rank_, dim, constness, P, Number>::operator[](
+        const unsigned int i) const
     {
       return Accessor<rank_, dim, constness, P - 1, Number>(
         tensor, merge(previous_indices, i, rank_ - P));
@@ -1375,13 +1231,10 @@ namespace internal
 
 
     template <int rank_, int dim, bool constness, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      constexpr DEAL_II_ALWAYS_INLINE
-      Accessor<rank_, dim, constness, 1, Number>::Accessor(
-        tensor_type &              tensor,
-        const TableIndices<rank_> &previous_indices)
+    DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
+    Accessor<rank_, dim, constness, 1, Number>::Accessor(
+      tensor_type &              tensor,
+      const TableIndices<rank_> &previous_indices)
       : tensor(tensor)
       , previous_indices(previous_indices)
     {}
@@ -1389,10 +1242,7 @@ namespace internal
 
 
     template <int rank_, int dim, bool constness, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      constexpr inline DEAL_II_ALWAYS_INLINE
+    DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
       typename Accessor<rank_, dim, constness, 1, Number>::reference
       Accessor<rank_, dim, constness, 1, Number>::operator[](
         const unsigned int i)
@@ -1402,10 +1252,7 @@ namespace internal
 
 
     template <int rank_, int dim, bool constness, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-    __host__
-#  endif
-      constexpr DEAL_II_ALWAYS_INLINE
+    DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
       typename Accessor<rank_, dim, constness, 1, Number>::reference
       Accessor<rank_, dim, constness, 1, Number>::operator[](
         const unsigned int i) const
@@ -1441,24 +1288,18 @@ SymmetricTensor<rank_, dim, Number>::SymmetricTensor(
 
 template <int rank_, int dim, typename Number>
 template <typename OtherNumber>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr DEAL_II_ALWAYS_INLINE
-  SymmetricTensor<rank_, dim, Number>::SymmetricTensor(
-    const SymmetricTensor<rank_, dim, OtherNumber> &initializer)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
+SymmetricTensor<rank_, dim, Number>::SymmetricTensor(
+  const SymmetricTensor<rank_, dim, OtherNumber> &initializer)
   : data(initializer.data)
 {}
 
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE
-  SymmetricTensor<rank_, dim, Number>::SymmetricTensor(
-    const Number (&array)[n_independent_components])
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+SymmetricTensor<rank_, dim, Number>::SymmetricTensor(
+  const Number (&array)[n_independent_components])
   : data(
       *reinterpret_cast<const typename base_tensor_type::array_type *>(array))
 {
@@ -1471,10 +1312,8 @@ __host__
 
 template <int rank_, int dim, typename Number>
 template <typename OtherNumber>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number> &
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, Number> &
   SymmetricTensor<rank_, dim, Number>::operator=(
     const SymmetricTensor<rank_, dim, OtherNumber> &t)
 {
@@ -1485,10 +1324,8 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number> &
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, Number> &
   SymmetricTensor<rank_, dim, Number>::operator=(const Number &d)
 {
   Assert(numbers::value_is_zero(d),
@@ -1850,12 +1687,9 @@ namespace internal
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr DEAL_II_ALWAYS_INLINE
-    SymmetricTensor<rank_, dim, Number>::operator Tensor<rank_, dim, Number>()
-      const
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, Number>::operator Tensor<rank_, dim, Number>()
+    const
 {
   return internal::SymmetricTensorImplementation::convert_to_tensor(*this);
 }
@@ -1863,12 +1697,9 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr bool
-  SymmetricTensor<rank_, dim, Number>::operator==(
-    const SymmetricTensor<rank_, dim, Number> &t) const
+DEAL_II_HOST constexpr bool
+SymmetricTensor<rank_, dim, Number>::operator==(
+  const SymmetricTensor<rank_, dim, Number> &t) const
 {
   return data == t.data;
 }
@@ -1876,12 +1707,9 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr bool
-  SymmetricTensor<rank_, dim, Number>::operator!=(
-    const SymmetricTensor<rank_, dim, Number> &t) const
+DEAL_II_HOST constexpr bool
+SymmetricTensor<rank_, dim, Number>::operator!=(
+  const SymmetricTensor<rank_, dim, Number> &t) const
 {
   return data != t.data;
 }
@@ -1890,10 +1718,8 @@ __host__
 
 template <int rank_, int dim, typename Number>
 template <typename OtherNumber>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number> &
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, Number> &
   SymmetricTensor<rank_, dim, Number>::operator+=(
     const SymmetricTensor<rank_, dim, OtherNumber> &t)
 {
@@ -1905,10 +1731,8 @@ __host__
 
 template <int rank_, int dim, typename Number>
 template <typename OtherNumber>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number> &
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, Number> &
   SymmetricTensor<rank_, dim, Number>::operator-=(
     const SymmetricTensor<rank_, dim, OtherNumber> &t)
 {
@@ -1920,10 +1744,8 @@ __host__
 
 template <int rank_, int dim, typename Number>
 template <typename OtherNumber>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number> &
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, Number> &
   SymmetricTensor<rank_, dim, Number>::operator*=(const OtherNumber &d)
 {
   data *= d;
@@ -1934,10 +1756,8 @@ __host__
 
 template <int rank_, int dim, typename Number>
 template <typename OtherNumber>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number> &
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, Number> &
   SymmetricTensor<rank_, dim, Number>::operator/=(const OtherNumber &d)
 {
   data /= d;
@@ -1947,10 +1767,8 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number>
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, Number>
   SymmetricTensor<rank_, dim, Number>::operator-() const
 {
   SymmetricTensor tmp = *this;
@@ -1961,11 +1779,8 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE void
-  SymmetricTensor<rank_, dim, Number>::clear()
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE void
+SymmetricTensor<rank_, dim, Number>::clear()
 {
   data.clear();
 }
@@ -1973,11 +1788,8 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr std::size_t
-  SymmetricTensor<rank_, dim, Number>::memory_consumption()
+DEAL_II_HOST constexpr std::size_t
+SymmetricTensor<rank_, dim, Number>::memory_consumption()
 {
   // all memory consists of statically allocated memory of the current
   // object, no pointers
@@ -1989,10 +1801,7 @@ __host__
 namespace internal
 {
   template <int dim, typename Number, typename OtherNumber = Number>
-#  if defined(__clang__) && defined(__CUDA__)
-  __host__
-#  endif
-    DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+  DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
     typename SymmetricTensorAccessors::
       double_contraction_result<2, 2, dim, Number, OtherNumber>::type
       perform_double_contraction(
@@ -2027,10 +1836,7 @@ namespace internal
 
 
   template <int dim, typename Number, typename OtherNumber = Number>
-#  if defined(__clang__) && defined(__CUDA__)
-  __host__
-#  endif
-    DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+  DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
     typename SymmetricTensorAccessors::
       double_contraction_result<4, 2, dim, Number, OtherNumber>::type
       perform_double_contraction(
@@ -2056,10 +1862,7 @@ namespace internal
 
 
   template <int dim, typename Number, typename OtherNumber = Number>
-#  if defined(__clang__) && defined(__CUDA__)
-  __host__
-#  endif
-    DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+  DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
     typename SymmetricTensorAccessors::StorageType<
       2,
       dim,
@@ -2097,10 +1900,7 @@ namespace internal
 
 
   template <int dim, typename Number, typename OtherNumber = Number>
-#  if defined(__clang__) && defined(__CUDA__)
-  __host__
-#  endif
-    DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+  DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
     typename SymmetricTensorAccessors::StorageType<
       4,
       dim,
@@ -2142,10 +1942,7 @@ namespace internal
 
 template <int rank_, int dim, typename Number>
 template <typename OtherNumber>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
   typename internal::SymmetricTensorAccessors::
     double_contraction_result<rank_, 2, dim, Number, OtherNumber>::type
     SymmetricTensor<rank_, dim, Number>::operator*(
@@ -2163,10 +1960,8 @@ __host__
 
 template <int rank_, int dim, typename Number>
 template <typename OtherNumber>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  DEAL_II_CONSTEXPR inline typename internal::SymmetricTensorAccessors::
+DEAL_II_HOST DEAL_II_CONSTEXPR inline
+  typename internal::SymmetricTensorAccessors::
     double_contraction_result<rank_, 4, dim, Number, OtherNumber>::type
     SymmetricTensor<rank_, dim, Number>::operator*(
       const SymmetricTensor<4, dim, OtherNumber> &s) const
@@ -2385,12 +2180,9 @@ namespace internal
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE Number &
-  SymmetricTensor<rank_, dim, Number>::operator()(
-    const TableIndices<rank_> &indices)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE Number &
+SymmetricTensor<rank_, dim, Number>::operator()(
+  const TableIndices<rank_> &indices)
 {
   for (unsigned int r = 0; r < rank; ++r)
     AssertIndexRange(indices[r], dimension);
@@ -2400,12 +2192,9 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE const Number &
-  SymmetricTensor<rank_, dim, Number>::operator()(
-    const TableIndices<rank_> &indices) const
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE const Number &
+SymmetricTensor<rank_, dim, Number>::operator()(
+  const TableIndices<rank_> &indices) const
 {
   for (unsigned int r = 0; r < rank; ++r)
     AssertIndexRange(indices[r], dimension);
@@ -2442,13 +2231,9 @@ namespace internal
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr DEAL_II_ALWAYS_INLINE internal::SymmetricTensorAccessors::
-    Accessor<rank_, dim, true, rank_ - 1, Number>
-    SymmetricTensor<rank_, dim, Number>::operator[](
-      const unsigned int row) const
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE internal::
+  SymmetricTensorAccessors::Accessor<rank_, dim, true, rank_ - 1, Number>
+  SymmetricTensor<rank_, dim, Number>::operator[](const unsigned int row) const
 {
   return internal::SymmetricTensorAccessors::
     Accessor<rank_, dim, true, rank_ - 1, Number>(
@@ -2460,12 +2245,9 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE internal::SymmetricTensorAccessors::
-    Accessor<rank_, dim, false, rank_ - 1, Number>
-    SymmetricTensor<rank_, dim, Number>::operator[](const unsigned int row)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE internal::
+  SymmetricTensorAccessors::Accessor<rank_, dim, false, rank_ - 1, Number>
+  SymmetricTensor<rank_, dim, Number>::operator[](const unsigned int row)
 {
   return internal::SymmetricTensorAccessors::
     Accessor<rank_, dim, false, rank_ - 1, Number>(
@@ -2477,12 +2259,9 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr DEAL_II_ALWAYS_INLINE const Number &
-  SymmetricTensor<rank_, dim, Number>::operator[](
-    const TableIndices<rank_> &indices) const
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE const Number &
+SymmetricTensor<rank_, dim, Number>::operator[](
+  const TableIndices<rank_> &indices) const
 {
   return operator()(indices);
 }
@@ -2490,12 +2269,9 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline DEAL_II_ALWAYS_INLINE Number &
-  SymmetricTensor<rank_, dim, Number>::operator[](
-    const TableIndices<rank_> &indices)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE Number &
+SymmetricTensor<rank_, dim, Number>::operator[](
+  const TableIndices<rank_> &indices)
 {
   return operator()(indices);
 }
@@ -2566,12 +2342,9 @@ namespace internal
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline const Number &
-  SymmetricTensor<rank_, dim, Number>::access_raw_entry(
-    const unsigned int index) const
+DEAL_II_HOST constexpr inline const Number &
+SymmetricTensor<rank_, dim, Number>::access_raw_entry(
+  const unsigned int index) const
 {
   AssertIndexRange(index, n_independent_components);
   return data[internal::SymmetricTensorImplementation::entry_to_indices(*this,
@@ -2581,12 +2354,8 @@ __host__
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr inline Number &
-  SymmetricTensor<rank_, dim, Number>::access_raw_entry(
-    const unsigned int index)
+DEAL_II_HOST constexpr inline Number &
+SymmetricTensor<rank_, dim, Number>::access_raw_entry(const unsigned int index)
 {
   AssertIndexRange(index, n_independent_components);
   return data[internal::SymmetricTensorImplementation::entry_to_indices(*this,
@@ -2687,11 +2456,8 @@ namespace internal
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr typename numbers::NumberTraits<Number>::real_type
-  SymmetricTensor<rank_, dim, Number>::norm() const
+DEAL_II_HOST constexpr typename numbers::NumberTraits<Number>::real_type
+SymmetricTensor<rank_, dim, Number>::norm() const
 {
   return internal::compute_norm<dim, Number>(data);
 }
@@ -2784,12 +2550,9 @@ namespace internal
 
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr unsigned int
-  SymmetricTensor<rank_, dim, Number>::component_to_unrolled_index(
-    const TableIndices<rank_> &indices)
+DEAL_II_HOST constexpr unsigned int
+SymmetricTensor<rank_, dim, Number>::component_to_unrolled_index(
+  const TableIndices<rank_> &indices)
 {
   return internal::SymmetricTensorImplementation::component_to_unrolled_index<
     dim>(indices);
@@ -2889,12 +2652,9 @@ namespace internal
 } // namespace internal
 
 template <int rank_, int dim, typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  constexpr DEAL_II_ALWAYS_INLINE TableIndices<rank_>
-  SymmetricTensor<rank_, dim, Number>::unrolled_to_component_indices(
-    const unsigned int i)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE TableIndices<rank_>
+SymmetricTensor<rank_, dim, Number>::unrolled_to_component_indices(
+  const unsigned int i)
 {
   return internal::SymmetricTensorImplementation::unrolled_to_component_indices<
     dim>(i, std::integral_constant<int, rank_>());
@@ -2929,13 +2689,10 @@ SymmetricTensor<rank_, dim, Number>::serialize(Archive &ar, const unsigned int)
  * @relatesalso SymmetricTensor
  */
 template <int rank_, int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE
-    SymmetricTensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
-    operator+(const SymmetricTensor<rank_, dim, Number> &     left,
-              const SymmetricTensor<rank_, dim, OtherNumber> &right)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
+  operator+(const SymmetricTensor<rank_, dim, Number> &     left,
+            const SymmetricTensor<rank_, dim, OtherNumber> &right)
 {
   SymmetricTensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
     tmp = left;
@@ -2957,13 +2714,10 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int rank_, int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE
-    SymmetricTensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
-    operator-(const SymmetricTensor<rank_, dim, Number> &     left,
-              const SymmetricTensor<rank_, dim, OtherNumber> &right)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
+  operator-(const SymmetricTensor<rank_, dim, Number> &     left,
+            const SymmetricTensor<rank_, dim, OtherNumber> &right)
 {
   SymmetricTensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
     tmp = left;
@@ -2980,13 +2734,10 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int rank_, int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE
-    Tensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
-    operator+(const SymmetricTensor<rank_, dim, Number> &left,
-              const Tensor<rank_, dim, OtherNumber> &    right)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
+  Tensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
+  operator+(const SymmetricTensor<rank_, dim, Number> &left,
+            const Tensor<rank_, dim, OtherNumber> &    right)
 {
   return Tensor<rank_, dim, Number>(left) + right;
 }
@@ -3000,13 +2751,10 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int rank_, int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE
-    Tensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
-    operator+(const Tensor<rank_, dim, Number> &              left,
-              const SymmetricTensor<rank_, dim, OtherNumber> &right)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
+  Tensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
+  operator+(const Tensor<rank_, dim, Number> &              left,
+            const SymmetricTensor<rank_, dim, OtherNumber> &right)
 {
   return left + Tensor<rank_, dim, OtherNumber>(right);
 }
@@ -3020,13 +2768,10 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int rank_, int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE
-    Tensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
-    operator-(const SymmetricTensor<rank_, dim, Number> &left,
-              const Tensor<rank_, dim, OtherNumber> &    right)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
+  Tensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
+  operator-(const SymmetricTensor<rank_, dim, Number> &left,
+            const Tensor<rank_, dim, OtherNumber> &    right)
 {
   return Tensor<rank_, dim, Number>(left) - right;
 }
@@ -3040,13 +2785,10 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int rank_, int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE
-    Tensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
-    operator-(const Tensor<rank_, dim, Number> &              left,
-              const SymmetricTensor<rank_, dim, OtherNumber> &right)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
+  Tensor<rank_, dim, typename ProductType<Number, OtherNumber>::type>
+  operator-(const Tensor<rank_, dim, Number> &              left,
+            const SymmetricTensor<rank_, dim, OtherNumber> &right)
 {
   return left - Tensor<rank_, dim, OtherNumber>(right);
 }
@@ -3054,11 +2796,8 @@ __host__
 
 
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE Number
-  determinant(const SymmetricTensor<2, dim, Number> &t)
+DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE Number
+determinant(const SymmetricTensor<2, dim, Number> &t)
 {
   switch (dim)
     {
@@ -3097,11 +2836,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR DEAL_II_ALWAYS_INLINE Number
-  third_invariant(const SymmetricTensor<2, dim, Number> &t)
+DEAL_II_HOST DEAL_II_CONSTEXPR DEAL_II_ALWAYS_INLINE Number
+third_invariant(const SymmetricTensor<2, dim, Number> &t)
 {
   return determinant(t);
 }
@@ -3109,11 +2845,8 @@ __host__
 
 
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE Number
-  trace(const SymmetricTensor<2, dim, Number> &d)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE Number
+trace(const SymmetricTensor<2, dim, Number> &d)
 {
   Number t = d.data[0];
   for (unsigned int i = 1; i < dim; ++i)
@@ -3134,11 +2867,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr Number
-  first_invariant(const SymmetricTensor<2, dim, Number> &t)
+DEAL_II_HOST constexpr Number
+first_invariant(const SymmetricTensor<2, dim, Number> &t)
 {
   return trace(t);
 }
@@ -3156,11 +2886,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE Number
-  second_invariant(const SymmetricTensor<2, 1, Number> &)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE Number
+second_invariant(const SymmetricTensor<2, 1, Number> &)
 {
   return internal::NumberType<Number>::value(0.0);
 }
@@ -3186,11 +2913,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE Number
-  second_invariant(const SymmetricTensor<2, 2, Number> &t)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE Number
+second_invariant(const SymmetricTensor<2, 2, Number> &t)
 {
   return t[0][0] * t[1][1] - t[0][1] * t[0][1];
 }
@@ -3206,11 +2930,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE Number
-  second_invariant(const SymmetricTensor<2, 3, Number> &t)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE Number
+second_invariant(const SymmetricTensor<2, 3, Number> &t)
 {
   return (t[0][0] * t[1][1] + t[1][1] * t[2][2] + t[2][2] * t[0][0] -
           t[0][1] * t[0][1] - t[0][2] * t[0][2] - t[1][2] * t[1][2]);
@@ -3614,11 +3335,8 @@ eigenvectors(const SymmetricTensor<2, dim, Number> &T,
  * @relatesalso SymmetricTensor
  */
 template <int rank_, int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number>
-  transpose(const SymmetricTensor<rank_, dim, Number> &t)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number>
+transpose(const SymmetricTensor<rank_, dim, Number> &t)
 {
   return t;
 }
@@ -3626,10 +3344,8 @@ __host__
 
 
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<2, dim, Number>
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<2, dim, Number>
   deviator(const SymmetricTensor<2, dim, Number> &t)
 {
   SymmetricTensor<2, dim, Number> tmp = t;
@@ -3645,11 +3361,9 @@ __host__
 
 
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE SymmetricTensor<2, dim, Number>
-                                                 unit_symmetric_tensor()
+DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<2, dim, Number>
+  unit_symmetric_tensor()
 {
   // create a default constructed matrix filled with
   // zeros, then set the diagonal elements to one
@@ -3676,11 +3390,8 @@ __host__
 
 
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR inline SymmetricTensor<4, dim, Number>
-  deviator_tensor()
+DEAL_II_HOST DEAL_II_CONSTEXPR inline SymmetricTensor<4, dim, Number>
+             deviator_tensor()
 {
   SymmetricTensor<4, dim, Number> tmp;
 
@@ -3707,11 +3418,9 @@ __host__
 
 
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE SymmetricTensor<4, dim, Number>
-                                                 identity_tensor()
+DEAL_II_HOST DEAL_II_CONSTEXPR inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<4, dim, Number>
+  identity_tensor()
 {
   SymmetricTensor<4, dim, Number> tmp;
 
@@ -3745,11 +3454,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<2, dim, Number>
-  invert(const SymmetricTensor<2, dim, Number> &t)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<2, dim, Number>
+invert(const SymmetricTensor<2, dim, Number> &t)
 {
   return internal::SymmetricTensorImplementation::Inverse<2, dim, Number>::
     value(t);
@@ -3768,11 +3474,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr SymmetricTensor<4, dim, Number>
-  invert(const SymmetricTensor<4, dim, Number> &t)
+DEAL_II_HOST constexpr SymmetricTensor<4, dim, Number>
+invert(const SymmetricTensor<4, dim, Number> &t)
 {
   return internal::SymmetricTensorImplementation::Inverse<4, dim, Number>::
     value(t);
@@ -3802,12 +3505,9 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline SymmetricTensor<4, dim, Number>
-  outer_product(const SymmetricTensor<2, dim, Number> &t1,
-                const SymmetricTensor<2, dim, Number> &t2)
+DEAL_II_HOST constexpr inline SymmetricTensor<4, dim, Number>
+outer_product(const SymmetricTensor<2, dim, Number> &t1,
+              const SymmetricTensor<2, dim, Number> &t2)
 {
   SymmetricTensor<4, dim, Number> tmp;
 
@@ -3831,10 +3531,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<2, dim, Number>
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<2, dim, Number>
   symmetrize(const Tensor<2, dim, Number> &t)
 {
   SymmetricTensor<2, dim, Number> result;
@@ -3858,10 +3556,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int rank_, int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number>
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
+  SymmetricTensor<rank_, dim, Number>
   operator*(const SymmetricTensor<rank_, dim, Number> &t, const Number &factor)
 {
   SymmetricTensor<rank_, dim, Number> tt = t;
@@ -3879,11 +3575,8 @@ __host__
  * @relatesalso SymmetricTensor
  */
 template <int rank_, int dim, typename Number>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number>
-  operator*(const Number &factor, const SymmetricTensor<rank_, dim, Number> &t)
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim, Number>
+operator*(const Number &factor, const SymmetricTensor<rank_, dim, Number> &t)
 {
   // simply forward to the other operator
   return t * factor;
@@ -3916,16 +3609,13 @@ __host__
  * @relates SymmetricTensor
  */
 template <int rank_, int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<
-    rank_,
-    dim,
-    typename ProductType<Number,
-                         typename EnableIfScalar<OtherNumber>::type>::type>
-  operator*(const SymmetricTensor<rank_, dim, Number> &t,
-            const OtherNumber &                        factor)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<
+  rank_,
+  dim,
+  typename ProductType<Number,
+                       typename EnableIfScalar<OtherNumber>::type>::type>
+operator*(const SymmetricTensor<rank_, dim, Number> &t,
+          const OtherNumber &                        factor)
 {
   // form the product. we have to convert the two factors into the final
   // type via explicit casts because, for awkward reasons, the C++
@@ -3948,16 +3638,13 @@ __host__
  * @relates SymmetricTensor
  */
 template <int rank_, int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<
-    rank_,
-    dim,
-    typename ProductType<OtherNumber,
-                         typename EnableIfScalar<Number>::type>::type>
-  operator*(const Number &                                  factor,
-            const SymmetricTensor<rank_, dim, OtherNumber> &t)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<
+  rank_,
+  dim,
+  typename ProductType<OtherNumber,
+                       typename EnableIfScalar<Number>::type>::type>
+operator*(const Number &                                  factor,
+          const SymmetricTensor<rank_, dim, OtherNumber> &t)
 {
   // simply forward to the other operator with switched arguments
   return (t * factor);
@@ -3971,16 +3658,13 @@ __host__
  * @relates SymmetricTensor
  */
 template <int rank_, int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline SymmetricTensor<
-    rank_,
-    dim,
-    typename ProductType<Number,
-                         typename EnableIfScalar<OtherNumber>::type>::type>
-  operator/(const SymmetricTensor<rank_, dim, Number> &t,
-            const OtherNumber &                        factor)
+DEAL_II_HOST constexpr inline SymmetricTensor<
+  rank_,
+  dim,
+  typename ProductType<Number,
+                       typename EnableIfScalar<OtherNumber>::type>::type>
+operator/(const SymmetricTensor<rank_, dim, Number> &t,
+          const OtherNumber &                        factor)
 {
   using product_type = typename ProductType<Number, OtherNumber>::type;
   SymmetricTensor<rank_, dim, product_type> tt(t);
@@ -3997,11 +3681,8 @@ __host__
  * @relates SymmetricTensor
  */
 template <int rank_, int dim>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim>
-  operator*(const SymmetricTensor<rank_, dim> &t, const double factor)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim>
+operator*(const SymmetricTensor<rank_, dim> &t, const double factor)
 {
   SymmetricTensor<rank_, dim> tt(t);
   tt *= factor;
@@ -4017,11 +3698,8 @@ __host__
  * @relates SymmetricTensor
  */
 template <int rank_, int dim>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim>
-  operator*(const double factor, const SymmetricTensor<rank_, dim> &t)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE SymmetricTensor<rank_, dim>
+operator*(const double factor, const SymmetricTensor<rank_, dim> &t)
 {
   SymmetricTensor<rank_, dim> tt(t);
   tt *= factor;
@@ -4036,11 +3714,8 @@ __host__
  * @relates SymmetricTensor
  */
 template <int rank_, int dim>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline SymmetricTensor<rank_, dim>
-  operator/(const SymmetricTensor<rank_, dim> &t, const double factor)
+DEAL_II_HOST constexpr inline SymmetricTensor<rank_, dim>
+operator/(const SymmetricTensor<rank_, dim> &t, const double factor)
 {
   SymmetricTensor<rank_, dim> tt(t);
   tt /= factor;
@@ -4057,10 +3732,7 @@ __host__
  * @relates SymmetricTensor
  */
 template <int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
   typename ProductType<Number, OtherNumber>::type
   scalar_product(const SymmetricTensor<2, dim, Number> &     t1,
                  const SymmetricTensor<2, dim, OtherNumber> &t2)
@@ -4082,10 +3754,7 @@ __host__
  * @relates SymmetricTensor
  */
 template <int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE
   typename ProductType<Number, OtherNumber>::type
   scalar_product(const SymmetricTensor<2, dim, Number> &t1,
                  const Tensor<2, dim, OtherNumber> &    t2)
@@ -4112,10 +3781,7 @@ __host__
  * @relates SymmetricTensor
  */
 template <int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
   typename ProductType<Number, OtherNumber>::type
   scalar_product(const Tensor<2, dim, Number> &              t1,
                  const SymmetricTensor<2, dim, OtherNumber> &t2)
@@ -4139,14 +3805,11 @@ __host__
  * @relates SymmetricTensor
  */
 template <typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline DEAL_II_ALWAYS_INLINE void
-  double_contract(
-    SymmetricTensor<2, 1, typename ProductType<Number, OtherNumber>::type> &tmp,
-    const SymmetricTensor<4, 1, Number> &                                   t,
-    const SymmetricTensor<2, 1, OtherNumber> &                              s)
+DEAL_II_HOST constexpr inline DEAL_II_ALWAYS_INLINE void
+double_contract(
+  SymmetricTensor<2, 1, typename ProductType<Number, OtherNumber>::type> &tmp,
+  const SymmetricTensor<4, 1, Number> &                                   t,
+  const SymmetricTensor<2, 1, OtherNumber> &                              s)
 {
   tmp[0][0] = t[0][0][0][0] * s[0][0];
 }
@@ -4168,14 +3831,11 @@ __host__
  * @relates SymmetricTensor
  */
 template <typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline void
-  double_contract(
-    SymmetricTensor<2, 1, typename ProductType<Number, OtherNumber>::type> &tmp,
-    const SymmetricTensor<2, 1, Number> &                                   s,
-    const SymmetricTensor<4, 1, OtherNumber> &                              t)
+DEAL_II_HOST constexpr inline void
+double_contract(
+  SymmetricTensor<2, 1, typename ProductType<Number, OtherNumber>::type> &tmp,
+  const SymmetricTensor<2, 1, Number> &                                   s,
+  const SymmetricTensor<4, 1, OtherNumber> &                              t)
 {
   tmp[0][0] = t[0][0][0][0] * s[0][0];
 }
@@ -4197,14 +3857,11 @@ __host__
  * @relates SymmetricTensor
  */
 template <typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline void
-  double_contract(
-    SymmetricTensor<2, 2, typename ProductType<Number, OtherNumber>::type> &tmp,
-    const SymmetricTensor<4, 2, Number> &                                   t,
-    const SymmetricTensor<2, 2, OtherNumber> &                              s)
+DEAL_II_HOST constexpr inline void
+double_contract(
+  SymmetricTensor<2, 2, typename ProductType<Number, OtherNumber>::type> &tmp,
+  const SymmetricTensor<4, 2, Number> &                                   t,
+  const SymmetricTensor<2, 2, OtherNumber> &                              s)
 {
   const unsigned int dim = 2;
 
@@ -4231,14 +3888,11 @@ __host__
  * @relates SymmetricTensor
  */
 template <typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline void
-  double_contract(
-    SymmetricTensor<2, 2, typename ProductType<Number, OtherNumber>::type> &tmp,
-    const SymmetricTensor<2, 2, Number> &                                   s,
-    const SymmetricTensor<4, 2, OtherNumber> &                              t)
+DEAL_II_HOST constexpr inline void
+double_contract(
+  SymmetricTensor<2, 2, typename ProductType<Number, OtherNumber>::type> &tmp,
+  const SymmetricTensor<2, 2, Number> &                                   s,
+  const SymmetricTensor<4, 2, OtherNumber> &                              t)
 {
   const unsigned int dim = 2;
 
@@ -4265,14 +3919,11 @@ __host__
  * @relates SymmetricTensor
  */
 template <typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline void
-  double_contract(
-    SymmetricTensor<2, 3, typename ProductType<Number, OtherNumber>::type> &tmp,
-    const SymmetricTensor<4, 3, Number> &                                   t,
-    const SymmetricTensor<2, 3, OtherNumber> &                              s)
+DEAL_II_HOST constexpr inline void
+double_contract(
+  SymmetricTensor<2, 3, typename ProductType<Number, OtherNumber>::type> &tmp,
+  const SymmetricTensor<4, 3, Number> &                                   t,
+  const SymmetricTensor<2, 3, OtherNumber> &                              s)
 {
   const unsigned int dim = 3;
 
@@ -4300,14 +3951,11 @@ __host__
  * @relates SymmetricTensor
  */
 template <typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr inline void
-  double_contract(
-    SymmetricTensor<2, 3, typename ProductType<Number, OtherNumber>::type> &tmp,
-    const SymmetricTensor<2, 3, Number> &                                   s,
-    const SymmetricTensor<4, 3, OtherNumber> &                              t)
+DEAL_II_HOST constexpr inline void
+double_contract(
+  SymmetricTensor<2, 3, typename ProductType<Number, OtherNumber>::type> &tmp,
+  const SymmetricTensor<2, 3, Number> &                                   s,
+  const SymmetricTensor<4, 3, OtherNumber> &                              t)
 {
   const unsigned int dim = 3;
 
@@ -4327,12 +3975,11 @@ __host__
  * @relates SymmetricTensor
  */
 template <int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr Tensor<1, dim, typename ProductType<Number, OtherNumber>::type>
-  operator*(const SymmetricTensor<2, dim, Number> &src1,
-            const Tensor<1, dim, OtherNumber> &    src2)
+DEAL_II_HOST constexpr Tensor<1,
+                              dim,
+                              typename ProductType<Number, OtherNumber>::type>
+operator*(const SymmetricTensor<2, dim, Number> &src1,
+          const Tensor<1, dim, OtherNumber> &    src2)
 {
   Tensor<1, dim, typename ProductType<Number, OtherNumber>::type> dest;
   for (unsigned int i = 0; i < dim; ++i)
@@ -4349,12 +3996,11 @@ __host__
  * @relates SymmetricTensor
  */
 template <int dim, typename Number, typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr Tensor<1, dim, typename ProductType<Number, OtherNumber>::type>
-  operator*(const Tensor<1, dim, Number> &              src1,
-            const SymmetricTensor<2, dim, OtherNumber> &src2)
+DEAL_II_HOST constexpr Tensor<1,
+                              dim,
+                              typename ProductType<Number, OtherNumber>::type>
+operator*(const Tensor<1, dim, Number> &              src1,
+          const SymmetricTensor<2, dim, OtherNumber> &src2)
 {
   // this is easy for symmetric tensors:
   return src2 * src1;
@@ -4386,10 +4032,7 @@ template <int rank_1,
           int dim,
           typename Number,
           typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
   typename Tensor<rank_1 + rank_2 - 2,
                   dim,
                   typename ProductType<Number, OtherNumber>::type>::tensor_type
@@ -4425,10 +4068,7 @@ template <int rank_1,
           int dim,
           typename Number,
           typename OtherNumber>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  constexpr DEAL_II_ALWAYS_INLINE
+DEAL_II_HOST constexpr DEAL_II_ALWAYS_INLINE
   typename Tensor<rank_1 + rank_2 - 2,
                   dim,
                   typename ProductType<Number, OtherNumber>::type>::tensor_type
diff --git a/include/deal.II/base/table_indices.h b/include/deal.II/base/table_indices.h
index 50d640d9bb..bd5e3cae68 100644
--- a/include/deal.II/base/table_indices.h
+++ b/include/deal.II/base/table_indices.h
@@ -89,15 +89,13 @@ public:
   constexpr bool
   operator!=(const TableIndices<N> &other) const;
 
-/**
- * Sort the indices in ascending order. While this operation is not very
- * useful for Table objects, it is used for the SymmetricTensor class.
- */
-#if defined(__clang__) && defined(__CUDA__)
-  __host__
-#endif
-    DEAL_II_CONSTEXPR void
-    sort();
+  /**
+   * Sort the indices in ascending order. While this operation is not very
+   * useful for Table objects, it is used for the SymmetricTensor class.
+   */
+  DEAL_II_HOST
+  DEAL_II_CONSTEXPR void
+  sort();
 
   /**
    * Write or read the data of this object to or from a stream for the purpose
@@ -168,11 +166,8 @@ TableIndices<N>::operator!=(const TableIndices<N> &other) const
 
 
 template <int N>
-#if defined(__clang__) && defined(__CUDA__)
-__host__
-#endif
-  DEAL_II_CONSTEXPR inline void
-  TableIndices<N>::sort()
+DEAL_II_HOST DEAL_II_CONSTEXPR inline void
+TableIndices<N>::sort()
 {
   std::sort(std::begin(indices), std::end(indices));
 }
diff --git a/include/deal.II/physics/elasticity/standard_tensors.h b/include/deal.II/physics/elasticity/standard_tensors.h
index 069a1ad743..c2cdc62f4f 100644
--- a/include/deal.II/physics/elasticity/standard_tensors.h
+++ b/include/deal.II/physics/elasticity/standard_tensors.h
@@ -231,12 +231,8 @@ namespace Physics
        * @dealiiHolzapfelA{229,6.83}
        */
       template <typename Number>
-      static
-#if defined(__clang__) && defined(__CUDA__)
-        __host__
-#endif
-          DEAL_II_CONSTEXPR SymmetricTensor<4, dim, Number>
-                            Dev_P(const Tensor<2, dim, Number> &F);
+      static DEAL_II_HOST DEAL_II_CONSTEXPR SymmetricTensor<4, dim, Number>
+      Dev_P(const Tensor<2, dim, Number> &F);
 
       /**
        * Return the transpose of the fourth-order referential deviatoric tensor,
@@ -250,12 +246,8 @@ namespace Physics
        * @f]
        */
       template <typename Number>
-      static
-#if defined(__clang__) && defined(__CUDA__)
-        __host__
-#endif
-          DEAL_II_CONSTEXPR SymmetricTensor<4, dim, Number>
-                            Dev_P_T(const Tensor<2, dim, Number> &F);
+      static DEAL_II_HOST DEAL_II_CONSTEXPR SymmetricTensor<4, dim, Number>
+      Dev_P_T(const Tensor<2, dim, Number> &F);
 
       /** @} */
 
@@ -281,12 +273,8 @@ namespace Physics
        * @dealiiHolzapfelA{228,6.82}
        */
       template <typename Number>
-      static
-#if defined(__clang__) && defined(__CUDA__)
-        __host__
-#endif
-          DEAL_II_CONSTEXPR SymmetricTensor<2, dim, Number>
-                            ddet_F_dC(const Tensor<2, dim, Number> &F);
+      static DEAL_II_HOST DEAL_II_CONSTEXPR SymmetricTensor<2, dim, Number>
+      ddet_F_dC(const Tensor<2, dim, Number> &F);
 
       /** @} */
 
@@ -310,12 +298,8 @@ namespace Physics
        * @dealiiWriggersA{76,3.255}
        */
       template <typename Number>
-      static
-#if defined(__clang__) && defined(__CUDA__)
-        __host__
-#endif
-          DEAL_II_CONSTEXPR SymmetricTensor<4, dim, Number>
-                            dC_inv_dC(const Tensor<2, dim, Number> &F);
+      static DEAL_II_HOST DEAL_II_CONSTEXPR SymmetricTensor<4, dim, Number>
+      dC_inv_dC(const Tensor<2, dim, Number> &F);
 
       /** @} */
     };
@@ -332,12 +316,9 @@ namespace Physics
 
 template <int dim>
 template <typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  DEAL_II_CONSTEXPR inline SymmetricTensor<4, dim, Number>
-  Physics::Elasticity::StandardTensors<dim>::Dev_P(
-    const Tensor<2, dim, Number> &F)
+DEAL_II_HOST DEAL_II_CONSTEXPR inline SymmetricTensor<4, dim, Number>
+             Physics::Elasticity::StandardTensors<dim>::Dev_P(
+  const Tensor<2, dim, Number> &F)
 {
   // Make things work with AD types
   using std::pow;
@@ -362,12 +343,9 @@ __host__
 
 template <int dim>
 template <typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  DEAL_II_CONSTEXPR inline SymmetricTensor<4, dim, Number>
-  Physics::Elasticity::StandardTensors<dim>::Dev_P_T(
-    const Tensor<2, dim, Number> &F)
+DEAL_II_HOST DEAL_II_CONSTEXPR inline SymmetricTensor<4, dim, Number>
+             Physics::Elasticity::StandardTensors<dim>::Dev_P_T(
+  const Tensor<2, dim, Number> &F)
 {
   // Make things work with AD types
   using std::pow;
@@ -393,12 +371,9 @@ __host__
 
 template <int dim>
 template <typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  DEAL_II_CONSTEXPR SymmetricTensor<2, dim, Number>
-                    Physics::Elasticity::StandardTensors<dim>::ddet_F_dC(
-    const Tensor<2, dim, Number> &F)
+DEAL_II_HOST DEAL_II_CONSTEXPR SymmetricTensor<2, dim, Number>
+Physics::Elasticity::StandardTensors<dim>::ddet_F_dC(
+  const Tensor<2, dim, Number> &F)
 {
   return internal::NumberType<Number>::value(0.5 * determinant(F)) *
          symmetrize(invert(transpose(F) * F));
@@ -408,12 +383,9 @@ __host__
 
 template <int dim>
 template <typename Number>
-#  if defined(__clang__) && defined(__CUDA__)
-__host__
-#  endif
-  DEAL_II_CONSTEXPR inline SymmetricTensor<4, dim, Number>
-  Physics::Elasticity::StandardTensors<dim>::dC_inv_dC(
-    const Tensor<2, dim, Number> &F)
+DEAL_II_HOST DEAL_II_CONSTEXPR inline SymmetricTensor<4, dim, Number>
+             Physics::Elasticity::StandardTensors<dim>::dC_inv_dC(
+  const Tensor<2, dim, Number> &F)
 {
   const SymmetricTensor<2, dim, Number> C_inv =
     symmetrize(invert(transpose(F) * F));