• Stars
    star
    159
  • Rank 235,916 (Top 5 %)
  • Language
    Fortran
  • License
    MIT License
  • Created almost 4 years ago
  • Updated 12 months ago

Reviews

There are no reviews yet. Be the first to send feedback to the community and the maintainers!

Repository Details

GPUFORT: S2S translation tool for CUDA Fortran and Fortran+X in the spirit of hipify

GPUFORT

This project develops a source to source translation tool that is able to convert:

  1. Fortran+OpenACC and CUDA Fortran -> Fortran + OpenMP 4.5+
  2. Fortran+OpenACC and CUDA Fortran -> Fortran + [GCC/AOMP OpenACC/MP runtime calls] + HIP C++

The result of the first translation process, can be compiled with AOMP, which has a Fortran frontend. The result of the second translation process can be compiled with hipfort or a combination of hipcc and gfortran. Note that a OpenACC runtime is only necessary for translating OpenACC code.

An overview of the different translation paths that we work on is shown below:

Image

NOTE: GPUFORT is a research project. We made it publicly available because we believe that it might be helpful for some. We want to stress that the code translation and code generation outputs produced by GPUFORT will in most cases require manual reviewing and fixing.

Installation and usage

Please take a look at the (slightly outdated) user guide.

Implementation details

This presentation gives an overview of GPUFORT's building blocks.

Limitations

  • GPUFORT is not a compiler (yet)

GPUFORT is not intended to be a compiler. It's main purpose is to be a translator that allows an experienced user to fix and tune the outcomes of the translation process. However, we believe GPUFORT can develop into an early-outlining compiler if enough effort is put into the project. Given that all code and especially the grammar is written in python3, GPUFORT can be developed at a quick pace.

  • GPUFORT assumes syntactically and functionally correct input

GPUFORT does only perform a small number of syntax checks as we assume that developers apply GPUFORT to code that can be run correctly on CUDA devices. (We plan to add the option to prescribe a user-specified syntax checker tool.)

  • GPUFORT does a bad job in analyzing what code parts can be offloaded and which ones not
  • GPUFORT does a bad job in reorganizing loops and assignments in order to maximize the available parallelism

While both would be possible as the translator works with a tree structure, we simply have not started to implement much in this direction yet.

  • GPUFORT does not implement the full OpenACC standard (yet)

GPUFORT was developed to translate a number of HPC apps to code formats that are well supported by AMD's ROCm ecosystem. The development of GPUFORT is steered by the requirements of these applications.

Fortran-C Interoperablity Limitations

To interface generated HIP C++ kernels with the original Fortran code, GPUFORT relies on the iso_c_binding interoperability mechanisms that were added to the Fortran language with the Fortran 2003 standard. Please be aware that the interoperability of C structs and Fortran derived types is quite limited till this date:

  • "Derived types with the C binding attribute shall not have the sequence attribute, type parameters, the extends attribute, nor type-bound procedures."
  • "Every component must be of interoperable type and kind and may not have the pointer or allocatable attribute. The names of the components are irrelevant for interoperability."

(Source: https://gcc.gnu.org/onlinedocs/gfortran/Derived-Types-and-struct.html)

We are currently investigating what workarounds could be automatically applied. Until then, you have to modify your code manually to circumvent the above limitations.

Currently supported features:

  • ACC:
    • ACC2OMP & ACC2HIP
    • Translation of data directives: !$acc enter data, !$acc exit data, !$acc data
    • Synchronization directives: !$acc wait, !$acc update self/host/device
    • Kernel and loop constructs !$acc kernels plus !$acc loop in subsequent line, !$acc kernels loop, !$acc parallel plus !$acc loop in subsequent line, !$acc parallel loop, !$acc loop
    • Support for !$acc routine seq functions with scalar arguments
  • CUF:
    • CUF2HIP
      • Majority of CUDA libary functionality via HIPFORT
      • Kernel and loop constructs: !$cuf kernel do
      • Overloaded intrinsics: allocate, allocated, deallocate, deallocated, =
      • Support for CUDA Fortran attributes(global) (array and scalar arguments), and attributes(host,device), attributes(device) procedures (only scalar arguments supported for the latter)

(List is not complete ...)

Planned features (or: "more limitations")

  • Current work focuses on:
    • ACC:
      • Initial support for !$acc declare (detected but not considered in codegen yet)
      • Improve support for!$acc parallel (loop)
      • Add support for !$acc parallel without !$acc loop in next line)
        • Results in gang parallelism
      • Add support for !$acc kernels without !$acc loop in next line)
        • Auto detection of offloadable code parts
      • Rewrite of GPUFORT Fortran runtime in (HIP) C++
    • ACC/CUF:
      • Support of derived types with allocatable, pointer members
  • Planned:
    • Add option for prescribing syntax checker (e.g. use other compiler for syntax checks.)

