Add type related files

new file:   include/native_types.hpp
new file:   include/native_types_impl.hpp
new file:   include/type_traits.hpp
new file:   include/types_ext.hpp
new file:   include/xfloat32.hpp

library/include/hiptensor/internal/hiptensor_utility.hpp

@@ -131,11 +131,6 @@ namespace std
 
         return os;
     }
-    static ostream& operator<<(ostream& os, const _Float16 value)
-    {
-        os << static_cast<float>(value);
-        return os;
-    }
 }
 
 #endif // HIPTENSOR_UTILITY_INTERNAL_HPP

library/src/CMakeLists.txt

@@ -62,7 +62,7 @@ set(HIPTENSOR_CORE_SOURCES
    ${CMAKE_CURRENT_SOURCE_DIR}/hiptensor.cpp
    ${CMAKE_CURRENT_SOURCE_DIR}/logger.cpp
    ${CMAKE_CURRENT_SOURCE_DIR}/performance.cpp
-   ${CMAKE_CURRENT_SOURCE_DIR}/types.cpp
+   ${CMAKE_CURRENT_SOURCE_DIR}/data_types.cpp
    ${CMAKE_CURRENT_SOURCE_DIR}/hip_device.cpp
    ${CMAKE_CURRENT_SOURCE_DIR}/handle.cpp
 )

library/src/contraction/contraction_meta_traits.hpp

@@ -34,8 +34,8 @@
 #include <element_wise_operation.hpp>
 
 // hiptensor includes
+#include "data_types.hpp"
 #include "meta_traits.hpp"
