cuda : fix HIP build

Author: ggerganov

ggml-common.h

@@ -12,8 +12,8 @@ typedef uint32_t ggml_half2;
 #elif defined(GGML_COMMON_DECL_METAL)
 #include <metal_stdlib>
 
-typedef half   ggml_half;
-typedef half2  ggml_half2;
+typedef half  ggml_half;
+typedef half2 ggml_half2;
 
 #define GGML_COMMON_AGGR
 
@@ -22,8 +22,18 @@ typedef half2  ggml_half2;
 #include <cuda_fp16.h>
 #include <cstdint>
 
-typedef half   ggml_half;
-typedef half2  ggml_half2;
+typedef half  ggml_half;
+typedef half2 ggml_half2;
+
+#define GGML_COMMON_AGGR data
+
+#define GGML_COMMON_DECL
+#elif defined(GGML_COMMON_DECL_HIP)
+#include <hip/hip_fp16.h>
+#include <cstdint>
+
+typedef half  ggml_half;
+typedef half2 ggml_half2;
 
 #define GGML_COMMON_AGGR data
 
@@ -373,7 +383,7 @@ static_assert(sizeof(block_iq4_xs) == sizeof(ggml_half) + sizeof(uint16_t) + QK_
 #define GGML_TABLE_END() };
 
 #define GGML_COMMON_IMPL
-#elif defined(GGML_COMMON_IMPL_CUDA)
+#elif defined(GGML_COMMON_IMPL_CUDA) || defined(GGML_COMMON_IMPL_HIP)
 #include <cstdint>
 
 #define GGML_TABLE_BEGIN(type, name, size) static const __device__ __constant__ type name[size] = {

ggml-cuda.cu

@@ -2,8 +2,13 @@
 #include "ggml.h"
 #include "ggml-backend-impl.h"
 
+#if defined(GGML_USE_HIPBLAS)
+#define GGML_COMMON_DECL_HIP
+#define GGML_COMMON_IMPL_HIP
+#else
 #define GGML_COMMON_DECL_CUDA
 #define GGML_COMMON_IMPL_CUDA
+#endif
 #include "ggml-common.h"
 
 #include <algorithm>