OpenACC: Difference between revisions

From Alliance Doc
Jump to navigation Jump to search
(Tabs and file format for easy download)
 
(18 intermediate revisions by 2 users not shown)
Line 1: Line 1:
{{draft}}
{{draft}}


OpenACC makes it relatively easy to offload vectorized code to accelerators (like GPUs, for example). Unlike [[CUDA]] and OpenCL where kernels need to be coded explicitly, OpenACC minimizes the amount of modifications to do on a serial or [[OpenMP]] code. The compiler converts the OpenACC code into a binary executable that can make use of accelerators. The performance of OpenACC codes can be similar to the one of a [[CUDA]] code, except that OpenACC requires less code development.
OpenACC makes it relatively easy to offload vectorized code to accelerators such as GPUs, for example. Unlike [[CUDA]] and OpenCL where kernels need to be coded explicitly, OpenACC minimizes the amount of modifications to do on a serial or [[OpenMP]] code. The compiler converts the OpenACC code into a binary executable that can make use of accelerators. The performance of OpenACC codes can be similar to the one of a [[CUDA]] code, except that OpenACC requires less code development.


= OpenACC compiler instructions =
= OpenACC directives =
Similar to [[OpenMP]], OpenACC can convert a for-loop into parallel code that would run on an accelerator. This can be achieved with compiler instructions <code>#pragma acc ...</code> before a for-loop.
Similar to [[OpenMP]], OpenACC can convert a <code>for</code> loop into parallel code that would run on an accelerator. This can be achieved with compiler directives <code>#pragma acc ...</code> before structured blocks of code like, for example, a <code>for</code> loop. All supported <code>pragma</code> directives are described in the [https://www.openacc.org/specification OpenACC specification].
 
== <code>pragma acc parallel</code> ==
This will start the following structured block of code in parallel execution on the accelerator.
 
Optional clauses:
* <code>vector_length(size)</code>: determines the vector length to use for vector or SIMD operations.
** For example, <code>size</code> can be: 256, 512
 
== <code>pragma acc loop</code> ==
The following for-loop will be converted for an execution on the accelerator.
 
Optional clauses:
* <code>reduction(op:variable)</code>: the loop is doing a reduction in a variable defined outside of the loop. For each thread in parallel, a private version of that variable will be used.
** Typical operators: <code>+</code> (sum), <code>*</code> (product)


= Code examples =
= Code examples =
OpenACC can be used in FORTRAN, C and C++.
OpenACC can be used in [[Fortran]], [[C]] and [[C++]], which we illustrate here using a simple program that computes a decimal approximation to π based on a definite integral which is equal to arctan(1), i.e. π/4.
 
<tabs>
<tabs>
<tab name="C">
<tab name="C">
The following example computes an approximation of Pi:
{{File
{{File
   |name=pi.c
   |name=pi.c
Line 32: Line 16:
#include <stdio.h>
#include <stdio.h>


#define N 2000000000
const int vl = 512;
#define vl 512
const long long N = 2000000000;


int main(void) {
int main(int argc,char** argv) {
   double pi = 0.0f;
   double pi = 0.0f;
   long long i;
   long long i;
Line 46: Line 30:
   }
   }
   printf("pi = %11.10f\n", pi / N);
   printf("pi = %11.10f\n", pi / N);
  return 0;
}
}}
</tab>
<tab name="C++">
{{File
  |name=pi.cxx
  |lang="C++"
  |contents=
#include <iostream>
#include <iomanip>
const int vl = 512;
const long long N = 2000000000;
int main(int argc,char** argv) {
  double pi = 0.0f;
  long long i;
  #pragma acc parallel vector_length(vl)
  #pragma acc loop reduction(+:pi)
  for (i = 0; i < N; i++) {
    double t = double((i + 0.5) / N);
    pi += 4.0/(1.0 + t * t);
  }
  std::cout << std::fixed;
  std::cout << std::setprecision(10);
  std::cout << "pi = " << pi/double(N) << std::endl;
   return 0;
   return 0;
}
}

Latest revision as of 17:54, 30 November 2020


This article is a draft

This is not a complete article: This is a draft, a work in progress that is intended to be published into an article, which may or may not be ready for inclusion in the main wiki. It should not necessarily be considered factual or authoritative.




OpenACC makes it relatively easy to offload vectorized code to accelerators such as GPUs, for example. Unlike CUDA and OpenCL where kernels need to be coded explicitly, OpenACC minimizes the amount of modifications to do on a serial or OpenMP code. The compiler converts the OpenACC code into a binary executable that can make use of accelerators. The performance of OpenACC codes can be similar to the one of a CUDA code, except that OpenACC requires less code development.

OpenACC directives[edit]

Similar to OpenMP, OpenACC can convert a for loop into parallel code that would run on an accelerator. This can be achieved with compiler directives #pragma acc ... before structured blocks of code like, for example, a for loop. All supported pragma directives are described in the OpenACC specification.

Code examples[edit]

OpenACC can be used in Fortran, C and C++, which we illustrate here using a simple program that computes a decimal approximation to π based on a definite integral which is equal to arctan(1), i.e. π/4.

File : pi.c

#include <stdio.h>

const int vl = 512;
const long long N = 2000000000;

int main(int argc,char** argv) {
  double pi = 0.0f;
  long long i;

  #pragma acc parallel vector_length(vl) 
  #pragma acc loop reduction(+:pi)
  for (i = 0; i < N; i++) {
    double t = (double)((i + 0.5) / N);
    pi += 4.0 / (1.0 + t * t);
  }
  printf("pi = %11.10f\n", pi / N);
  return 0;
}


File : pi.cxx

#include <iostream>
#include <iomanip>

const int vl = 512;
const long long N = 2000000000;

int main(int argc,char** argv) {
  double pi = 0.0f;
  long long i;

  #pragma acc parallel vector_length(vl)
  #pragma acc loop reduction(+:pi)
  for (i = 0; i < N; i++) {
    double t = double((i + 0.5) / N);
    pi += 4.0/(1.0 + t * t);
  }
  std::cout << std::fixed;
  std::cout << std::setprecision(10);
  std::cout << "pi = " << pi/double(N) << std::endl;
  return 0;
}


Compilers[edit]

PGI[edit]

  • Module pgi, any version from 13.10
    • Newer versions support newest GPU capabilities.

Compilation example:

# TODO

GCC[edit]

  • Module gcc, any version from 9.3.0
    • Newer versions support newest GPU capabilities.

Compilation example:

gcc -fopenacc -march=native -O3 pi.c -o pi

Tutorial[edit]

See our OpenACC_Tutorial.

References[edit]