More Repositories

1

ROCm

AMD ROCmâ„¢ Software - GitHub Home
Shell
4,583
star
2

HIP

HIP: C++ Heterogeneous-Compute Interface for Portability
C++
3,398
star
3

MIOpen

AMD's Machine Intelligence Library
Assembly
1,060
star
4

HIPIFY

HIPIFY: Convert CUDA to Portable C++ Code
C++
505
star
5

hcc

HCC is an Open Source, Optimizing C++ Compiler for Heterogeneous Compute currently for the ROCm GPU Computing Platform
C++
425
star
6

rocBLAS

Next generation BLAS implementation for ROCm platform
C++
308
star
7

composable_kernel

Composable Kernel: Performance Portable Programming Model for Machine Learning Tensor Operators
C++
285
star
8

omnitrace

Omnitrace: Application Profiling, Tracing, and Analysis
C++
283
star
9

rccl

ROCm Communication Collectives Library (RCCL)
C++
231
star
10

ROCR-Runtime

ROCm Platform Runtime: ROCr a HPC market enhanced HSA based runtime
C++
217
star
11

Tensile

Stretching GPU performance for GEMMs and tensor contractions.
Python
214
star
12

aomp

AOMP is an open source Clang/LLVM based compiler with added support for the OpenMP® API on Radeon™ GPUs. Use this repository for releases, issues, documentation, packaging, and examples.
Fortran
203
star
13

AMDMIGraphX

AMD's graph optimization engine.
C++
185
star
14

rocFFT

Next generation FFT implementation for ROCm
C++
174
star
15

MIVisionX

MIVisionX toolkit is a set of comprehensive computer vision and machine intelligence libraries, utilities, and applications bundled into a single toolkit. AMD MIVisionX also delivers a highly optimized open-source implementation of the Khronos OpenVXâ„¢ and OpenVXâ„¢ Extensions.
C++
168
star
16

rocPRIM

ROCm Parallel Primitives
C++
157
star
17

rocm-examples

A collection of examples for the ROCm software stack
C++
154
star
18

omniperf

Advanced Profiling and Analytics for AMD Hardware
Python
132
star
19

rocprofiler

ROC profiler library. Profiling with perf-counters and derived metrics.
C
126
star
20

rocMLIR

C++
120
star
21

rocSPARSE

Next generation SPARSE implementation for ROCm platform
C++
117
star
22

rocm_smi_lib

ROCm SMI LIB
C++
116
star
23

rocRAND

RAND library for HIP programming language
C++
110
star
24

HIP-CPU

An implementation of HIP that works on CPUs, across OSes.
C++
107
star
25

rocThrust

ROCm Thrust - run Thrust dependent software on AMD GPUs
C++
100
star
26

ROCm-Device-Libs

ROCm Device Libraries
C
97
star
27

rocSOLVER

Next generation LAPACK implementation for ROCm platform
C++
91
star
28

rocWMMA

rocWMMA
C++
86
star
29

hipCUB

Reusable software components for ROCm developers
C++
81
star
30

rocALUTION

Next generation library for iterative sparse solvers for ROCm platform
C++
74
star
31

hipfort

Fortran interfaces for ROCm libraries
Fortran
69
star
32

roctracer

ROCm Tracer Callback/Activity Library for Performance tracing AMD GPUs
C++
69
star
33

hipSPARSE

ROCm SPARSE marshalling library
C++
67
star
34

atmi

Asynchronous Task and Memory Interface, or ATMI, is a runtime framework and programming model for heterogeneous CPU-GPU systems. It provides a consistent, declarative API to create task graphs on CPUs and GPUs (integrated and discrete).
C++
66
star
35

ROCmValidationSuite

The ROCm Validation Suite is a system administrator’s and cluster manager's tool for detecting and troubleshooting common problems affecting AMD GPU(s) running in a high-performance computing environment, enabled using the ROCm software stack on a compatible platform.
C++
61
star
36

rocm-cmake

CMake modules used within the ROCm libraries
CMake
59
star
37

hipFFT

hipFFT is a FFT marshalling library.
C++
52
star
38

ROCgdb

This is ROCgdb, the ROCm source-level debugger for Linux, based on GDB, the GNU source-level debugger.
C
50
star
39

amd_matrix_instruction_calculator

A tool for generating information about the matrix multiplication instructions in AMD Radeonâ„¢ and AMD Instinctâ„¢ accelerators
Python
48
star
40

