Skip to content

Developer's Guide

Jump to bottom
Charlie Lin edited this page Jul 12, 2024 · 13 revisions

Developer's Guide

Introduction to MIGraphX

See the dev_intro document: https://github.com/ROCm/AMDMIGraphX/blob/develop/docs/dev/dev_intro.rst

Contributing

All pull requests submitted will go through jenkins testing and review. To pass jenkin, please make sure:

  1. There are no warnings when building.
  2. The output of make analyze produces no diagnostics.
  3. The code is formatted with clang-format-5.0.
  4. All unit tests pass on both gcc and hcc when running make check.

First, a pull request should be for one feature. Do not submit pull requests with multiple features in it. Consider splitting it into multiple smaller pull requests.

Secondly, make sure you have proper tests to verify your pull request works as expected. Pull requests will not be accepted without the proper tests!

Third, the pull request should include the related documentation changes.

Tests

A test can be added by adding a .cpp file under the test/ directory, such as:

#include <test.hpp>

int main()
{
    int a = 1;
    EXPECT(a == 1);
}

The test.hpp header provides some tools useful for testing values such as EXPECT macro which will check the value and then print out a message if the predicate is false.

All tests under test/ are ran without using the gpu backend since we can use more tools when running the tests(such as address sanitizers or coverage reports). If a test needs the gpu backend it can be placed in the gpu/ directory.

If a test uses onnx it can be placed in the onnx/ directory.

Tests need to be thorough, fast, isolated, consistently repeatable, and as simple as possible. We strive to have as much coverage as possible in our unit tests. As such, we try to have tests both for normal behavior and for error conditions.

Style

We use standard C++ naming scheme. So everything is named with snake_case except template parameters are CamelCase and macros use UPPER_CASE.

All macros must be prefixed with MIGRAPHX_. A macro should only be used unless absolutely necessary. We don't use a macro when it can be written as a function or variable.

See also:

We use C++11 using type alias instead of typedef. So we will write using my_int = int instead of typedef int my_int. This is to make it consistent with templated type aliases, which must be written with using.

See also:

We use nullptr instead of NULL or 0.

See also:

Manage Pointers

We don't manually manage resources or pointers as this can lead to memory leaks especially in the presence of exceptions.

For raw memory allocations, they can be allocated with either std::make_unique or std::make_shared. If an array of elements need to be allocated then a std::vector can be used.

For non-memory resources such as FILE*, then the MIGRAPHX_MANAGE_PTR macro can be used to create a std::unique_ptr which will call the correct function to free the resource. It takes the pointer, and the function that should be called to release the resource:

using file_ptr = MIGRAPHX_MANAGE_PTR(FILE*, fclose);

It can be constructed by passing the newly created pointer to its constructor:

file_ptr f{fopen("some_file", "r")};

The raw pointer can be extracted from it using the .get() method.

Now some APIs do not return the new pointer. Instead, a raw pointer needs to be created first. You can see an example of this when using hipMalloc:

using hip_ptr = MIGRAPHX_MANAGE_PTR(void, hipFree);

hip_ptr allocate_gpu(std::size_t sz)
{
    void* result;
    auto status = hipMalloc(&result, sz);
    if(status != hipSuccess)
        MIGRAPHX_THROW("Gpu allocation failed");
    return hip_ptr{result};
}

See also:

Raw loops

The standard algorithms should be preferred over raw loops. Raw loop suffer from many issues:

  • Difficult to reason about and difficult to prove post conditions
  • Error prone and likely to fail under non-obvious conditions
  • Introduce non-obvious performance problems
  • Complicates reasoning about the surrounding code

Also, algorithms can be optimized much further than just basic raw loops, so there can be a performance benefit to using algorithms.

If there isn't an algorithm able to replace a raw loop, it might be a good idea to add a new algorithm to handle it.

See also:

Type Erasure

We use type erasure for dynamic polymorphism instead of inheritance because inheritance-based polymorphism forces indirection through pointers which has problems including:

  • Changes the semantics of copy, assignment, and equality
  • Leads to incidental data structures
  • Inefficient due to heap allocation and aliasing
  • Can be good as a global variable

To help with writing the boilerplate needed for type erasure we use a python template script to generate the type erasure code. This can be found under tools/ directory. This requires python 3.6 to use. Calling make generate will generate code for all headers under tools/include and place them in src/include/migraph. This can be necessary when there are interface changes or new type erased classes are added.

Clone this wiki locally