From b5d286a6d3dca489777c505fa0a187a890110d6c Mon Sep 17 00:00:00 2001 From: Satyanvesh Dittakavi Date: Tue, 7 Nov 2023 09:15:32 +0000 Subject: [PATCH] SWDEV-430963 - Address MIGraphX failures due to unsafe buffer access usage Compiler seem to be stricter in compiler staging builds related to safe buffer programming when compared to other component staging builds. This seem to result in additional errors when -Werror is enabled in MIGraphX tests. Removes all the clang pragmas to ignore several type of warnings in all the headers and adds a single pragma which ignores all warnings using #pragma clang diagnostic ignored "-Weverything" in hiprtc builtins. Change-Id: I95f302bb285b2451b19dd5dfdb7df29164b0f750 --- .../hip/amd_detail/amd_channel_descriptor.h | 9 ------- .../hip/amd_detail/amd_device_functions.h | 27 ------------------- .../include/hip/amd_detail/amd_hip_atomic.h | 12 --------- .../include/hip/amd_detail/amd_hip_complex.h | 16 ----------- .../amd_detail/amd_hip_cooperative_groups.h | 13 --------- hipamd/include/hip/amd_detail/amd_hip_fp16.h | 18 ------------- .../hip/amd_detail/amd_hip_unsafe_atomics.h | 9 ------- .../hip/amd_detail/amd_math_functions.h | 10 ------- .../hip/amd_detail/amd_surface_functions.h | 17 ------------ .../hip/amd_detail/amd_warp_functions.h | 15 ----------- .../hip/amd_detail/device_library_decls.h | 11 -------- .../hip_cooperative_groups_helper.h | 11 -------- hipamd/include/hip/amd_detail/math_fwd.h | 10 ------- hipamd/include/hip/amd_detail/ockl_image.h | 9 ------- .../hip/amd_detail/texture_fetch_functions.h | 16 ----------- .../amd_detail/texture_indirect_functions.h | 15 ----------- hipamd/src/hiprtc/cmake/HIPRTC.cmake | 10 ++++--- 17 files changed, 7 insertions(+), 221 deletions(-) diff --git a/hipamd/include/hip/amd_detail/amd_channel_descriptor.h b/hipamd/include/hip/amd_detail/amd_channel_descriptor.h index f5ba75ebe6..c6b150d4b3 100644 --- a/hipamd/include/hip/amd_detail/amd_channel_descriptor.h +++ b/hipamd/include/hip/amd_detail/amd_channel_descriptor.h @@ -23,11 +23,6 @@ THE SOFTWARE. #ifndef HIP_INCLUDE_HIP_AMD_DETAIL_CHANNEL_DESCRIPTOR_H #define HIP_INCLUDE_HIP_AMD_DETAIL_CHANNEL_DESCRIPTOR_H -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wold-style-cast" -#endif - #if !defined(__HIPCC_RTC__) #include #include @@ -360,8 +355,4 @@ struct hipChannelFormatDesc hipCreateChannelDesc(int x, int y, int z, int w, #endif /* __cplusplus */ -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif /* !HIP_INCLUDE_HIP_AMD_DETAIL_CHANNEL_DESCRIPTOR_H */ diff --git a/hipamd/include/hip/amd_detail/amd_device_functions.h b/hipamd/include/hip/amd_detail/amd_device_functions.h index 5c2945a1f2..19f39e16ed 100644 --- a/hipamd/include/hip/amd_detail/amd_device_functions.h +++ b/hipamd/include/hip/amd_detail/amd_device_functions.h @@ -23,22 +23,6 @@ THE SOFTWARE. #ifndef HIP_INCLUDE_HIP_AMD_DETAIL_DEVICE_FUNCTIONS_H #define HIP_INCLUDE_HIP_AMD_DETAIL_DEVICE_FUNCTIONS_H -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wreserved-macro-identifier" -#pragma clang diagnostic ignored "-Wreserved-identifier" -#pragma clang diagnostic ignored "-Wsign-conversion" -#pragma clang diagnostic ignored "-Wc++98-compat-pedantic" -#pragma clang diagnostic ignored "-Wold-style-cast" -#pragma clang diagnostic ignored "-Wshorten-64-to-32" -#pragma clang diagnostic ignored "-Wimplicit-int-conversion" -#pragma clang diagnostic ignored "-Wimplicit-float-conversion" -#pragma clang diagnostic ignored "-Wmissing-noreturn" -#pragma clang diagnostic ignored "-Wimplicit-fallthrough" -#pragma clang diagnostic ignored "-Wunneeded-internal-declaration" -#pragma clang diagnostic ignored "-Wshift-count-overflow" -#endif - #if !defined(__HIPCC_RTC__) #include #include "host_defines.h" @@ -459,10 +443,6 @@ __device__ static inline unsigned long long int __double2ull_ru(double x) { __device__ static inline unsigned long long int __double2ull_rz(double x) { return (unsigned long long int)x; } -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wc++98-compat-pedantic" -#endif __device__ static inline long long int __double_as_longlong(double x) { static_assert(sizeof(long long) == sizeof(double), ""); @@ -471,9 +451,6 @@ __device__ static inline long long int __double_as_longlong(double x) { return tmp; } -#if defined(__clang__) -#pragma clang diagnostic pop -#endif /* __device__ unsigned short __float2half_rn(float x); @@ -1121,8 +1098,4 @@ static inline __device__ void* memset(void* ptr, int val, size_t size) { } #endif // !__OPENMP_AMDGCN__ -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif diff --git a/hipamd/include/hip/amd_detail/amd_hip_atomic.h b/hipamd/include/hip/amd_detail/amd_hip_atomic.h index 3f55831f77..c02a57b079 100644 --- a/hipamd/include/hip/amd_detail/amd_hip_atomic.h +++ b/hipamd/include/hip/amd_detail/amd_hip_atomic.h @@ -22,14 +22,6 @@ THE SOFTWARE. #pragma once -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wc++98-compat" -#pragma clang diagnostic ignored "-Wold-style-cast" -#pragma clang diagnostic ignored "-Wc++98-compat-pedantic" -#pragma clang diagnostic ignored "-Wsign-conversion" -#endif - #if !defined(__HIPCC_RTC__) #include "amd_device_functions.h" #endif @@ -1635,8 +1627,4 @@ unsigned long long atomicXor( return __atomic_fetch_xor(address, val, __ATOMIC_RELAXED); } -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif // __hip_atomic_compare_exchange_strong diff --git a/hipamd/include/hip/amd_detail/amd_hip_complex.h b/hipamd/include/hip/amd_detail/amd_hip_complex.h index 7dc014b6fd..9ca29eafd2 100644 --- a/hipamd/include/hip/amd_detail/amd_hip_complex.h +++ b/hipamd/include/hip/amd_detail/amd_hip_complex.h @@ -25,18 +25,6 @@ THE SOFTWARE. #ifndef HIP_INCLUDE_HIP_AMD_DETAIL_HIP_COMPLEX_H #define HIP_INCLUDE_HIP_AMD_DETAIL_HIP_COMPLEX_H -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wfloat-equal" -#pragma clang diagnostic ignored "-Wc++98-compat" -#pragma clang diagnostic ignored "-Wold-style-cast" -#pragma clang diagnostic ignored "-Wreserved-macro-identifier" -#pragma clang diagnostic ignored "-Wimplicit-int-float-conversion" -#pragma clang diagnostic ignored "-Wimplicit-float-conversion" -#pragma clang diagnostic ignored "-Wdouble-promotion" -#pragma clang diagnostic ignored "-Wc++98-compat-pedantic" -#endif - #if !defined(__HIPCC_RTC__) #include "hip/amd_detail/amd_hip_vector_types.h" #endif @@ -183,8 +171,4 @@ __HOST_DEVICE__ static inline hipDoubleComplex hipCfma(hipDoubleComplex p, hipDo return make_hipDoubleComplex(real, imag); } -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif //HIP_INCLUDE_HIP_AMD_DETAIL_HIP_COMPLEX_H diff --git a/hipamd/include/hip/amd_detail/amd_hip_cooperative_groups.h b/hipamd/include/hip/amd_detail/amd_hip_cooperative_groups.h index 04ecca70d2..eeb67bd075 100644 --- a/hipamd/include/hip/amd_detail/amd_hip_cooperative_groups.h +++ b/hipamd/include/hip/amd_detail/amd_hip_cooperative_groups.h @@ -32,15 +32,6 @@ THE SOFTWARE. #ifndef HIP_INCLUDE_HIP_AMD_DETAIL_HIP_COOPERATIVE_GROUPS_H #define HIP_INCLUDE_HIP_AMD_DETAIL_HIP_COOPERATIVE_GROUPS_H -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wc++98-compat" -#pragma clang diagnostic ignored "-Wsign-conversion" -#pragma clang diagnostic ignored "-Wunused-parameter" -#pragma clang diagnostic ignored "-Wreserved-macro-identifier" -#pragma clang diagnostic ignored "-Wpadded" -#endif - #if __cplusplus #if !defined(__HIPCC_RTC__) #include @@ -842,9 +833,5 @@ __CG_QUALIFIER__ thread_block_tile tiled_partition(const Paren } } // namespace cooperative_groups -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif // __cplusplus #endif // HIP_INCLUDE_HIP_AMD_DETAIL_HIP_COOPERATIVE_GROUPS_H diff --git a/hipamd/include/hip/amd_detail/amd_hip_fp16.h b/hipamd/include/hip/amd_detail/amd_hip_fp16.h index 5124afbe39..5ff2662570 100644 --- a/hipamd/include/hip/amd_detail/amd_hip_fp16.h +++ b/hipamd/include/hip/amd_detail/amd_hip_fp16.h @@ -24,20 +24,6 @@ THE SOFTWARE. #ifndef HIP_INCLUDE_HIP_AMD_DETAIL_HIP_FP16_H #define HIP_INCLUDE_HIP_AMD_DETAIL_HIP_FP16_H -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wreserved-identifier" -#pragma clang diagnostic ignored "-Wreserved-macro-identifier" -#pragma clang diagnostic ignored "-Wc++98-compat" -#pragma clang diagnostic ignored "-Wc++98-compat-pedantic" -#pragma clang diagnostic ignored "-Wsign-conversion" -#pragma clang diagnostic ignored "-Wfloat-conversion" -#pragma clang diagnostic ignored "-Wdouble-promotion" -#pragma clang diagnostic ignored "-Wnested-anon-types" -#pragma clang diagnostic ignored "-Wgnu-anonymous-struct" -#pragma clang diagnostic ignored "-Wfloat-equal" -#endif - #if defined(__HIPCC_RTC__) #define __HOST_DEVICE__ __device__ #else @@ -1820,8 +1806,4 @@ THE SOFTWARE. #endif #endif // !defined(__clang__) && defined(__GNUC__) -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif // HIP_INCLUDE_HIP_AMD_DETAIL_HIP_FP16_H diff --git a/hipamd/include/hip/amd_detail/amd_hip_unsafe_atomics.h b/hipamd/include/hip/amd_detail/amd_hip_unsafe_atomics.h index 836e60e9df..59841ab9b3 100644 --- a/hipamd/include/hip/amd_detail/amd_hip_unsafe_atomics.h +++ b/hipamd/include/hip/amd_detail/amd_hip_unsafe_atomics.h @@ -24,11 +24,6 @@ THE SOFTWARE. #ifdef __cplusplus -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wold-style-cast" -#endif - /** * @brief Unsafe floating point rmw atomic add. * @@ -567,8 +562,4 @@ __device__ inline double safeAtomicMin(double* addr, double val) { #endif } -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif diff --git a/hipamd/include/hip/amd_detail/amd_math_functions.h b/hipamd/include/hip/amd_detail/amd_math_functions.h index f2e6b9a512..fca4d7e408 100644 --- a/hipamd/include/hip/amd_detail/amd_math_functions.h +++ b/hipamd/include/hip/amd_detail/amd_math_functions.h @@ -22,12 +22,6 @@ THE SOFTWARE. #pragma once -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wundef" -#pragma clang diagnostic ignored "-Wreserved-macro-identifier" -#endif - #if !defined(__HIPCC_RTC__) #include "hip_fp16_math_fwd.h" #include "amd_hip_vector_types.h" @@ -108,7 +102,3 @@ uint amd_mixed_dot(uint a, uint b, uint c, bool saturate) { #if !defined(__HIPCC_RTC__) #include #endif - -#if defined(__clang__) -#pragma clang diagnostic pop -#endif diff --git a/hipamd/include/hip/amd_detail/amd_surface_functions.h b/hipamd/include/hip/amd_detail/amd_surface_functions.h index b14e3d6520..108d67c4e5 100644 --- a/hipamd/include/hip/amd_detail/amd_surface_functions.h +++ b/hipamd/include/hip/amd_detail/amd_surface_functions.h @@ -23,19 +23,6 @@ THE SOFTWARE. #ifndef HIP_INCLUDE_HIP_AMD_DETAIL_SURFACE_FUNCTIONS_H #define HIP_INCLUDE_HIP_AMD_DETAIL_SURFACE_FUNCTIONS_H -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wreserved-macro-identifier" -#pragma clang diagnostic ignored "-Wc++17-extensions" -#pragma clang diagnostic ignored "-Wreserved-identifier" -#pragma clang diagnostic ignored "-Wc++98-compat" -#pragma clang diagnostic ignored "-Wold-style-cast" -#pragma clang diagnostic ignored "-Wextra-semi-stmt" -#pragma clang diagnostic ignored "-Wunused-variable" -#pragma clang diagnostic ignored "-Wunused-parameter" -#pragma clang diagnostic ignored "-Wunused-template" -#endif - #if defined(__cplusplus) #if !defined(__HIPCC_RTC__) @@ -254,8 +241,4 @@ static __device__ __hip_img_chk__ void surfCubemapLayeredwrite(T* data, hipSurfa #endif -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif diff --git a/hipamd/include/hip/amd_detail/amd_warp_functions.h b/hipamd/include/hip/amd_detail/amd_warp_functions.h index c43b81f969..559ab20b33 100644 --- a/hipamd/include/hip/amd_detail/amd_warp_functions.h +++ b/hipamd/include/hip/amd_detail/amd_warp_functions.h @@ -23,17 +23,6 @@ THE SOFTWARE. #ifndef HIP_INCLUDE_HIP_AMD_DETAIL_WARP_FUNCTIONS_H #define HIP_INCLUDE_HIP_AMD_DETAIL_WARP_FUNCTIONS_H -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wreserved-identifier" -#pragma clang diagnostic ignored "-Wreserved-macro-identifier" -#pragma clang diagnostic ignored "-Wsign-conversion" -#pragma clang diagnostic ignored "-Wold-style-cast" -#pragma clang diagnostic ignored "-Wc++98-compat" -#pragma clang diagnostic ignored "-Wc++98-compat-pedantic" -#pragma clang diagnostic ignored "-Wunsafe-buffer-usage" -#endif - __device__ static inline unsigned __hip_ds_bpermute(int index, unsigned src) { union { int i; unsigned u; float f; } tmp; tmp.u = src; tmp.i = __builtin_amdgcn_ds_bpermute(index, tmp.i); @@ -502,8 +491,4 @@ unsigned long long __shfl_xor(unsigned long long var, int lane_mask, int width = return tmp1; } -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif diff --git a/hipamd/include/hip/amd_detail/device_library_decls.h b/hipamd/include/hip/amd_detail/device_library_decls.h index c055ad1d8a..edc4692c83 100644 --- a/hipamd/include/hip/amd_detail/device_library_decls.h +++ b/hipamd/include/hip/amd_detail/device_library_decls.h @@ -31,13 +31,6 @@ THE SOFTWARE. #ifndef HIP_INCLUDE_HIP_AMD_DETAIL_DEVICE_LIBRARY_DECLS_H #define HIP_INCLUDE_HIP_AMD_DETAIL_DEVICE_LIBRARY_DECLS_H -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wc++98-compat-pedantic" -#pragma clang diagnostic ignored "-Wreserved-identifier" -#pragma clang diagnostic ignored "-Wold-style-cast" -#endif - #if !defined(__HIPCC_RTC__) #include "hip/amd_detail/host_defines.h" #endif @@ -137,8 +130,4 @@ __device__ inline static __local void* __to_local(unsigned x) { return (__local #define __CLK_LOCAL_MEM_FENCE 0x01 typedef unsigned __cl_mem_fence_flags; -#if defined(__clang__) -#pragma clang diagnostic pop -#endif - #endif diff --git a/hipamd/include/hip/amd_detail/hip_cooperative_groups_helper.h b/hipamd/include/hip/amd_detail/hip_cooperative_groups_helper.h index 358cbe14ad..15607b9683 100644 --- a/hipamd/include/hip/amd_detail/hip_cooperative_groups_helper.h +++ b/hipamd/include/hip/amd_detail/hip_cooperative_groups_helper.h @@ -40,14 +40,6 @@ THE SOFTWARE. #define __align__(x) __attribute__((aligned(x))) #endif -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wreserved-macro-identifier" -#pragma clang diagnostic ignored "-Wc++98-compat" -#pragma clang diagnostic ignored "-Wc++98-compat-pedantic" -#pragma clang diagnostic ignored "-Wshorten-64-to-32" -#endif - #if !defined(__CG_QUALIFIER__) #define __CG_QUALIFIER__ __device__ __forceinline__ #endif @@ -247,9 +239,6 @@ __CG_STATIC_QUALIFIER__ unsigned int masked_bit_count(lane_mask x, unsigned int /** * @} */ -#if defined(__clang__) -#pragma clang diagnostic pop -#endif #endif // __cplusplus #endif // HIP_INCLUDE_HIP_AMD_DETAIL_HIP_COOPERATIVE_GROUPS_HELPER_H diff --git a/hipamd/include/hip/amd_detail/math_fwd.h b/hipamd/include/hip/amd_detail/math_fwd.h index dbfd6e061a..9951f8fc34 100644 --- a/hipamd/include/hip/amd_detail/math_fwd.h +++ b/hipamd/include/hip/amd_detail/math_fwd.h @@ -22,12 +22,6 @@ THE SOFTWARE. #pragma once -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wundef" -#pragma clang diagnostic ignored "-Wreserved-identifier" -#endif - #if !defined(__HIPCC_RTC__) #include "host_defines.h" #include "amd_hip_vector_types.h" // For Native_vec_ @@ -702,7 +696,3 @@ double __ocml_fma_rtz_f64(double, double, double); #if defined(__cplusplus) } // extern "C" #endif - -#if defined(__clang__) -#pragma clang diagnostic pop -#endif diff --git a/hipamd/include/hip/amd_detail/ockl_image.h b/hipamd/include/hip/amd_detail/ockl_image.h index 136c6e352f..50223add4f 100644 --- a/hipamd/include/hip/amd_detail/ockl_image.h +++ b/hipamd/include/hip/amd_detail/ockl_image.h @@ -22,11 +22,6 @@ THE SOFTWARE. #pragma once -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wreserved-identifier" -#endif - #if !defined(__HIPCC_RTC__) #include #endif @@ -180,7 +175,3 @@ __device__ int __ockl_image_channel_order_CM(unsigned int ADDRESS_SPACE_CONSTANT __device__ int __ockl_image_channel_order_CMa(unsigned int ADDRESS_SPACE_CONSTANT* i); } - -#if defined(__clang__) -#pragma clang diagnostic pop -#endif \ No newline at end of file diff --git a/hipamd/include/hip/amd_detail/texture_fetch_functions.h b/hipamd/include/hip/amd_detail/texture_fetch_functions.h index dc45561ff6..0e95b7a7c5 100644 --- a/hipamd/include/hip/amd_detail/texture_fetch_functions.h +++ b/hipamd/include/hip/amd_detail/texture_fetch_functions.h @@ -22,18 +22,6 @@ THE SOFTWARE. #pragma once -#if defined(__clang__) -#pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wc++17-extensions" -#pragma clang diagnostic ignored "-Wreserved-identifier" -#pragma clang diagnostic ignored "-Wc++98-compat" -#pragma clang diagnostic ignored "-Wold-style-cast" -#pragma clang diagnostic ignored "-Wextra-semi-stmt" -#pragma clang diagnostic ignored "-Wunused-variable" -#pragma clang diagnostic ignored "-Wunused-parameter" -#pragma clang diagnostic ignored "-Wunused-template" -#endif - #if defined(__cplusplus) #if !defined(__HIPCC_RTC__) @@ -501,7 +489,3 @@ static __forceinline__ __device__ __hip_img_chk__ __hip_tex2dgather_ret_t