ROCm-CompilerSupport

The compiler support repository provides various Lightning Compiler related services.
C++
46
star
41

rpp

AMD ROCm Performance Primitives (RPP) library is a comprehensive high-performance computer vision library for AMD processors with HIP/OpenCL/CPU back-ends.
C++
46
star
42

ROCclr

44
star
43

rocm_bandwidth_test

Bandwidth test for ROCm
C++
41
star
44

amdsmi

AMD SMI
C++
39
star
45

HIPCC

HIPCC: HIP compiler driver
C++
39
star
46

aotriton

Ahead of Time (AOT) Triton Math Library
Python
37
star
47

Experimental_ROC

Experimental and Intriguing Tools for ROCm
Shell
35
star
48

rocHPCG

HPCG benchmark based on ROCm platform
C++
35
star
49

ROC_SHMEM

ROC_SHMEM intra-kernel networking runtime for AMD dGPUs on the ROCm platform.
C++
34
star
50

MISA

Machine Intelligence Shader Autogen. AMDGPU ML shader code generator. (previously iGEMMgen)
Python
34
star
51

ROCm.github.io

ROCm Website
32
star
52

TransferBench

TransferBench is a utility capable of benchmarking simultaneous copies between user-specified devices (CPUs/GPUs)
C++
29
star
53

rocm-blogs

Jupyter Notebook
28
star
54

clang-ocl

OpenCL compilation with clang compiler.
CMake
26
star
55

hipSOLVER

ROCm SOLVER marshalling library
C++
24
star
56

ROCm-OpenCL-Driver

ROCm OpenCL Compiler Tool Driver
C++
24
star
57

rdc

RDC
C++
23
star
58

hipRAND

Random number library that generate pseudo-random and quasi-random numbers.
C++
23
star
59

rccl-tests

RCCL Performance Benchmark Tests
Cuda
21
star
60

ROCdbgapi

The AMD Debugger API is a library that provides all the support necessary for a debugger and other tools to perform low level control of the execution and inspection of execution state of AMD's commercially available GPU architectures.
C++
19
star
61

pyrsmi

python package of rocm-smi-lib
Python
18
star
62

hip-python

HIP Python Low-level Bindings
Shell
17
star
63

hip-tests

C++
15
star
64

roc-stdpar

C++
14
star
65

pytorch-micro-benchmarking

Python
14
star
66

hipify_torch

Python
13
star
67

rocmProfileData

C++
13
star
68

rocm-docs-core

ROCm Documentation Python package for ReadTheDocs build standardization
CSS
12
star
69

rocAL

The AMD rocAL is designed to efficiently decode and process images and videos from a variety of storage formats and modify them through a processing graph programmable by the user.
C++
11
star
70

half

C++
9
star
71

rocprofiler-sdk

C++
9
star
72

rocBLAS-Examples

Examples illustrating usage of the rocBLAS library
C++
9
star
73

OSU_Microbenchmarks

ROCm - UCX enabled OSU_Benchmarks
C
8
star
74

MITuna

Python
7
star
75

rtg_tracer

C++
7
star
76

Gromacs

ROCm's implementation of Gromacs
C++
6
star
77

rocm-spack-pkgs

Repository to host spack recipes for ROCm
Python
6
star
78

rbuild

Rocm build tool
Python
6
star
79

rocm-core

CMake
5
star
80

rocm-llvm-python

Low-level Cython and Python bindings to the (ROCm) LLVM and AMD COMGR C API. Also ships the official LLVM Clang bindings.
Shell
4
star
81

hip-testsuite

Python
4
star
82

MIFin

Tuna centric MIOpen client
C++
4
star
83

flang

Mirror of flang repo: The source repo is https://github.com/flang-compiler/flang . Once a day the master branch is updated from the upstream source repo and then locked. AOMP or ROCm developers may commit or create PRs on branch aomp-dev.
C++
3
star
84

numba-hip

HIP backend patch for Numba, the NumPy aware dynamic Python compiler using LLVM.
Python
3
star
85

tensorcast

Python
3
star
86

hipSPARSELt

C++
2
star
87

aomp-extras

hostcall services library, math library, and utilities
Shell
2
star
88

MIOpenExamples

MIOpen examples
C++
2
star
89

rocprofiler-register

CMake
2
star
90

rocm-install-on-windows

2
star
91

hipOMB

OSU MPI benchmarks with ROCm support
C
1
star
92

migraphx-benchmark

1
star
93

rocm-recipes

Recipes for rocm
CMake
1
star
94

hipBLAS-common

Common files shared by hipBLAS and hipBLASLt
CMake
1
star