OpenMP Accelerator Model Getting Started

Environment Variables

This OpenMP Accelerator Model runtime implementation utilizes TI's OpenCL Runtime. It also relies on TI's C6000 Code Generation Tools version 8.0.x. In order to ensure that OpenCL is set up correctly, build and run any OpenCL example following the instructions specified in the OpenCL User's Guide. Set the following environment variables to the correct paths:

Environment Variable	Value
TI_OCL_CGT_INSTALL	Location of the TI CGT DSP compiler tools. Only required if the tools are not installed in the default location specified by the installer.
TARGET_ROOTDIR	Location of the K2H file-system if cross-compiling on x86/x86_64

If cross-compiling, source the ompacc_env.sh environment setup script provided in $TARGET_ROOTDIR/usr/share/ti/examples/openmpacc/

clacc

clacc is a tool used to compile OpenMP 4.0 accelerator model source code. clacc in turn relies on the host gcc toolchain, C6000 Code Generation Tools (CGT) and the TI OpenMP Accelerator Model and OpenCL host libraries.

clacc wrapper compiler for OpenMP Accelerator Model compilation

Building a program using clacc is accomplished with the command:

clacc [options] [object-files] C-source-files

A summary of clacc options are specified below. On Ubuntu based K2H systems, a man page is available.

Option	Description
-h [ --help ]	Show help message
-v [ --verbose ]	Show debug output during clacc compilation
-k [ --keep_files ]	Keep intermediate temporary files during clacc compilation
-d [ --runtime_verbose ]	Enable runtime debugging information display. Setting this enables detailed runtime debug information printing from the accelerator model ARM runtime library.
-p [ --runtime_performance ]	Enable runtime performance information display. Setting this displays time taken to execute OpenMP target regions and time taken to transfer data to and from the accelerator device at runtime.
-g [ --debug ]	Generate target debug symbols that may be used with gdb-c6x for runtime debugging.
--make_lib	This is used to create a CLACC static library which packages the ARM and DSP object files and required dependencies. This library may only be used with CLACC and cannot be passed to the host or accelerator device side compiler directly.
--host_cc <arg>	Host compiler to use. Default is arm-linux-gnueabihf-g++.
--hc <arg>	Host compiler options to be passed to the final compile and link stage. E.g. --hc="-O3 -Wall -Wextra -fopenmp"
--hc <arg>	Target compiler options to be passed to the accelerator device compile stage. E.g. --tc="-O3 -k -symdebug:none"
--hl arg	Host linker options to be passed to the final link stage. E.g. --hl="-lrt -L/path/to/atlas/library/install/for/ARM -lptcblas -latlas"
--tl arg	Target linker options to be passed to the accelerator model device link stage. E.g. --tl="-L/path/to/static/library/for/C66x/DSP -ldsplib.ae66"
--show_cc_opts	Show host and target compiler options being used.
-o [ --exe ] <arg>	Name of final executable fat binary to be created. In case of library mode, this is the name of the CLACC library to be generated.

Example Applications

This product is shipped with a number of examples that demonstrate the use of OpenMP constructs and the TI-specific extensions. The examples typically have the following source directory file structure:

• Host C/C++ file(s)
  • File with main() function are named <example_name>_main.cpp/c
  • Files with host equivalents of functions used within target regions are named <example_name>_host.cpp/c
• Target C file(s)
  • Files with #pragma omp target and #pragma omp declare target regions are named <example_name>_target.c
• Makefile
  • Includes a top level make.inc file

At present target regions are only supported in C source files. However, host code may be written in C or C++ and are handled accordingly in the Makefile.

The Makefiles used in the examples have OA_SHELL_OPTS set to -v -k by default, which shows debugging output and does not remove temporary files that are created respectively. After compiling an example the directory may be populated with temporary files: *.out __TI_CLACC_KERNEL.c *.asm *.cl *.dsp_h *.bc *.objc *.if *.map *.opt *.int.c *.o *.obj. Please remove -k option to disable keeping temporary files.

In order to enable display of runtime DEBUG information, add -d flag to OA_SHELL_OPTS

In order to enable display of runtime PERFORMANCE information add -p flag to OA_SHELL_OPTS

To compile an example:

• $ cd $TARGET_ROOTDIR/usr/share/ti/examples/openmpacc/
• $ cd vecadd
• $ make

To run an example on a TI 66AK2H platform:

• Ensure the OpenCL package is installed and an OpenCL example has been built and tested
• Ensure the OpenMPAcc package is installed
• Ensure the compiled example executable is present on platform file-system
• Run executable

See OpenMP Accelerator Model Examples for brief descriptions of the examples.

Known Issues & Limitations (v1.1.1)

• OpenMP 4.0 device constructs may only be used with C code.
• Only one target device is supported. The target device is a collection of 8 Texas Instruments' C66x DSP cores.
• Only [lower-bound : length ] syntax for specifying array section is currently supported. See the OpenMP Specification Section 2.4 for more information on array sections.
• The implementation does not support using a target update construct to synchronize variables that are mapped to a device using a declare target construct. However, variables that are mapped to a device using the target data construct may be synchronized using the target update construct.
• When mapping array sections, sections specified in map clauses of an enclosing target data must be replicated in the enclosed target construct for correct behavior.
• OpenMP pragmas may not immediately precede declaration statements. This especially applies to OpenMP point pragmas. For example:

<source lang="cpp" strict enclose="div">

  #pragma omp target update from(c[0:size])
  int a = 0;

</source>

• The sum of sizes of all mapped array sections for a single target region must not exceed 1.5 Gb. This is a target DSP device restriction and cannot be surpassed. More information is available here.
• Concurrent target regions are not supported. If host code is multi-threaded, it must be ensured that only one target region is active across the threads. For example, in an OpenMP parallel region, place the target region code in a critical section.
• Any GCC pragmas or attributes in host code compiled using CLACC are not retained.
• Creating dynamic libraries with target regions using CLACC is not supported.
• Avoid placing double precision floating point constants in host source code compiled using CLACC - a loss of precision could occur during source-to-source translation performed by CLACC.

Migration Guide

Version 0.3.3 to version 1.1.0

• Modified OMP device construct #pragma omp declare target end to #pragma omp end declare target to be consistent with the OpenMP specification, v4.0.
• The names and interfaces for the functions used for dynamic memory management from shared memory have changed.
  • __TI_omp_device_alloc() has been replaced with __malloc_ddr() or __malloc_msmc()
  • __TI_omp_device_free() has been replaced with __free_ddr() or __free_msmc()

• Changes have been made to CLACC command options. The changes and additions to the options are listed below. Please refer to the section on clacc for more details on the updated CLACC options.
  • --lib/-l used to specify host linker options has been replaced with --hl
  • --host_cc_opts used to specify host compiler options has been replaced with --hc
  • --runtime_debug used to enable runtime debug information has been replaced with --runtime_verbose
  • --target_cc_opts used to specify host compiler options has been replaced with --tc
  • --lib_path and --inc_path are no longer supported. Compiler and and linker include paths may be specified using --lib and --hc/--tc.
  • New option --debug/-g has been added and may be used to generate target debug symbols
  • New option --make_lib has been added and may be used to create a CLACC static library
  • New option --tl has been added and may be used to specify target linker options