Merge branch 'develop' into qdot_clip_opts

CMakeLists.txt

@@ -228,14 +228,12 @@ rocm_enable_cppcheck(
         shadowVar
         shadowVariable
         unsafeClassDivZero
+        # Disable because of too many FPs
+        arithOperationsOnVoidPointer
         definePrefix:*test/include/test.hpp
         ctuOneDefinitionRuleViolation:*test/*
         useSmartPointer:*src/api/api.cpp
         useSmartPointer:*make_shared_array.hpp
-        constParameter:*src/targets/gpu/*.cpp
-        constParameter:*src/targets/gpu/*.hpp
-        # Suppress mlir_conv.cpp since this file will be deleted
-        *:*src/targets/gpu/mlir_conv.cpp
     FORCE
     INCONCLUSIVE
     RULE_FILE
@@ -252,10 +250,13 @@ rocm_enable_cppcheck(
         ${CMAKE_CURRENT_SOURCE_DIR}/src/targets/gpu/kernels/include
         ${CMAKE_CURRENT_SOURCE_DIR}/test/include
     DEFINE
+        MIGRAPHX_MLIR=1
         CPPCHECK=1
         __device__=
         __host__=
         __global__=
+    UNDEFINE
+        MIGRAPHX_USE_CLANG_TIDY
 )
 
 enable_testing()

cppcheck.rules

@@ -107,6 +107,24 @@
         <summary>Use make_shared or make_unique instead of new</summary>
     </message>
 </rule>
+<rule>
+    <tokenlist>raw</tokenlist>
+    <pattern><![CDATA[ [^\(,;{}:]+ \w+ && (\w+|\(|\+|\-|\*)]]></pattern>
+    <message>
+        <id>UseNamedLogicOperator</id>
+        <severity>style</severity>
+        <summary>Use 'and' instead of &&</summary>
+    </message>
+</rule>
+<rule>
+    <tokenlist>raw</tokenlist>
+    <pattern><![CDATA[ (if|while) \([^\)&]+&& ]]></pattern>
+    <message>
+        <id>UseNamedLogicOperator</id>
+        <severity>style</severity>
+        <summary>Use 'and' instead of &&</summary>
+    </message>
+</rule>
 <rule>
     <tokenlist>raw</tokenlist>
     <pattern><![CDATA[ \|\| ]]></pattern>

dev-requirements.txt

@@ -25,6 +25,6 @@ ROCmSoftwarePlatform/rocm-recipes
 facebook/zstd@v1.4.5 -X subdir -DCMAKE_DIR=build/cmake
 ccache@v4.1 -DENABLE_TESTING=OFF
 pcre,pfultz2/pcre@8.45 -H sha256:d6f7182602a775a7d500a0cedca6449af0400c6493951513046d17615ed0bf11
-danmar/cppcheck@2.9 -DHAVE_RULES=1
+danmar/cppcheck@bb2711c22a0be09efe7f1a8da3030876471026c8 -DHAVE_RULES=1 # 2.11
 RadeonOpenCompute/rocm-cmake@027404a8326da6e7e9338e0b81f9428660190724 --build
 -f requirements.txt

examples/migraphx/cpp_parse_load_save/parse_load_save.cpp

@@ -114,7 +114,7 @@ int main(int argc, char** argv)
 char* getCmdOption(char** begin, char** end, const std::string& option)
 {
     char** itr = std::find(begin, end, option);
-    if(itr != end && ++itr != end)
+    if(itr != end and ++itr != end)
     {
         return *itr;
     }

examples/vision/cpp_mnist/mnist_inference.cpp

@@ -160,9 +160,9 @@ int main(int argc, char** argv)
     auto lengths = shape.lengths();
     auto num_results =
         std::accumulate(lengths.begin(), lengths.end(), 1, std::multiplies<size_t>());
-    float* results = reinterpret_cast<float*>(outputs[0].data());
-    float* max     = std::max_element(results, results + num_results);
-    int answer     = max - results;
+    float* results   = reinterpret_cast<float*>(outputs[0].data());
+    const float* max = std::max_element(results, results + num_results);
+    int answer       = max - results;
 
     std::cout << std::endl
               << "Randomly chosen digit: " << rand_digit << std::endl
@@ -192,12 +192,12 @@ void read_nth_digit(const int n, std::vector<float>& digit)
         for(int i = 0; i < HEIGHT * WIDTH; ++i)
         {
             unsigned char temp = 0;
-            file.read((char*)&temp, sizeof(temp));
+            file.read(reinterpret_cast<char*>(&temp), sizeof(temp));
             if(d == n)
             {
                 float data = temp / 255.0;
                 digit.push_back(data);
-                std::cout << SYMBOLS[(int)(data * 10) % 11];
+                std::cout << SYMBOLS[static_cast<int>(data * 10) % 11];
                 if((i + 1) % WIDTH == 0)
                     std::cout << std::endl;
             }

src/api/api.cpp

@@ -899,7 +899,7 @@ migraphx_dynamic_dimensions_assign_to(migraphx_dynamic_dimensions_t output,
 
 extern "C" migraphx_status
 migraphx_dynamic_dimensions_create(migraphx_dynamic_dimensions_t* dynamic_dimensions,
-                                   const_migraphx_dynamic_dimension_t* ptr,
+                                   const const_migraphx_dynamic_dimension_t* ptr,
                                    size_t size)
 {
     auto api_error_result = migraphx::try_([&] {
@@ -1432,7 +1432,7 @@ extern "C" migraphx_status migraphx_instructions_assign_to(migraphx_instructions
 }
 
 extern "C" migraphx_status migraphx_instructions_create(migraphx_instructions_t* instructions,
-                                                        const_migraphx_instruction_t* ptr,
+                                                        const const_migraphx_instruction_t* ptr,
                                                         size_t size)
 {
     auto api_error_result = migraphx::try_([&] {

src/api/include/migraphx/migraphx.h

@@ -209,7 +209,7 @@ MIGRAPHX_C_EXPORT migraphx_status migraphx_dynamic_dimensions_assign_to(
 
 MIGRAPHX_C_EXPORT migraphx_status
 migraphx_dynamic_dimensions_create(migraphx_dynamic_dimensions_t* dynamic_dimensions,
-                                   const_migraphx_dynamic_dimension_t* ptr,
+                                   const const_migraphx_dynamic_dimension_t* ptr,
                                    size_t size);
 
 MIGRAPHX_C_EXPORT migraphx_status
@@ -377,7 +377,7 @@ MIGRAPHX_C_EXPORT migraphx_status migraphx_instructions_assign_to(
     migraphx_instructions_t output, const_migraphx_instructions_t input);
 
 MIGRAPHX_C_EXPORT migraphx_status migraphx_instructions_create(
-    migraphx_instructions_t* instructions, const_migraphx_instruction_t* ptr, size_t size);
+    migraphx_instructions_t* instructions, const const_migraphx_instruction_t* ptr, size_t size);
 
 MIGRAPHX_C_EXPORT migraphx_status migraphx_modules_destroy(migraphx_modules_t modules);
 

src/api/migraphx.py

@@ -79,7 +79,8 @@ def dynamic_dimension(h):
 def dynamic_dimensions(h):
     h.constructor(
         'create',
-        api.params(ptr='const_migraphx_dynamic_dimension_t*', size='size_t'),
+        api.params(ptr='const const_migraphx_dynamic_dimension_t*',
+                   size='size_t'),
         fname='migraphx::to_obj_vector<const_migraphx_dynamic_dimension_t>')
     h.method('size', returns='size_t')
     h.method('get',
@@ -215,7 +216,7 @@ def instruction(h):
 def instructions(h):
     h.constructor(
         'create',
-        api.params(ptr='const_migraphx_instruction_t*', size='size_t'),
+        api.params(ptr='const const_migraphx_instruction_t*', size='size_t'),
         fname='migraphx::to_obj_vector<const_migraphx_instruction_t>')
 
 

src/driver/argument_parser.hpp

@@ -338,7 +338,7 @@ struct argument_parser
 
     MIGRAPHX_DRIVER_STATIC auto file_exist()
     {
-        return validate([](auto&, auto&, auto& params) {
+        return validate([](auto&, auto&, const auto& params) {
             if(params.empty())
                 throw std::runtime_error("No argument passed.");
             if(not fs::exists(params.back()))
@@ -348,13 +348,12 @@ struct argument_parser
 
     MIGRAPHX_DRIVER_STATIC auto matches(const std::unordered_set<std::string>& names)
     {
-        return validate([=](auto&, auto&, auto& params) {
-            for(const auto& p : params)
-            {
-                if(names.count(p) == 0)
-                    throw std::runtime_error("Invalid argument: " + p + ". Valid arguments are {" +
-                                             to_string_range(names) + "}");
-            }
+        return validate([=](auto&, auto&, const auto& params) {
+            auto invalid_param = std::find_if(
+                params.begin(), params.end(), [&](const auto& p) { return names.count(p) == 0; });
+            if(invalid_param != params.end())
+                throw std::runtime_error("Invalid argument: " + *invalid_param +
+                                         ". Valid arguments are {" + to_string_range(names) + "}");
         });
     }
 

src/fuse_pointwise.cpp

@@ -41,7 +41,7 @@ static literal get_scalar(instruction_ref ins)
     if(ins->name() == "contiguous")
         return get_scalar(ins->inputs().front());
     const auto& s = ins->get_shape();
-    if(s.elements() != 1 && not(s.scalar()))
+    if(s.elements() != 1 and not(s.scalar()))
         return {};
     if(not ins->can_eval())
         return {};

src/fuse_reduce.cpp

@@ -52,7 +52,7 @@ struct fused_reduce
     {
         if(mods.size() != 1)
             MIGRAPHX_THROW("should have one submodule.");
-        auto* sm = mods.front();
+        const auto* sm = mods.front();
         if(sm->get_output_shapes().size() != 1)
             MIGRAPHX_THROW("Only one output supported");
         auto names = sm->get_parameter_names();
@@ -143,7 +143,7 @@ insert_module_in_submodule(module_ref sm,
 }
 
 static std::vector<instruction_ref>
-find_inputs(module_ref sm,
+find_inputs(const_module_ref sm,
             const module& parent,
             const std::unordered_map<instruction_ref, instruction_ref>& map_ins)
 {

src/include/migraphx/algorithm.hpp

@@ -100,8 +100,7 @@ inline size_t levenshtein_distance(const std::string& s1, const std::string& s2)
 
     std::vector<size_t> d(l2 + 1);
 
-    for(size_t j = 1; j <= l2; j++)
-        d[j] = j;
+    std::iota(d.begin(), d.end(), 0);
 
     for(size_t i = 1; i <= l1; i++)
     {

src/include/migraphx/op/convolution.hpp

@@ -82,7 +82,7 @@ struct convolution
         const auto input_ndim   = inputs[0].ndim();
         const auto padding_size = padding.size();
 
-        if(input_ndim != padding_size / 2 + 2 && input_ndim != padding_size + 2)
+        if(input_ndim != padding_size / 2 + 2 and input_ndim != padding_size + 2)
         {
             MIGRAPHX_THROW("CONVOLUTION: input and attribute size mismatch!");
         }

src/include/migraphx/op/if_op.hpp

@@ -71,7 +71,7 @@ struct if_op
         std::unordered_map<std::string, argument> params;
 
         std::set<std::string> pnames;
-        for(const auto& smod : mods)
+        for(const_module_ref smod : mods)
         {
             auto names = smod->get_parameter_names();
             pnames.insert(names.begin(), names.end());

src/include/migraphx/op/loop.hpp

@@ -59,9 +59,9 @@ struct loop
             MIGRAPHX_THROW("LOOP: operator should have one submodule.");
         }
 
-        const auto& mod     = mods.front();
-        auto mod_out_shapes = mod->get_output_shapes();
-        auto dep_param_num  = inputs.size() - 2;
+        const_module_ref mod = mods.front();
+        auto mod_out_shapes  = mod->get_output_shapes();
+        auto dep_param_num   = inputs.size() - 2;
 
         // first item of the mod output shapes is condition used in loop,
         // which is not needed to compute output shape

src/instruction.cpp

@@ -389,7 +389,7 @@ void instruction::print(std::ostream& os,
     if(not ins->module_inputs().empty())
     {
         std::string delim = ", [";
-        for(auto&& mod_arg : ins->module_inputs())
+        for(const const_module_ref& mod_arg : ins->module_inputs())
         {
             os << delim << mod_arg->name();
             delim = ", ";

src/module.cpp

@@ -873,12 +873,11 @@ module::print_py(std::ostream& os,
             if(ins->name() == "@literal")
             {
                 os << mname << ".add_literal(";
-                bool use_abs = false;
-                ins->get_literal().visit([&](auto v) {
-                    use_abs = std::none_of(v.begin(), v.end(), [](auto x) { return x < 0; });
-                });
+                const bool use_abs = false;
                 // Disable abs for now
-                use_abs = false;
+                // ins->get_literal().visit([&](auto v) {
+                //     use_abs = std::none_of(v.begin(), v.end(), [](auto x) { return x < 0; });
+                // });
                 if(use_abs)
                     os << "migraphx.abs_literal(";
                 os << "migraphx.generate_argument(";

src/onnx/parse_constant_of_shape.cpp

@@ -55,16 +55,15 @@ struct parse_constant_of_shape : op_parser<parse_constant_of_shape>
             l_val = literal({shape::float_type, {1}, {0}}, {0.0f});
         }
 
-        // input is empty, output is a scalar
-        auto type = l_val.get_shape().type();
-
         if(args.empty())
         {
             MIGRAPHX_THROW("ConstantOfShape : must have 1 input!");
         }
         else
         {
             migraphx::shape s;
+            // input is empty, output is a scalar
+            auto type = l_val.get_shape().type();
             // empty input tensor, output is a scalar
             if(args[0]->get_shape().elements() == 0)
             {

src/program.cpp

@@ -223,7 +223,7 @@ void program::compile(const std::vector<target>& targets, std::vector<compile_op
     // Gather all the target roots
     std::unordered_multimap<std::size_t, module_ref> roots;
     auto mods = this->get_modules();
-    for(auto* mod : mods)
+    for(const auto* mod : mods)
     {
         for(const auto& ins : *mod)
         {
@@ -548,7 +548,7 @@ std::vector<argument> program::eval(parameter_map params, execution_environment
             ins_out[x] = ss.str();
         });
         ret = generic_eval(*this, contexts, std::move(params), [&](instruction_ref ins, auto f) {
-            auto& ctx = contexts[ins->get_target_id()];
+            const auto& ctx = contexts[ins->get_target_id()];
             ctx.finish();
             std::cout << "Run instruction: " << ins_out.at(ins) << std::endl;
             timer t{};
@@ -728,7 +728,7 @@ static void mod_from_val(module_ref mod,
                                std::back_inserter(module_inputs),
                                [&](const value& i) { return map_mods.at(i.to<std::string>()); });
 
-                for(auto& smod : module_inputs)
+                for(const auto& smod : module_inputs)
                 {
                     mod_from_val(smod, v, instructions, map_mods);
                 }
@@ -1186,7 +1186,7 @@ void program::remove_unused_modules()
     std::vector<module*> unused;
     generic_get_unused_modules(
         impl->modules, generic_get_modules(this->get_main_module()), std::back_inserter(unused));
-    for(auto* m : unused)
+    for(const auto* m : unused)
         this->remove_module(m->name());
 }
 

src/sqlite.cpp

@@ -48,6 +48,7 @@ struct sqlite_impl
     template <class F>
     void exec(const char* sql, F f)
     {
+        // cppcheck-suppress constParameterPointer
         auto callback = [](void* obj, auto... xs) -> int {
             try
             {

src/targets/cpu/target.cpp

@@ -61,7 +61,7 @@ namespace cpu {
 
 std::string target::name() const { return "cpu"; }
 
-// cppcheck-suppress constParameter
+// cppcheck-suppress constParameterReference
 std::vector<pass> target::get_passes(migraphx::context& gctx, const compile_options&) const
 {
     auto& ctx = any_cast<context>(gctx);

src/targets/gpu/device/include/migraphx/gpu/device/nary.hpp

@@ -124,7 +124,7 @@ void nary_broadcast_vec_impl(
                     buffer[i] = binput.data()[i];
                 }
                 __syncthreads();
-                auto* bp = as_pointer(buffer);
+                const auto* bp = as_pointer(buffer);
                 // Process the data
                 for(size_t i = idx.global; i < nelements; i += nglobal)
                 {
@@ -219,7 +219,7 @@ void nary_double_broadcast_vec_impl(
                     buffer[i + bdim_vec_len] = binput2.data()[i];
                 }
                 __syncthreads();
-                auto* bp = as_pointer(buffer);
+                const auto* bp = as_pointer(buffer);
                 // Process the data
                 for(size_t i = idx.global; i < nelements; i += nglobal)
                 {

src/targets/gpu/device/topk.cpp

@@ -72,12 +72,12 @@ struct hip_heap_vector
             index_int l    = 2 * index + 1;
             index_int r    = 2 * index + 2;
 
-            if(l < n && compare(data[data_index(l)], data[data_index(index)]))
+            if(l < n and compare(data[data_index(l)], data[data_index(index)]))
             {
                 index = l;
             }
 
-            if(r < n && compare(data[data_index(r)], data[data_index(index)]))
+            if(r < n and compare(data[data_index(r)], data[data_index(index)]))
             {
                 index = r;
                 if(compare(data[data_index(l)], data[data_index(r)]))