mirror of
https://github.com/ROCm/composable_kernel.git
synced 2026-06-29 11:16:59 +00:00
c24e528481
[CK] suppress compiler warnings while building pytorch. (#7760) ## Motivation Recently added compiler flags that are required to suppress false warnings by latest staging compiler are not recognized by older compiler versions and are triggering an avalanche of warnings. Previous attempt to suppress them by using -Wno-unknown-warning-option flag didn't help, because that flag wasn't recognized either and just added more warnings. I've verified that current approach by checking the clang version actually works as intended and makes the warnings go away. ## Technical Details <!-- Explain the changes along with any relevant GitHub links. --> ## Test Plan <!-- Explain any relevant testing done to verify this PR. --> ## Test Result <!-- Briefly summarize test outcomes. --> ## Submission Checklist - [ ] Look over the contributing guidelines at https://github.com/ROCm/ROCm/blob/develop/CONTRIBUTING.md#pull-requests.
54 lines
1.3 KiB
C++
54 lines
1.3 KiB
C++
// Copyright (c) Advanced Micro Devices, Inc., or its affiliates.
|
|
// SPDX-License-Identifier: MIT
|
|
|
|
#pragma once
|
|
|
|
#include "ck/ck.hpp"
|
|
#include "c_style_pointer_cast.hpp"
|
|
|
|
// Address Space for AMDGCN
|
|
// https://llvm.org/docs/AMDGPUUsage.html#address-space
|
|
|
|
namespace ck {
|
|
|
|
enum struct AddressSpaceEnum
|
|
{
|
|
Generic,
|
|
Global,
|
|
Lds,
|
|
Sgpr,
|
|
Vgpr,
|
|
};
|
|
|
|
template <typename T>
|
|
__device__ T* cast_pointer_to_generic_address_space(T CK_CONSTANT_ADDRESS_SPACE* p)
|
|
{
|
|
// cast a pointer in "Constant" address space (4) to "Generic" address space (0)
|
|
// only c-style pointer cast seems be able to be compiled
|
|
#ifdef __clang__
|
|
#pragma clang diagnostic push
|
|
#pragma clang diagnostic ignored "-Wold-style-cast"
|
|
#endif
|
|
return (T*)p; // NOLINT(old-style-cast)
|
|
#ifdef __clang__
|
|
#pragma clang diagnostic pop
|
|
#endif
|
|
}
|
|
|
|
template <typename T>
|
|
__host__ __device__ T CK_CONSTANT_ADDRESS_SPACE* cast_pointer_to_constant_address_space(T* p)
|
|
{
|
|
// cast a pointer in "Generic" address space (0) to "Constant" address space (4)
|
|
// only c-style pointer cast seems be able to be compiled
|
|
#ifdef __clang__
|
|
#pragma clang diagnostic push
|
|
#pragma clang diagnostic ignored "-Wold-style-cast"
|
|
#endif
|
|
return (T CK_CONSTANT_ADDRESS_SPACE*)p; // NOLINT(old-style-cast)
|
|
#ifdef __clang__
|
|
#pragma clang diagnostic pop
|
|
#endif
|
|
}
|
|
|
|
} // namespace ck
|