-#include "types.hpp"
 
 namespace hiptensor
 {

library/src/contraction/contraction_solution_params.hpp

@@ -30,7 +30,7 @@
 #include <hiptensor/hiptensor_types.hpp>
 
 #include "contraction_types.hpp"
-#include "types.hpp"
+#include "data_types.hpp"
 
 namespace hiptensor
 {

library/src/contraction/contraction_solution_params_impl.hpp

@@ -29,8 +29,8 @@
 
 #include "contraction_meta_traits.hpp"
 #include "contraction_solution_params.hpp"
+#include "data_types.hpp"
 #include "hash.hpp"
-#include "types.hpp"
 
 namespace std
 {

library/src/contraction/contraction_solution_registry.hpp

@@ -32,8 +32,8 @@
 #include <vector>
 
 #include "contraction_types.hpp"
+#include "data_types.hpp"
 #include "singleton.hpp"
-#include "types.hpp"
 
 namespace hiptensor
 {

library/src/types.cpp → library/src/data_types.cpp

@@ -24,7 +24,7 @@
  *
  *******************************************************************************/
 
-#include "types.hpp"
+#include "data_types.hpp"
 
 namespace hiptensor
 {

library/src/hiptensor.cpp

@@ -27,9 +27,9 @@
 
 #include <hiptensor/hiptensor.hpp>
 
+#include "data_types.hpp"
 #include "handle.hpp"
 #include "logger.hpp"
-#include "types.hpp"
 #include "util.hpp"
 
 hiptensorStatus_t hiptensorCreate(hiptensorHandle_t** handle)

library/src/include/config.hpp

@@ -0,0 +1,104 @@
+/*******************************************************************************
+ *
+ * MIT License
+ *
+ * Copyright (C) 2021-2024 Advanced Micro Devices, Inc. All rights reserved.
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to deal
+ * in the Software without restriction, including without limitation the rights
+ * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+ * copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in
+ * all copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ *
+ *******************************************************************************/
+#ifndef HIPTENSOR_CONFIG_HPP
+#define HIPTENSOR_CONFIG_HPP
+
+namespace hiptensor
+{
+
+///
+/// Architecture support
+/// Guaranteed symbols:
+/// HIPTENSOR_ARCH_GFX908
+/// HIPTENSOR_ARCH_GFX90a
+/// HIPTENSOR_ARCH_GFX940
+/// HIPTENSOR_ARCH_GFX941
+/// HIPTENSOR_ARCH_GFX942
+#if defined(__gfx908__)
+#define HIPTENSOR_ARCH_GFX908 __gfx908__
+#elif defined(__gfx90a__)
+#define HIPTENSOR_ARCH_GFX90A __gfx90a__
+#elif defined(__gfx940__)
+#define HIPTENSOR_ARCH_GFX940 __gfx940__
+#elif defined(__gfx941__)
+#define HIPTENSOR_ARCH_GFX941 __gfx941__
+#elif defined(__gfx942__)
+#define HIPTENSOR_ARCH_GFX942 __gfx942__
+#else
+#define HIPTENSOR_ARCH_HOST 1
+#endif
+
+#if !defined(HIPTENSOR_ARCH_GFX908)
+#define HIPTENSOR_ARCH_GFX908 0
+#endif
+#if !defined(HIPTENSOR_ARCH_GFX90A)
+#define HIPTENSOR_ARCH_GFX90A 0
+#endif
+#if !defined(HIPTENSOR_ARCH_GFX940)
+#define HIPTENSOR_ARCH_GFX940 0
+#endif
+#if !defined(HIPTENSOR_ARCH_GFX941)
+#define HIPTENSOR_ARCH_GFX941 0
+#endif
+#if !defined(HIPTENSOR_ARCH_GFX942)
+#define HIPTENSOR_ARCH_GFX942 0
+#endif
+#if !defined(HIPTENSOR_ARCH_HOST)
+#define HIPTENSOR_ARCH_HOST 0
+#endif
+
+#if defined(NDEBUG)
+#define HIPTENSOR_UNSUPPORTED_IMPL(MSG)
+#else
+#define HIPTENSOR_UNSUPPORTED_IMPL(MSG) __attribute__((deprecated(MSG)))
+#endif
+
+#if defined(HIP_NO_HALF)
+#define HIPTENSOR_NO_HALF 1
+#else
+#define HIPTENSOR_NO_HALF 0
+#endif // HIP_NO_HALF
+
+#if HIPTENSOR_NO_HALF || (!HIPTENSOR_NO_HALF && defined(__HIP_NO_HALF_CONVERSIONS__))
+#define HIPTENSOR_TESTS_NO_HALF 1
+#else
+#define HIPTENSOR_TESTS_NO_HALF 0
+#endif // !HIPTENSOR_NO_HALF && defined(__HIP_NO_HALF_CONVERSIONS__)
+
+///
+/// Host and Device symbols
+///
+#define HIPTENSOR_DEVICE __device__
+
+#define HIPTENSOR_HOST __host__
+
+#define HIPTENSOR_HOST_DEVICE HIPTENSOR_HOST HIPTENSOR_DEVICE
+
+#define HIPTENSOR_KERNEL __global__
+
+} // namespace hiptensor
+
+#endif // HIPTENSOR_CONFIG_HPP

library/src/include/data_types.hpp

@@ -0,0 +1,78 @@
+/*******************************************************************************
+ *
+ * MIT License
+ *
+ * Copyright (C) 2023-2024 Advanced Micro Devices, Inc. All rights reserved.
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to deal
+ * in the Software without restriction, including without limitation the rights
+ * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+ * copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in
+ * all copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
+ * THE SOFTWARE.
+ *
+ *******************************************************************************/
+
+#ifndef HIPTENSOR_LIBRARY_DATA_TYPES_HPP
+#define HIPTENSOR_LIBRARY_DATA_TYPES_HPP
+
+// clang-format off
+// Include order needs to be preserved
+#include <hip/library_types.h>
+#include <hip/hip_bfloat16.h>
+#include <hip/hip_fp16.h>
+#include <iostream>
+
+#include <hiptensor/hiptensor_types.hpp>
+
+// clang-format on
+
+namespace hiptensor
+{
+    // Used to map to empty tensors
+    struct NoneType;
+
+    static constexpr hipDataType NONE_TYPE = (hipDataType)31;
+
+    // Map type to runtime HipDataType
+    template <typename T>
+    struct HipDataType;
+
+    template <typename T>
+    static constexpr auto HipDataType_v = HipDataType<T>::value;
+
+    // Get data size in bytes from id
+    uint32_t hipDataTypeSize(hipDataType id);
+
+    // Convert hipDataType to hiptensorComputeType_t
+    hiptensorComputeType_t convertToComputeType(hipDataType hipType);
+
+    // Read a single value from void pointer, casted to T
+    template <typename T>
+    T readVal(void const* value, hipDataType id);
+
+    template <typename T>
+    T readVal(void const* value, hiptensorComputeType_t id);
+
+} // namespace hiptensor
+
+bool operator==(hipDataType hipType, hiptensorComputeType_t computeType);
+bool operator==(hiptensorComputeType_t computeType, hipDataType hipType);
+
+bool operator!=(hipDataType hipType, hiptensorComputeType_t computeType);
+bool operator!=(hiptensorComputeType_t computeType, hipDataType hipType);
+
+#include "data_types_impl.hpp"
+
+#endif // HIPTENSOR_LIBRARY_DATA_TYPES_HPP

library/src/include/types_impl.hpp → library/src/include/data_types_impl.hpp

@@ -24,10 +24,10 @@
  *
  *******************************************************************************/
 
-#ifndef HIPTENSOR_LIBRARY_TYPES_IMPL_HPP
-#define HIPTENSOR_LIBRARY_TYPES_IMPL_HPP
+#ifndef HIPTENSOR_LIBRARY_DATA_TYPES_IMPL_HPP
+#define HIPTENSOR_LIBRARY_DATA_TYPES_IMPL_HPP
 
-#include "types.hpp"
+#include "data_types.hpp"
 
 namespace hiptensor
 {
@@ -217,4 +217,4 @@ namespace hiptensor
 
 } // namespace hiptensor
 
-#endif // HIPTENSOR_LIBRARY_TYPES_IMPL_HPP
+#endif // HIPTENSOR_LIBRARY_DATA_TYPES_IMPL_HPP

library/src/include/native_types.hpp

@@ -0,0 +1,102 @@
+/*******************************************************************************
+ *
+ * MIT License
+ *
+ * Copyright (C) 2021-2024 Advanced Micro Devices, Inc. All rights reserved.
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to deal
+ * in the Software without restriction, including without limitation the rights
+ * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+ * copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in
+ * all copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ *
+ *******************************************************************************/
+#ifndef HIPTENSOR_NATIVE_TYPES_HPP
+#define HIPTENSOR_NATIVE_TYPES_HPP
+
+#include <array>
+#include <hip/hip_bfloat16.h>
+#include <hip/hip_fp16.h>
+#include <hip/hip_vector_types.h>
+#include <type_traits>
+#include <utility>
+
+#include "xfloat32.hpp"
+
+namespace hiptensor
+{
+
+    /**
+ * \defgroup DataTypes Data Type Metadata
+ *
+ * @brief Definition and metadata on supported data types of matrices.
+ *
+ * @{
+ *
+ * Native Data Types:
+ * float64_t = f64 = double
+ * float = f32
+ * _Float16 = f16
+ * int8
+ * uint8
+ * int16
+ * int32
+ * uint32
+ *
+ *
+ * Non-Native Data Types:
+ * h16 = __half
+ * bf16 = bfloat16
+ *
+ */
+
+    // Native types
+    using float16_t = _Float16;
+    using float32_t = float;
+    using float64_t = double;
+
+    using int8_t    = ::int8_t;
+    using uint8_t   = ::uint8_t;
+    using int16_t   = ::int16_t;
+    using uint16_t  = ::uint16_t;
+    using int32_t   = ::int32_t;
+    using uint32_t  = ::uint32_t;
+    using int64_t   = ::int64_t;
+    using uint64_t  = ::uint64_t;
+    using index_t   = ::int32_t;
+    using index64_t = ::int64_t;
+
+    // Non-native types
+    using bfloat16_t = hip_bfloat16;
+
+#if !HIPTENSOR_NO_HALF
+    using hfloat16_t = __half;
+#endif // !HIPTENSOR_NO_HALF
+
+    using xfloat32_t = hiptensor_xfloat32;
+
+    // clang-format off
+
+
+} // namespace hiptensor
+
+// Add in some extensions to basic type support.
+// Some of these are required for vector implementations.
+// #include "type_traits.hpp"
+// #include "types_ext.hpp"
+
+#include "native_types_impl.hpp"
+
+#endif // HIPTENSOR_NATIVE_TYPES_HPP