[flang][runtime] Enable more APIs in the offload build. #75996
Merged
Conversation
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This patch enables more numeric (mod, sum, matmul, etc.) APIs, and some others. I added new macros to disable warnings about using C++ STD methods like operators of std::complex, which do not have __device__ attribute. This may probably result in unresolved references, if the header files implementation relies on libstdc++. I will need to follow up on this.
llvmbot
added
flang:runtime
flang
|
@llvm/pr-subscribers-flang-runtime Author: Slava Zakharin (vzakhari) ChangesThis patch enables more numeric (mod, sum, matmul, etc.) APIs, I added new macros to disable warnings about using C++ STD methods Patch is 158.51 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/75996.diff 19 Files Affected:
diff --git a/flang/include/flang/ISO_Fortran_binding.h b/flang/include/flang/ISO_Fortran_binding.h
index dd384c516263e5..4a28d3322a38f0 100644
--- a/flang/include/flang/ISO_Fortran_binding.h
+++ b/flang/include/flang/ISO_Fortran_binding.h
@@ -189,8 +189,8 @@ RT_API_ATTRS void *CFI_address(
RT_API_ATTRS int CFI_allocate(CFI_cdesc_t *, const CFI_index_t lower_bounds[],
const CFI_index_t upper_bounds[], size_t elem_len);
RT_API_ATTRS int CFI_deallocate(CFI_cdesc_t *);
-int CFI_establish(CFI_cdesc_t *, void *base_addr, CFI_attribute_t, CFI_type_t,
- size_t elem_len, CFI_rank_t, const CFI_index_t extents[]);
+RT_API_ATTRS int CFI_establish(CFI_cdesc_t *, void *base_addr, CFI_attribute_t,
+ CFI_type_t, size_t elem_len, CFI_rank_t, const CFI_index_t extents[]);
RT_API_ATTRS int CFI_is_contiguous(const CFI_cdesc_t *);
RT_API_ATTRS int CFI_section(CFI_cdesc_t *, const CFI_cdesc_t *source,
const CFI_index_t lower_bounds[], const CFI_index_t upper_bounds[],
diff --git a/flang/include/flang/Runtime/allocatable.h b/flang/include/flang/Runtime/allocatable.h
index 4169483398f6a7..58061d9862095e 100644
--- a/flang/include/flang/Runtime/allocatable.h
+++ b/flang/include/flang/Runtime/allocatable.h
@@ -26,22 +26,22 @@ extern "C" {
// A descriptor must be initialized before being used for any purpose,
// but needs reinitialization in a deallocated state only when there is
// a change of type, rank, or corank.
-void RTNAME(AllocatableInitIntrinsic)(
+void RTDECL(AllocatableInitIntrinsic)(
Descriptor &, TypeCategory, int kind, int rank = 0, int corank = 0);
-void RTNAME(AllocatableInitCharacter)(Descriptor &, SubscriptValue length = 0,
+void RTDECL(AllocatableInitCharacter)(Descriptor &, SubscriptValue length = 0,
int kind = 1, int rank = 0, int corank = 0);
-void RTNAME(AllocatableInitDerived)(
+void RTDECL(AllocatableInitDerived)(
Descriptor &, const typeInfo::DerivedType &, int rank = 0, int corank = 0);
// Initializes the descriptor for an allocatable of intrinsic or derived type.
// These functions are meant to be used in the allocate statement lowering. If
// the descriptor is allocated, the initialization is skiped so the error
// handling can be done by AllocatableAllocate.
-void RTNAME(AllocatableInitIntrinsicForAllocate)(
+void RTDECL(AllocatableInitIntrinsicForAllocate)(
Descriptor &, TypeCategory, int kind, int rank = 0, int corank = 0);
-void RTNAME(AllocatableInitCharacterForAllocate)(Descriptor &,
+void RTDECL(AllocatableInitCharacterForAllocate)(Descriptor &,
SubscriptValue length = 0, int kind = 1, int rank = 0, int corank = 0);
-void RTNAME(AllocatableInitDerivedForAllocate)(
+void RTDECL(AllocatableInitDerivedForAllocate)(
Descriptor &, const typeInfo::DerivedType &, int rank = 0, int corank = 0);
// Checks that an allocatable is not already allocated in statements
@@ -50,29 +50,29 @@ void RTNAME(AllocatableInitDerivedForAllocate)(
// (If there's no STAT=, the error will be caught later anyway, but
// this API allows the error to be caught before descriptor is modified.)
// Return 0 on success (deallocated state), else the STAT= value.
-int RTNAME(AllocatableCheckAllocated)(Descriptor &,
+int RTDECL(AllocatableCheckAllocated)(Descriptor &,
const Descriptor *errMsg = nullptr, const char *sourceFile = nullptr,
int sourceLine = 0);
// For MOLD= allocation; sets bounds, cobounds, and length type
// parameters from another descriptor. The destination descriptor must
// be initialized and deallocated.
-void RTNAME(AllocatableApplyMold)(
+void RTDECL(AllocatableApplyMold)(
Descriptor &, const Descriptor &mold, int rank = 0);
// Explicitly sets the bounds and length type parameters of an initialized
// deallocated allocatable.
-void RTNAME(AllocatableSetBounds)(
+void RTDECL(AllocatableSetBounds)(
Descriptor &, int zeroBasedDim, SubscriptValue lower, SubscriptValue upper);
// The upper cobound is ignored for the last codimension.
-void RTNAME(AllocatableSetCoBounds)(Descriptor &, int zeroBasedCoDim,
+void RTDECL(AllocatableSetCoBounds)(Descriptor &, int zeroBasedCoDim,
SubscriptValue lower, SubscriptValue upper = 0);
// Length type parameters are indexed in declaration order; i.e., 0 is the
// first length type parameter in the deepest base type. (Not for use
// with CHARACTER; see above.)
-void RTNAME(AllocatableSetDerivedLength)(
+void RTDECL(AllocatableSetDerivedLength)(
Descriptor &, int which, SubscriptValue);
// When an explicit type-spec appears in an ALLOCATE statement for an
@@ -80,7 +80,7 @@ void RTNAME(AllocatableSetDerivedLength)(
// a derived type or CHARACTER value, the explicit value has to match
// the length type parameter's value. This API checks that requirement.
// Returns 0 for success, or the STAT= value on failure with hasStat==true.
-int RTNAME(AllocatableCheckLengthParameter)(Descriptor &,
+int RTDECL(AllocatableCheckLengthParameter)(Descriptor &,
int which /* 0 for CHARACTER length */, SubscriptValue other,
bool hasStat = false, const Descriptor *errMsg = nullptr,
const char *sourceFile = nullptr, int sourceLine = 0);
@@ -94,10 +94,10 @@ int RTNAME(AllocatableCheckLengthParameter)(Descriptor &,
// Successfully allocated memory is initialized if the allocatable has a
// derived type, and is always initialized by AllocatableAllocateSource().
// Performs all necessary coarray synchronization and validation actions.
-int RTNAME(AllocatableAllocate)(Descriptor &, bool hasStat = false,
+int RTDECL(AllocatableAllocate)(Descriptor &, bool hasStat = false,
const Descriptor *errMsg = nullptr, const char *sourceFile = nullptr,
int sourceLine = 0);
-int RTNAME(AllocatableAllocateSource)(Descriptor &, const Descriptor &source,
+int RTDECL(AllocatableAllocateSource)(Descriptor &, const Descriptor &source,
bool hasStat = false, const Descriptor *errMsg = nullptr,
const char *sourceFile = nullptr, int sourceLine = 0);
@@ -105,7 +105,7 @@ int RTNAME(AllocatableAllocateSource)(Descriptor &, const Descriptor &source,
// but note the order of first two arguments is reversed for consistency
// with the other APIs for allocatables.) The destination descriptor
// must be initialized.
-std::int32_t RTNAME(MoveAlloc)(Descriptor &to, Descriptor &from,
+std::int32_t RTDECL(MoveAlloc)(Descriptor &to, Descriptor &from,
const typeInfo::DerivedType *, bool hasStat = false,
const Descriptor *errMsg = nullptr, const char *sourceFile = nullptr,
int sourceLine = 0);
@@ -113,19 +113,19 @@ std::int32_t RTNAME(MoveAlloc)(Descriptor &to, Descriptor &from,
// Deallocates an allocatable. Finalizes elements &/or components as needed.
// The allocatable is left in an initialized state suitable for reallocation
// with the same bounds, cobounds, and length type parameters.
-int RTNAME(AllocatableDeallocate)(Descriptor &, bool hasStat = false,
+int RTDECL(AllocatableDeallocate)(Descriptor &, bool hasStat = false,
const Descriptor *errMsg = nullptr, const char *sourceFile = nullptr,
int sourceLine = 0);
// Same as AllocatableDeallocate but also set the dynamic type as the declared
// type as mentioned in 7.3.2.3 note 7.
-int RTNAME(AllocatableDeallocatePolymorphic)(Descriptor &,
+int RTDECL(AllocatableDeallocatePolymorphic)(Descriptor &,
const typeInfo::DerivedType *, bool hasStat = false,
const Descriptor *errMsg = nullptr, const char *sourceFile = nullptr,
int sourceLine = 0);
// Variant of above that does not finalize; for intermediate results
-void RTNAME(AllocatableDeallocateNoFinal)(
+void RTDECL(AllocatableDeallocateNoFinal)(
Descriptor &, const char *sourceFile = nullptr, int sourceLine = 0);
} // extern "C"
} // namespace Fortran::runtime
diff --git a/flang/include/flang/Runtime/api-attrs.h b/flang/include/flang/Runtime/api-attrs.h
index 61da2c06d3a4dc..9c8a67ffc34a82 100644
--- a/flang/include/flang/Runtime/api-attrs.h
+++ b/flang/include/flang/Runtime/api-attrs.h
@@ -121,4 +121,15 @@
#undef RT_DEVICE_COMPILATION
#endif
+#if defined(__CUDACC__)
+#define RT_DIAG_PUSH _Pragma("nv_diagnostic push")
+#define RT_DIAG_POP _Pragma("nv_diagnostic pop")
+#define RT_DIAG_DISABLE_CALL_HOST_FROM_DEVICE_WARN \
+ _Pragma("nv_diag_suppress 20011") _Pragma("nv_diag_suppress 20014")
+#else /* !defined(__CUDACC__) */
+#define RT_DIAG_PUSH
+#define RT_DIAG_POP
+#define RT_DIAG_DISABLE_CALL_HOST_FROM_DEVICE_WARN
+#endif /* !defined(__CUDACC__) */
+
#endif /* !FORTRAN_RUNTIME_API_ATTRS_H_ */
diff --git a/flang/include/flang/Runtime/derived-api.h b/flang/include/flang/Runtime/derived-api.h
index decba9f686d920..79aa7d82de8819 100644
--- a/flang/include/flang/Runtime/derived-api.h
+++ b/flang/include/flang/Runtime/derived-api.h
@@ -29,37 +29,37 @@ extern "C" {
// Initializes and allocates an object's components, if it has a derived type
// with any default component initialization or automatic components.
// The descriptor must be initialized and non-null.
-void RTNAME(Initialize)(
+void RTDECL(Initialize)(
const Descriptor &, const char *sourceFile = nullptr, int sourceLine = 0);
// Finalizes an object and its components. Deallocates any
// allocatable/automatic components. Does not deallocate the descriptor's
// storage.
-void RTNAME(Destroy)(const Descriptor &);
+void RTDECL(Destroy)(const Descriptor &);
// Finalizes the object and its components.
-void RTNAME(Finalize)(
+void RTDECL(Finalize)(
const Descriptor &, const char *sourceFile = nullptr, int sourceLine = 0);
/// Deallocates any allocatable/automatic components.
/// Does not deallocate the descriptor's storage.
/// Does not perform any finalization.
-void RTNAME(DestroyWithoutFinalization)(const Descriptor &);
+void RTDECL(DestroyWithoutFinalization)(const Descriptor &);
// Intrinsic or defined assignment, with scalar expansion but not type
// conversion.
-void RTNAME(Assign)(const Descriptor &, const Descriptor &,
+void RTDECL(Assign)(const Descriptor &, const Descriptor &,
const char *sourceFile = nullptr, int sourceLine = 0);
// Perform the test of the CLASS IS type guard statement of the SELECT TYPE
// construct.
-bool RTNAME(ClassIs)(const Descriptor &, const typeInfo::DerivedType &);
+bool RTDECL(ClassIs)(const Descriptor &, const typeInfo::DerivedType &);
// Perform the test of the SAME_TYPE_AS intrinsic.
-bool RTNAME(SameTypeAs)(const Descriptor &, const Descriptor &);
+bool RTDECL(SameTypeAs)(const Descriptor &, const Descriptor &);
// Perform the test of the EXTENDS_TYPE_OF intrinsic.
-bool RTNAME(ExtendsTypeOf)(const Descriptor &, const Descriptor &);
+bool RTDECL(ExtendsTypeOf)(const Descriptor &, const Descriptor &);
} // extern "C"
} // namespace Fortran::runtime
diff --git a/flang/include/flang/Runtime/matmul-transpose.h b/flang/include/flang/Runtime/matmul-transpose.h
index 7cfb189863df82..5eb5896972e0fd 100644
--- a/flang/include/flang/Runtime/matmul-transpose.h
+++ b/flang/include/flang/Runtime/matmul-transpose.h
@@ -18,12 +18,12 @@ extern "C" {
// The most general MATMUL(TRANSPOSE()). All type and shape information is
// taken from the arguments' descriptors, and the result is dynamically
// allocated.
-void RTNAME(MatmulTranspose)(Descriptor &, const Descriptor &,
+void RTDECL(MatmulTranspose)(Descriptor &, const Descriptor &,
const Descriptor &, const char *sourceFile = nullptr, int line = 0);
// A non-allocating variant; the result's descriptor must be established
// and have a valid base address.
-void RTNAME(MatmulTransposeDirect)(const Descriptor &, const Descriptor &,
+void RTDECL(MatmulTransposeDirect)(const Descriptor &, const Descriptor &,
const Descriptor &, const char *sourceFile = nullptr, int line = 0);
} // extern "C"
} // namespace Fortran::runtime
diff --git a/flang/include/flang/Runtime/matmul.h b/flang/include/flang/Runtime/matmul.h
index 4598c487a12ca1..40581d44de9e2c 100644
--- a/flang/include/flang/Runtime/matmul.h
+++ b/flang/include/flang/Runtime/matmul.h
@@ -17,12 +17,12 @@ extern "C" {
// The most general MATMUL. All type and shape information is taken from the
// arguments' descriptors, and the result is dynamically allocated.
-void RTNAME(Matmul)(Descriptor &, const Descriptor &, const Descriptor &,
+void RTDECL(Matmul)(Descriptor &, const Descriptor &, const Descriptor &,
const char *sourceFile = nullptr, int line = 0);
// A non-allocating variant; the result's descriptor must be established
// and have a valid base address.
-void RTNAME(MatmulDirect)(const Descriptor &, const Descriptor &,
+void RTDECL(MatmulDirect)(const Descriptor &, const Descriptor &,
const Descriptor &, const char *sourceFile = nullptr, int line = 0);
} // extern "C"
} // namespace Fortran::runtime
diff --git a/flang/include/flang/Runtime/numeric.h b/flang/include/flang/Runtime/numeric.h
index e4e11a61731a61..3d9cb8b5b0acdc 100644
--- a/flang/include/flang/Runtime/numeric.h
+++ b/flang/include/flang/Runtime/numeric.h
@@ -20,280 +20,280 @@ namespace Fortran::runtime {
extern "C" {
// CEILING
-CppTypeFor<TypeCategory::Integer, 1> RTNAME(Ceiling4_1)(
+CppTypeFor<TypeCategory::Integer, 1> RTDECL(Ceiling4_1)(
CppTypeFor<TypeCategory::Real, 4>);
-CppTypeFor<TypeCategory::Integer, 2> RTNAME(Ceiling4_2)(
+CppTypeFor<TypeCategory::Integer, 2> RTDECL(Ceiling4_2)(
CppTypeFor<TypeCategory::Real, 4>);
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Ceiling4_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Ceiling4_4)(
CppTypeFor<TypeCategory::Real, 4>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Ceiling4_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Ceiling4_8)(
CppTypeFor<TypeCategory::Real, 4>);
#ifdef __SIZEOF_INT128__
-CppTypeFor<TypeCategory::Integer, 16> RTNAME(Ceiling4_16)(
+CppTypeFor<TypeCategory::Integer, 16> RTDECL(Ceiling4_16)(
CppTypeFor<TypeCategory::Real, 4>);
#endif
-CppTypeFor<TypeCategory::Integer, 1> RTNAME(Ceiling8_1)(
+CppTypeFor<TypeCategory::Integer, 1> RTDECL(Ceiling8_1)(
CppTypeFor<TypeCategory::Real, 8>);
-CppTypeFor<TypeCategory::Integer, 2> RTNAME(Ceiling8_2)(
+CppTypeFor<TypeCategory::Integer, 2> RTDECL(Ceiling8_2)(
CppTypeFor<TypeCategory::Real, 8>);
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Ceiling8_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Ceiling8_4)(
CppTypeFor<TypeCategory::Real, 8>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Ceiling8_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Ceiling8_8)(
CppTypeFor<TypeCategory::Real, 8>);
#ifdef __SIZEOF_INT128__
-CppTypeFor<TypeCategory::Integer, 16> RTNAME(Ceiling8_16)(
+CppTypeFor<TypeCategory::Integer, 16> RTDECL(Ceiling8_16)(
CppTypeFor<TypeCategory::Real, 8>);
#endif
#if LDBL_MANT_DIG == 64
-CppTypeFor<TypeCategory::Integer, 1> RTNAME(Ceiling10_1)(
+CppTypeFor<TypeCategory::Integer, 1> RTDECL(Ceiling10_1)(
CppTypeFor<TypeCategory::Real, 10>);
-CppTypeFor<TypeCategory::Integer, 2> RTNAME(Ceiling10_2)(
+CppTypeFor<TypeCategory::Integer, 2> RTDECL(Ceiling10_2)(
CppTypeFor<TypeCategory::Real, 10>);
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Ceiling10_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Ceiling10_4)(
CppTypeFor<TypeCategory::Real, 10>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Ceiling10_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Ceiling10_8)(
CppTypeFor<TypeCategory::Real, 10>);
#ifdef __SIZEOF_INT128__
-CppTypeFor<TypeCategory::Integer, 16> RTNAME(Ceiling10_16)(
+CppTypeFor<TypeCategory::Integer, 16> RTDECL(Ceiling10_16)(
CppTypeFor<TypeCategory::Real, 10>);
#endif
#endif
#if LDBL_MANT_DIG == 113 || HAS_FLOAT128
-CppTypeFor<TypeCategory::Integer, 1> RTNAME(Ceiling16_1)(
+CppTypeFor<TypeCategory::Integer, 1> RTDECL(Ceiling16_1)(
CppTypeFor<TypeCategory::Real, 16>);
-CppTypeFor<TypeCategory::Integer, 2> RTNAME(Ceiling16_2)(
+CppTypeFor<TypeCategory::Integer, 2> RTDECL(Ceiling16_2)(
CppTypeFor<TypeCategory::Real, 16>);
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Ceiling16_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Ceiling16_4)(
CppTypeFor<TypeCategory::Real, 16>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Ceiling16_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Ceiling16_8)(
CppTypeFor<TypeCategory::Real, 16>);
#ifdef __SIZEOF_INT128__
-CppTypeFor<TypeCategory::Integer, 16> RTNAME(Ceiling16_16)(
+CppTypeFor<TypeCategory::Integer, 16> RTDECL(Ceiling16_16)(
CppTypeFor<TypeCategory::Real, 16>);
#endif
#endif
// EXPONENT is defined to return default INTEGER; support INTEGER(4 & 8)
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Exponent4_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Exponent4_4)(
CppTypeFor<TypeCategory::Real, 4>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Exponent4_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Exponent4_8)(
CppTypeFor<TypeCategory::Real, 4>);
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Exponent8_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Exponent8_4)(
CppTypeFor<TypeCategory::Real, 8>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Exponent8_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Exponent8_8)(
CppTypeFor<TypeCategory::Real, 8>);
#if LDBL_MANT_DIG == 64
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Exponent10_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Exponent10_4)(
CppTypeFor<TypeCategory::Real, 10>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Exponent10_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Exponent10_8)(
CppTypeFor<TypeCategory::Real, 10>);
#endif
#if LDBL_MANT_DIG == 113 || HAS_FLOAT
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Exponent16_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Exponent16_4)(
CppTypeFor<TypeCategory::Real, 16>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Exponent16_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Exponent16_8)(
CppTypeFor<TypeCategory::Real, 16>);
#endif
// FLOOR
-CppTypeFor<TypeCategory::Integer, 1> RTNAME(Floor4_1)(
+CppTypeFor<TypeCategory::Integer, 1> RTDECL(Floor4_1)(
CppTypeFor<TypeCategory::Real, 4>);
-CppTypeFor<TypeCategory::Integer, 2> RTNAME(Floor4_2)(
+CppTypeFor<TypeCategory::Integer, 2> RTDECL(Floor4_2)(
CppTypeFor<TypeCategory::Real, 4>);
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Floor4_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Floor4_4)(
CppTypeFor<TypeCategory::Real, 4>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Floor4_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Floor4_8)(
CppTypeFor<TypeCategory::Real, 4>);
#ifdef __SIZEOF_INT128__
-CppTypeFor<TypeCategory::Integer, 16> RTNAME(Floor4_16)(
+CppTypeFor<TypeCategory::Integer, 16> RTDECL(Floor4_16)(
CppTypeFor<TypeCategory::Real, 4>);
#endif
-CppTypeFor<TypeCategory::Integer, 1> RTNAME(Floor8_1)(
+CppTypeFor<TypeCategory::Integer, 1> RTDECL(Floor8_1)(
CppTypeFor<TypeCategory::Real, 8>);
-CppTypeFor<TypeCategory::Integer, 2> RTNAME(Floor8_2)(
+CppTypeFor<TypeCategory::Integer, 2> RTDECL(Floor8_2)(
CppTypeFor<TypeCategory::Real, 8>);
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Floor8_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Floor8_4)(
CppTypeFor<TypeCategory::Real, 8>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Floor8_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Floor8_8)(
CppTypeFor<TypeCategory::Real, 8>);
#ifdef __SIZEOF_INT128__
-CppTypeFor<TypeCategory::Integer, 16> RTNAME(Floor8_16)(
+CppTypeFor<TypeCategory::Integer, 16> RTDECL(Floor8_16)(
CppTypeFor<TypeCategory::Real, 8>);
#endif
#if LDBL_MANT_DIG == 64
-CppTypeFor<TypeCategory::Integer, 1> RTNAME(Floor10_1)(
+CppTypeFor<TypeCategory::Integer, 1> RTDECL(Floor10_1)(
CppTypeFor<TypeCategory::Real, 10>);
-CppTypeFor<TypeCategory::Integer, 2> RTNAME(Floor10_2)(
+CppTypeFor<TypeCategory::Integer, 2> RTDECL(Floor10_2)(
CppTypeFor<TypeCategory::Real, 10>);
-CppTypeFor<TypeCategory::Integer, 4> RTNAME(Floor10_4)(
+CppTypeFor<TypeCategory::Integer, 4> RTDECL(Floor10_4)(
CppTypeFor<TypeCategory::Real, 10>);
-CppTypeFor<TypeCategory::Integer, 8> RTNAME(Floor10_8)(
+CppTypeFor<TypeCategory::Integer, 8> RTDECL(Floor10_8)(
CppTypeFor<TypeCategory::Real, 10>);
#ifdef __SIZEOF_INT128__
-CppTypeFor<TypeCategory::Integer, 16> RTNAME(Floor10_16)(
+CppTypeFor<TypeCategory::Integer, 16> RTDECL(Floor10_16)(
CppTypeFor<TypeCategory::Real, 10>);
#endif
#endif
#if LDBL_MANT_DIG == 113 || HAS_FLOAT128
-CppTypeFor<TypeCategory...
[truncated]
|
klausler
approved these changes
Dec 20, 2023
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
This patch enables more numeric (mod, sum, matmul, etc.) APIs,
and some others.
I added new macros to disable warnings about using C++ STD methods
like operators of std::complex, which do not have device attribute.
This may probably result in unresolved references, if the header files
implementation relies on libstdc++. I will need to follow up on this.