Skip to content

Commit

Permalink
[MIOpen2.14.0] fix issues in release
Browse files Browse the repository at this point in the history
  • Loading branch information
@junliume
junliume committed Dec 17, 2021
1 parent 76ad155 commit dea91c9
Show file tree
Hide file tree
Showing 3 changed files with 43 additions and 103 deletions.
59 changes: 0 additions & 59 deletions Jenkinsfile
View file
Original file line number Diff line number Diff line change
@@ -1,5 +1,4 @@
def rocmnode(name) {
<<<<<<< HEAD
return 'rocmtest && miopen && ' + name
}

Expand Down Expand Up @@ -277,42 +276,12 @@ pipeline {
Tensile_build_env = "MIOPEN_DEBUG_HIP_KERNELS=0 "
Tensile_setup = " -DMIOPEN_TEST_MIOTENSILE=ON -DMIOPEN_USE_MIOPENTENSILE=ON -DMIOPEN_USE_ROCBLAS=OFF"
}
parameters {
booleanParam(
name: "STATIC_CHECKS",
defaultValue: true,
description: "")
booleanParam(
name: "SMOKE_TESTS",
defaultValue: true,
description: "")
booleanParam(
name: "FULL_TESTS",
defaultValue: true,
description: "")
booleanParam(
name: "MIOPENTENSILE",
defaultValue: false,
description: "")
booleanParam(
name: "PACKAGES",
defaultValue: true,
description: "")
}
stages{
stage("Static checks"){
when { expression { params.STATIC_CHECKS && !params.DISABLE_ALL_STAGES } }
parallel{
<<<<<<< HEAD
stage('Clang Tidy') {
agent{ label rocmnode("nogpu") }
||||||| merged common ancestors
stage('Clang Tidy') {
agent{ label rocmnode("rocmtest") }
=======
stage('Hip Tidy') {
agent{ label rocmnode("nogpu") }
>>>>>>> release/rocm-rel-4.3
environment{
setup_cmd = "CXX='/opt/rocm/llvm/bin/clang++' cmake -DMIOPEN_BACKEND=HIP -DBUILD_DEV=On .. "
build_cmd = "make -j\$(nproc) -k analyze"
Expand Down Expand Up @@ -368,7 +337,6 @@ pipeline {
parallel{
stage('Fp32 OpenCL Debug + Codecov') {
agent{ label rocmnode("vega") }
>>>>>>> release/rocm-rel-4.3
steps{
buildHipClangJobAndReboot(compiler: 'g++', build_type: 'debug', config_targets: Smoke_targets, codecov: true)
}
Expand Down Expand Up @@ -425,7 +393,6 @@ pipeline {
}
stage('Fp32 HipNoGPU Debug') {
agent{ label rocmnode("nogpu") }
>>>>>>> release/rocm-rel-4.3
environment{
HipNoGPU_flags = "-DMIOPEN_BACKEND=HIPNOGPU -DMIOPEN_INSTALL_CXX_HEADERS=On"
build_cmd = "make -j\$(nproc)"
Expand Down Expand Up @@ -522,16 +489,6 @@ pipeline {
}
}
}
<<<<<<< HEAD

// Run fp16, bfp16, and int8 quick tests
stage("Fast low precision"){
when { expression { params.SMOKE_TESTS } }
||||||| merged common ancestors

// Run fp16, bfp16, and int8 quick tests
stage("Fast low precision"){
=======
stage("Smoke Fp16/Bf16/Int8"){
when { expression { params.SMOKE_FP16_BF16_INT8 && !params.DISABLE_ALL_STAGES } }
environment{
Expand Down Expand Up @@ -885,30 +842,14 @@ pipeline {
stage("Packages"){
when { expression { params.PACKAGES && !params.DISABLE_ALL_STAGES } }
parallel {
<<<<<<< HEAD
stage('OpenCL Release Package') {
agent{ label rocmnode("nogpu") }
||||||| merged common ancestors
stage('OpenCL Release Package') {
agent{ label rocmnode("rocmtest") }
=======
stage('OpenCL Package') {
agent{ label rocmnode("nogpu") }
>>>>>>> release/rocm-rel-4.3
steps{
buildHipClangJobAndReboot(compiler: 'g++', package_build: "true", gpu_arch: "gfx900;gfx906;gfx908;gfx90a")
}
}
<<<<<<< HEAD
stage("HIP Release Package /opt/rocm"){
agent{ label rocmnode("nogpu") }
||||||| merged common ancestors
stage("HIP Release Package /opt/rocm"){
agent{ label rocmnode("rocmtest") }
=======
stage("HIP Package /opt/rocm"){
agent{ label rocmnode("nogpu") }
>>>>>>> release/rocm-rel-4.3
steps{
buildHipClangJobAndReboot( package_build: "true", prefixpath: '/opt/rocm', gpu_arch: "gfx900;gfx906;gfx908;gfx90a")
}
Expand Down
44 changes: 0 additions & 44 deletions src/comgr.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -202,20 +202,6 @@ static void AddCompilerOptions(OptionList& list, const miopen::TargetProperties&
#if ROCM_FEATURE_TARGETID_OFF
// It seems like these options are used only in codegen.
// However it seems ok to pass these to compiler.
<<<<<<< HEAD
if(target.Sramecc())
list.push_back("-msram-ecc");
else
list.push_back("-mno-sram-ecc");
#else
std::ignore = target;
#endif
||||||| merged common ancestors
if(IsEnabledFeatureSramEcc(target.Name()))
list.push_back("-msram-ecc");
else
list.push_back("-mno-sram-ecc");
=======
if(target.Sramecc())
{
if(*target.Sramecc())
Expand All @@ -226,7 +212,6 @@ static void AddCompilerOptions(OptionList& list, const miopen::TargetProperties&
#else
std::ignore = target;
#endif
>>>>>>> release/rocm-rel-4.3
list.push_back("-mllvm");
list.push_back("-amdgpu-internalize-symbols");
}
Expand Down Expand Up @@ -323,26 +308,13 @@ static void RemoveLinkOptionsUnwanted(OptionList& list)
/// \todo Get list of supported isa names from comgr and select.
static std::string GetIsaName(const miopen::TargetProperties& target)
{
<<<<<<< HEAD
#if ROCM_FEATURE_TARGETID_OFF
const char* const ecc_suffix = target.Sramecc() ? "+sram-ecc" : "";
return {"amdgcn-amd-amdhsa--" + target.Name() + ecc_suffix};
#else
const LcOptionTargetStrings lots(target);
return {"amdgcn-amd-amdhsa--" + lots.targetId};
#endif
||||||| merged common ancestors
const char* const ecc_suffix = IsEnabledFeatureSramEcc(device) ? "+sram-ecc" : "";
return {"amdgcn-amd-amdhsa--" + device + ecc_suffix};
=======
#if ROCM_FEATURE_TARGETID_OFF
const char* const ecc_suffix = (target.Sramecc() && *target.Sramecc()) ? "+sram-ecc" : "";
return {"amdgcn-amd-amdhsa--" + target.Name() + ecc_suffix};
#else
const LcOptionTargetStrings lots(target);
return {"amdgcn-amd-amdhsa--" + lots.targetId};
#endif
>>>>>>> release/rocm-rel-4.3
}

} // namespace lc
Expand Down Expand Up @@ -783,18 +755,10 @@ void BuildHip(const std::string& name,
+ " " + GetDebugCompilerOptionsInsert() //
+ " " + MIOPEN_STRINGIZE(HIP_COMPILER_FLAGS) +
(" -DHIP_PACKAGE_VERSION_FLAT=") + std::to_string(HIP_PACKAGE_VERSION_FLAT);
<<<<<<< HEAD
#if ROCM_FEATURE_LLVM_AMDGCN_BUFFER_ATOMIC_FADD_F32_RETURNS_FLOAT
if(miopen::solver::support_amd_buffer_atomic_fadd(target.Name()))
raw += "-DCK_AMD_BUFFER_ATOMIC_FADD_RETURNS_FLOAT=1";
#endif
||||||| merged common ancestors
=======
#if ROCM_FEATURE_LLVM_AMDGCN_BUFFER_ATOMIC_FADD_F32_RETURNS_FLOAT
if(miopen::solver::support_amd_buffer_atomic_fadd(target.Name()))
raw += " -DCK_AMD_BUFFER_ATOMIC_FADD_RETURNS_FLOAT=1";
#endif
>>>>>>> release/rocm-rel-4.3
auto optCompile = miopen::SplitSpaceSeparated(raw, compiler::lc::GetOptionsNoSplit());
compiler::lc::hip::RemoveCompilerOptionsUnwanted(optCompile);
action.SetOptionList(optCompile);
Expand All @@ -807,18 +771,10 @@ void BuildHip(const std::string& name,
+ " " + GetDebugCompilerOptionsInsert() //
+ " " + MIOPEN_STRINGIZE(HIP_COMPILER_FLAGS) +
(" -DHIP_PACKAGE_VERSION_FLAT=") + std::to_string(HIP_PACKAGE_VERSION_FLAT);
<<<<<<< HEAD
#if ROCM_FEATURE_LLVM_AMDGCN_BUFFER_ATOMIC_FADD_F32_RETURNS_FLOAT
if(miopen::solver::support_amd_buffer_atomic_fadd(target.Name()))
raw += "-DCK_AMD_BUFFER_ATOMIC_FADD_RETURNS_FLOAT=1";
#endif
||||||| merged common ancestors
=======
#if ROCM_FEATURE_LLVM_AMDGCN_BUFFER_ATOMIC_FADD_F32_RETURNS_FLOAT
if(miopen::solver::support_amd_buffer_atomic_fadd(target.Name()))
raw += " -DCK_AMD_BUFFER_ATOMIC_FADD_RETURNS_FLOAT=1";
#endif
>>>>>>> release/rocm-rel-4.3
#if COMGR_SUPPORTS_PCH
if(compiler::lc::hip::IsPchEnabled())
{
Expand Down
43 changes: 43 additions & 0 deletions src/kernels/static_composable_kernel/include/utility/static_kernel_ck_utils_type.hpp
View file
Original file line number Diff line number Diff line change
@@ -0,0 +1,43 @@
#ifndef CK_UTILS_TYPE_HPP
#define CK_UTILS_TYPE_HPP

#include "static_kernel_integral_constant.hpp"

namespace ck {

template <index_t... Is>
struct Sequence;

template <typename X, typename Y>
struct is_same : public integral_constant<bool, false>
{
};

template <typename X>
struct is_same<X, X> : public integral_constant<bool, true>
{
};

template <typename>
struct is_static : integral_constant<bool, false>
{
};

template <typename T, T X>
struct is_static<integral_constant<T, X>> : integral_constant<bool, true>
{
};

template <index_t... Is>
struct is_static<Sequence<Is...>> : integral_constant<bool, true>
{
};

template <typename T>
using remove_reference_t = typename std::remove_reference<T>::type;

template <typename T>
using remove_cv_t = typename std::remove_cv<T>::type;

} // namespace ck
#endif

0 comments on commit dea91c9

Please sign in to comment.