AI/rocm-systems
2
0
Forkuj 0
Kod Zgłoszenia Oczekujące zmiany Actions Packages Projekty Wydania Wiki Aktywność

Update documentation.

Przeglądaj źródła 
- Add more detailed feature comparison HIP vs specific CUDA SDK revs.
- Remove reference to old codenames.
- Fix TOC in some md files.

Change-Id: If0f6a05dc61237c4710c4d4bf2d933d556dcaeae


[ROCm/clr commit: 4b3102b578]
This commit is contained in:
Ben Sander
2016-07-18 13:56:13 -05:00
rodzic f58a6b4dd5
commit 5586cc02b4
6 zmienionych plików z 132 dodań i 142 usunięć
Ściągnij plik aktualizacji Ściągnij plik porównania Expand all files Collapse all files
projects/clr/hipamd/README.md
+2 -3
Wyświetl plik
@@ -12,16 +12,15 @@ Key features include:
New projects can be developed directly in the portable HIP C++ language and can run on either NVIDIA or AMD platforms. Additionally, HIP provides porting tools which make it easy to port existing CUDA codes to the HIP layer, with no loss of performance as compared to the original CUDA application. HIP is not intended to be a drop-in replacement for CUDA, and developers should expect to do some manual coding and performance tuning work to complete the port.
## More Info:
- [Installation](INSTALL.md) and [clang-hipify](clang-hipify/README.md)
- [Installation](INSTALL.md)
- [HIP FAQ](docs/markdown/hip_faq.md)
- [HIP Kernel Language](docs/markdown/hip_kernel_language.md)
- [HIP Runtime API (Doxygen)](http://gpuopen-professionalcompute-tools.github.io/HIP)
- [HIP Porting Guide](docs/markdown/hip_porting_guide.md)
- [HIP Terminology](docs/markdown/hip_terms.md) (including Rosetta Stone of GPU computing terms across CUDA/HIP/HC/AMP/OpenL)
- [clang-hipify](clang-hipify/README.md)
- [Developer/CONTRIBUTING Info](CONTRIBUTING.md)
- [Release Notes](RELEASE.md)
## How do I get set up?
See the [Installation](INSTALL.md) notes.
projects/clr/hipamd/RELEASE.md
+4 -5
Wyświetl plik
@@ -1,11 +1,10 @@
# Release notes
We have attempted to document known bugs and limitations - in particular the [HIP Kernel Language](docs/markdown/hip_kernel_language.md) document uses the phrase "Under Development", and the [HIP Runtime API bug list](http://gpuopen-professionalcompute-tools.github.io/HIP/bug.html) lists known bugs. Some of the key items we are working on:
- Tuning built-in functions, including shfl.
- Performance optimization.
We have attempted to document known bugs and limitations - in particular the [HIP Kernel Language](docs/markdown/hip_kernel_language.md) document uses the phrase "Under Development", and the [HIP Runtime API bug list](http://gpuopen-professionalcompute-tools.github.io/HIP/bug.html) lists known bugs.
Stay tuned - the work for many of these features is already in-flight.
===================================================================================================
Upcoming:
- hipLaunchKernel supports one-dimensional grid and/or block dims, without explicit cast to dim3 type.
===================================================================================================
Release:0.90.00
projects/clr/hipamd/docs/doxygen-input/mainpage.txt
+3 -3
Wyświetl plik
@@ -7,11 +7,11 @@
* The HIP interface makes it very easy to port existing CUDA apps to run on AMD GPUs,
* or to develop new apps that can run on either CUDA or AMD GPUs from a common source base.
*
* - HIP is very thin and has little or no performance impact over coding directly in CUDA NVCC or Kalmar HC mode.
* - HIP is very thin and has little or no performance impact over coding directly in CUDA NVCC or HCC HC mode.
* - HIP allows developers to use the "best" development environment and tools on each target platform.
* - HIP allows coding in a single-source C++ programming language including features such as templates, C++11 lambdas,and more.
* - "hipify" script automatically converts source from CUDA to HIP.
* - Developers can specialize for CUDA or Kalmar to tune for performance or handle tricky cases with #ifdef.
* - "hipify" tool automatically converts source from CUDA to HIP.
* - Developers can specialize for CUDA or HCC to tune for performance or handle tricky cases with #ifdef.
* - See the @ref API.
projects/clr/hipamd/docs/markdown/hip_faq.md
+65 -25
Wyświetl plik
@@ -1,13 +1,12 @@
# FAQ
<!-- START doctoc generated TOC please keep comment here to allow auto update -->
<!-- DON'T EDIT THIS SECTION, INSTEAD RE-RUN doctoc TO UPDATE -->
**Table of Contents** *generated with [DocToc](https://github.com/thlorenz/doctoc)*
- [What APIs does HIP support ?](#what-apis-does-hip-support-)
- [What is not supported?](#what-is-not-supported)
- [What APIs and features does HIP support ?](#Q1)
- [What is not supported?](#Q2)
- [Run-time features](#run-time-features)
- [Kernel language features](#kernel-language-features)
- [Is HIP a drop-in replacement for CUDA?](#Q3)
- [What version of CUDA is supported?](#Q4)
- [What libraries does HIP support?](#Q5)
- [How does HIP compare with OpenCL?](#how-does-hip-compare-with-opencl)
- [What hardware does HIP support?](#what-hardware-does-hip-support)
- [Does Hipify automatically convert all source code?](#does-hipify-automatically-convert-all-source-code)
@@ -20,38 +19,79 @@
- [What's the difference between HIP and hc?](#whats-the-difference-between-hip-and-hc)
- [HIP detected my platform (hcc vs nvcc) incorrectly - what should I do?](#hip-detected-my-platform-hcc-vs-nvcc-incorrectly---what-should-i-do)
<!-- END doctoc generated TOC please keep comment here to allow auto update -->
=================
### What APIs does HIP support ?
### What APIs and features does HIP support?<a name="Q1"></a>
HIP provides the following:
- Devices (hipSetDevice(), hipGetDeviceProperties(), etc)
- Memory management (hipMalloc(), hipMemcpy(), hipFree())
- Streams (hipStreamCreate(), etc.)---under development
- Devices (hipSetDevice(), hipGetDeviceProperties(), etc.)
- Memory management (hipMalloc(), hipMemcpy(), hipFree(), etc.)
- Streams (hipStreamCreate(), etc.)
- Events (hipEventRecord(), hipEventElapsedTime(), etc.)
- Kernel launching (hipLaunchKernel is a standard C/C++ function that replaces <<< >>>)
- CUDA-style kernel indexing
- Device-side math built-ins
- CUDA-style kernel coordinate functions (threadIdx, blockIdx, blockDim, gridDim)
- Most device-side math built-ins
- Error reporting (hipGetLastError(), hipGetErrorString())
The HIP documentation describes each API and its limitations, if any, compared with the equivalent CUDA API.
The HIP API documentation describes each API and its limitations, if any, compared with the equivalent CUDA API.
### What is not supported?
### What is not supported?<a name="Q2"></a>
#### Run-time features
- Textures
- Dynamic parallelism
- Managed memory
- MemcpyToSymbol functions
- Dynamic parallelism (CUDA 5.0)
- Managed memory (CUDA 6.5)
- Graphics interoperation with OpenGL or Direct3D
- CUDA array, mipmappedArray and pitched memory
- CUDA Driver API
#### Kernel language features
- Device-side dynamic memory allocations (malloc, free, new, delete)
- Virtual functions, indirect functions and try/catch
- Device-side dynamic memory allocations (malloc, free, new, delete) (CUDA 4.0)
- Virtual functions, indirect functions and try/catch (CUDA 4.0)
- `__prof_trigger`
- PTX assembly
- See the [HIP Kernel Language](hip_kernel_language.md) for more information.
- PTX assembly (CUDA 4.0)
- Several kernel features are under development. See the [HIP Kernel Language](hip_kernel_language.md) for more information.
### Is HIP a drop-in replacement for CUDA?<a name="Q3"></a>
No. HIP provides porting tools which do most of the work do convert CUDA code into portable C++ code that uses the HIP APIs.
Most developers will port their code from CUDA to HIP and then maintain the HIP version.
HIP code provides the same performance as coding in native CUDA, plus the benefit that the code can also run on AMD platforms.
### What version of CUDA is supported?<a name="Q4"></a>
HIP APIs and features do not map to a specific CUDA version. HIP provides a strong subset of functionality provided in CUDA, and the hipify tools can
scan code to identify any unsupported CUDA functions - this is very useful for identifying the specific features required by a given application.
However, we can provide a rough summary of the features included in each CUDA SDK and the support level in HIP:
- CUDA 4.0 and earlier :
- HIP supports CUDA 4.0 except for the limitations described [above](#Q2).
- CUDA 5.0 :
- Dynamic Parallelism (not supported)
- cuIpc functions (under development).
- CUDA 5.5 :
- CUPTI (not directly supported, [AMD GPUPerfAPI](http://developer.amd.com/tools-and-sdks/graphics-development/gpuperfapi/) can be used as an alternative in some cases)
- CUDA 6.0
- Managed memory (under development)
- CUDA 6.5
- __shfl instriniscs (supported)
- CUDA 7.0
- Per-thread-streams (under development)
- C++11 (HCC supports all of C++11, all of C++14 and some C++17 features)
- CUDA 7.5
- float16 (under development)
- CUDA 8.0
- No new language features.
### What libraries does HIP support? <a name="Q5"></a>
HIP includes growing support for the 4 key math libraries using hcBlas, hcFft, hcrng, and hcsparse).
These offer pointer-based memory interfaces (as opposed to opaque buffers) and can be easily interfaces with other HCC code. Developers should use conditional compliation if portability to nvcc systems is desired - using calls to cu* routines on one path and hc* routines on the other.
- [hcblas](https://bitbucket.org/multicoreware/hcblas)
- [hcfft](https://bitbucket.org/multicoreware/hcfft)
- [hcsparse](https://bitbucket.org/multicoreware/hcsparse)
- [hcrng](https://bitbucket.org/multicoreware/hcrng)
Additionally, some of the cublas routines are automatically converted to hipblas equivalents by the clang-hipify tool. These APIs use cublas or hcblas depending on the platform, and replace the need
to use conditional compilation.
### How does HIP compare with OpenCL?
Both AMD and Nvidia support OpenCL 1.2 on their devices, so developers can write portable code.
@@ -65,7 +105,7 @@ HIP offers several benefits over OpenCL:
- HIP offers an offline compilation model.
### What hardware does HIP support?
- For AMD platforms, HIP runs on the same hardware that the HCC "hc" mode supports - specifically AMD Kaveri, Carrizo and Fiji.
- For AMD platforms, HIP runs on the same hardware that the HCC "hc" mode supports. See the ROCM documentation for the list of supported platforms.
- For Nvidia platforms, HIP requires Unified Memory and should run on a device which runs the CUDA SDK 6.0 or newer. We have tested the Nvidia Titan and K40.
### Does Hipify automatically convert all source code?
@@ -79,7 +119,7 @@ Developers should always expect to perform some platform-specific tuning and opt
NVCC is Nvidia's compiler driver for compiling "CUDA C++" code into PTX or device code for Nvidia GPUs. It's a closed-source binary product that comes with CUDA SDKs.
### What is HCC?
HCC is AMD's compiler driver which compiles "heterogenous C++" code into HSAIL or GCN device code for AMD GPUs. HCC is an open-source compiler based on recent versions of CLANG/LLVM.
HCC is AMD's compiler driver which compiles "heterogenous C++" code into HSAIL or GCN device code for AMD GPUs. It's an open-source compiler based on recent versions of CLANG/LLVM.
### Why use HIP rather than supporting CUDA directly?
While HIP is a strong subset of the CUDA, it is a subset. The HIP layer allows that subset to be clearly defined and documented.
projects/clr/hipamd/docs/markdown/hip_kernel_language.md
+15 -14
Wyświetl plik
@@ -1,6 +1,4 @@
<!-- START doctoc generated TOC please keep comment here to allow auto update -->
<!-- DON'T EDIT THIS SECTION, INSTEAD RE-RUN doctoc TO UPDATE -->
**Table of Contents** *generated with [DocToc](https://github.com/thlorenz/doctoc)*
**Table of Contents**
- [Introduction](#introduction)
- [Function-Type Qualifiers](#function-type-qualifiers)
@@ -45,8 +43,6 @@
- [In-Line Assembly](#in-line-assembly)
- [C++ Support](#c-support)
<!-- END doctoc generated TOC please keep comment here to allow auto update -->
## Introduction
@@ -98,9 +94,9 @@ HIP parses the `__noinline__` and `__forceinline__` keywords and converts them t
`__global__` functions are often referred to as *kernels,* and calling one is termed *launching the kernel.* These functions require the caller to specify an "execution configuration" that includes the grid and block dimensions. The execution configuration can also include other information for the launch, such as the amount of additional shared memory to allocate and the stream where the kernel should execute. HIP introduces a standard C++ calling convention to pass the execution configuration to the kernel (this convention replaces the Cuda <<< >>> syntax). In HIP,
- Kernels launch with the "hipLaunchKernel" function
- The first five parameters to hipLaunchKernel are the following:
- **symbol kernelName**: the name of the kernel to launch
- **dim3 gridDim**: 3D-grid dimensions
- **dim3 blockDim**: 3D-block dimensions
- **symbol kernelName**: the name of the kernel to launch. To support template kernels which contains "," use the HIP_KERNEL_NAME macro. The hipify tools insert this automatically.
- **dim3 gridDim**: 3D-grid dimensions specifying the number of blocks to launch.
- **dim3 blockDim**: 3D-block dimensions specifying the number of threads in each block.
- **size_t dynamicShared**: amount of additional shared memory to allocate when launching the kernel (see [__shared__](#__shared__))
- **hipStream_t**: stream where the kernel should execute. A value of 0 corresponds to the NULL stream (see [Synchronization Functions](#synchronization-functions)).
- Kernel arguments follow these first five parameters
@@ -629,6 +625,9 @@ The compiler ensures that the kernel uses fewer registers than both allowed maxi
HIP/hcc will parse the `launch_bounds` attribute but silently ignores the performance hint. Full support is under development.
The hcc compiler does not support the "--maxregcount" option like nvcc. Instead, users are encouraged to use the hip_launch_bounds directive since the parameters are more intuitive and portable than
micro-architecture details like registers, and also the directive allows per-kernel control rather than an entire file. hip_launch_bounds works on both hcc and nvcc targets.
## Register Keyword
The register keyword affects code generation in neither nvcc nor hcc. It’s deprecated in standard C++, so hcc will generate a warning. (nvcc silently ignores use of this keyword.) To disable the warning, you can pass the option `-Wno-deprecated-register` to hcc.
@@ -636,23 +635,25 @@ The register keyword affects code generation in neither nvcc nor hcc. It’s de
## Pragma Unroll
hcc support for the unroll pragma is under development and is slated to arrive with the Lightning Compiler.
Unroll with a bounds that is known at compile-time is supported. For example:
```
#pragma unroll 16 /* hint to compiler to unroll next loop by 16 */
for (int i=0; i<16; i++) ...
```
```
#pragma unroll 1 /* tell compiler to never unroll the loop */
for (int i=0; i<16; i++) ...
```
Unbounded loop unroll is under development on HCC compiler.
```
#pragma unroll /* hint to compiler to completely unroll next loop. */
for (int i=0; i<16; i++) ...
```
```
#pragma unroll 1 /* tell compiler to never unroll the loop */
for (int i=0; i<16; i++) ...
```
## In-Line Assembly
projects/clr/hipamd/docs/markdown/hip_porting_guide.md
+43 -92
Wyświetl plik
@@ -1,95 +1,50 @@
<!-- START doctoc generated TOC please keep comment here to allow auto update -->
<!-- DON'T EDIT THIS SECTION, INSTEAD RE-RUN doctoc TO UPDATE -->
**Table of Contents** *generated with [DocToc](https://github.com/thlorenz/doctoc)*
- [HIP Porting Guide](#hip-porting-guide)
- [###Table of Contents](#table-of-contents)
- [Porting a New Cuda Project](#porting-a-new-cuda-project)
- [General Tips](#general-tips)
- [Scanning existing CUDA code to scope the porting effort](#scanning-existing-cuda-code-to-scope-the-porting-effort)
- [Converting a project "in-place"](#converting-a-project-in-place)
- [Distinguishing Compiler Modes](#distinguishing-compiler-modes)
- [Identifying HIP Target Platform](#identifying-hip-target-platform)
- [Identifying the Compiler: hcc or nvcc](#identifying-the-compiler-hcc-or-nvcc)
- [Identifying Current Compilation Pass: Host or Device](#identifying-current-compilation-pass-host-or-device)
- [Compiler Defines: Summary](#compiler-defines-summary)
- [Identifying Architecture Features](#identifying-architecture-features)
- [HIP_ARCH Defines](#hip_arch-defines)
- [Device-Architecture Properties](#device-architecture-properties)
- [Table of Architecture Properties](#table-of-architecture-properties)
- [Finding HIP](#finding-hip)
- [Compiler Options](#compiler-options)
- [Linking Issues](#linking-issues)
- [Linking With hipcc](#linking-with-hipcc)
- [-lm Option](#-lm-option)
- [Linking Code With Other Compilers](#linking-code-with-other-compilers)
- [libc++ and libstdc++](#libc-and-libstdc)
- [HIP Headers (hip_runtime.h, hip_runtime_api.h)](#hip-headers-hip_runtimeh-hip_runtime_apih)
- [Using a Standard C++ Compiler](#using-a-standard-c-compiler)
- [cuda.h](#cudah)
- [Choosing HIP File Extensions](#choosing-hip-file-extensions)
- [Workarounds](#workarounds)
- [warpSize](#warpsize)
- [Textures and Cache Control](#textures-and-cache-control)
- [More Tips](#more-tips)
- [hcc CPU Mode](#hcc-cpu-mode)
- [HIPTRACE Mode](#hiptrace-mode)
- [Environment Variables](#environment-variables)
- [Debugging hipcc](#debugging-hipcc)
- [What Does This Error Mean?](#what-does-this-error-mean)
- [/usr/include/c++/v1/memory:5172:15: error: call to implicitly deleted default constructor of 'std::__1::bad_weak_ptr' throw bad_weak_ptr();](#usrincludecv1memory517215-error-call-to-implicitly-deleted-default-constructor-of-std__1bad_weak_ptr-throw-bad_weak_ptr)
- [grid_launch kernel dispatch - fallback](#grid_launch-kernel-dispatch---fallback)
- [HIP Environment Variables](#hip-environment-variables)
- [Editor Highlighting](#editor-highlighting)
<!-- END doctoc generated TOC please keep comment here to allow auto update -->
# HIP Porting Guide
In addition to providing a portable C++ programmming environement for GPUs, HIP is designed to ease
the porting of existing CUDA code into the HIP environment. This section describes the available tools
and provides practical suggestions on how to port CUDA code and work through common issues.
###Table of Contents
## Table of Contents
=================
* [HIP Porting Guide](#hip-porting-guide)
* [Table of Contents](#table-of-contents)
* [Porting a New Cuda Project To HIP](#porting-a-new-cuda-project)
* [General Tips](#general-tips" aria-hidden="true"><span aria-hidden="true)
* [Scanning existing CUDA code to scope the porting effort](#scanning-existing-cuda-code-to-scope-the-porting-effort" aria-hidden="true"><span aria-hidden="true)
* [Distinguishing Compiler Modes](#distinguishing-compiler-modes" aria-hidden="true"><span aria-hidden="true)
* [Identifying HIP Target Platform](#identifying-hip-target-platform" aria-hidden="true"><span aria-hidden="true)
* [Identifying the Compiler: hcc or nvcc](#identifying-the-compiler-hcc-or-nvcc" aria-hidden="true"><span aria-hidden="true)
* [Identifying Current Compilation Pass: Host or Device](#identifying-current-compilation-pass-host-or-device" aria-hidden="true"><span aria-hidden="true)
* [Compiler Defines: Summary](#compiler-defines-summary" aria-hidden="true"><span aria-hidden="true)
* [Identifying Architecture Features](#identifying-architecture-features" aria-hidden="true"><span aria-hidden="true)
* [HIP_ARCH Defines](#hip_arch-defines" aria-hidden="true"><span aria-hidden="true)
* [Device-Architecture Properties](#device-architecture-properties" aria-hidden="true"><span aria-hidden="true)
* [Table of Architecture Properties](#table-of-architecture-properties" aria-hidden="true"><span aria-hidden="true)
* [Finding HIP](#finding-hip" aria-hidden="true"><span aria-hidden="true)
* [Compiler Options](#compiler-options" aria-hidden="true"><span aria-hidden="true)
* [Linking Issues](#linking-issues" aria-hidden="true"><span aria-hidden="true)
* [Linking With hipcc](#linking-with-hipcc" aria-hidden="true"><span aria-hidden="true)
* [-lm Option](#-lm-option" aria-hidden="true"><span aria-hidden="true)
* [Linking Code With Other Compilers](#linking-code-with-other-compilers" aria-hidden="true"><span aria-hidden="true)
* [libc and libstdc ](#libc-and-libstdc" aria-hidden="true"><span aria-hidden="true)
* [HIP Headers (hip_runtime.h, hip_runtime_api.h)](#hip-headers-hip_runtimeh-hip_runtime_apih" aria-hidden="true"><span aria-hidden="true)
* [Using a Standard C Compiler](#using-a-standard-c-compiler" aria-hidden="true"><span aria-hidden="true)
* [cuda.h](#cudah" aria-hidden="true"><span aria-hidden="true)
* [Choosing HIP File Extensions](#choosing-hip-file-extensions" aria-hidden="true"><span aria-hidden="true)
* [Workarounds](#workarounds" aria-hidden="true"><span aria-hidden="true)
* [warpSize](#warpsize" aria-hidden="true"><span aria-hidden="true)
* [Textures and Cache Control](#textures-and-cache-control" aria-hidden="true"><span aria-hidden="true)
* [More Tips](#more-tips" aria-hidden="true"><span aria-hidden="true)
* [hcc CPU Mode](#hcc-cpu-mode" aria-hidden="true"><span aria-hidden="true)
* [HIPTRACE Mode](#hiptrace-mode" aria-hidden="true"><span aria-hidden="true)
* [Environment Variables](#environment-variables" aria-hidden="true"><span aria-hidden="true)
* [Debugging hipcc](#debugging-hipcc" aria-hidden="true"><span aria-hidden="true)
* [What Does This Error Mean?](#what-does-this-error-mean" aria-hidden="true"><span aria-hidden="true)
* [/usr/include/c /v1/memory:5172:15: error: call to implicitly deleted default constructor of 'std::__1::bad_weak_ptr' throw bad_weak_ptr();](#usrincludecv1memory517215-error-call-to-implicitly-deleted-default-constructor-of-std__1bad_weak_ptr-throw-bad_weak_ptr" aria-hidden="true"><span aria-hidden="true)
* [grid_launch kernel dispatch - fallback](#grid_launch-kernel-dispatch---fallback" aria-hidden="true"><span aria-hidden="true)
* [Editor Highlighting](#editor-highlighting)
<!-- toc -->
- [Porting a New Cuda Project](#porting-a-new-cuda-project)
* [General Tips](#general-tips)
* [Scanning existing CUDA code to scope the porting effort](#scanning-existing-cuda-code-to-scope-the-porting-effort)
* [Converting a project "in-place"](#converting-a-project-in-place)
- [Distinguishing Compiler Modes](#distinguishing-compiler-modes)
* [Identifying HIP Target Platform](#identifying-hip-target-platform)
* [Identifying the Compiler: hcc or nvcc](#identifying-the-compiler-hcc-or-nvcc)
* [Identifying Current Compilation Pass: Host or Device](#identifying-current-compilation-pass-host-or-device)
* [Compiler Defines: Summary](#compiler-defines-summary)
- [Identifying Architecture Features](#identifying-architecture-features)
* [HIP_ARCH Defines](#hip_arch-defines)
* [Device-Architecture Properties](#device-architecture-properties)
* [Table of Architecture Properties](#table-of-architecture-properties)
- [Finding HIP](#finding-hip)
- [Compiler Options](#compiler-options)
- [Linking Issues](#linking-issues)
* [Linking With hipcc](#linking-with-hipcc)
* [-lm Option](#-lm-option)
- [Linking Code With Other Compilers](#linking-code-with-other-compilers)
* [libc++ and libstdc++](#libc-and-libstdc)
* [HIP Headers (hip_runtime.h, hip_runtime_api.h)](#hip-headers-hip_runtimeh-hip_runtime_apih)
* [Using a Standard C++ Compiler](#using-a-standard-c-compiler)
+ [cuda.h](#cudah)
* [Choosing HIP File Extensions](#choosing-hip-file-extensions)
* [Workarounds](#workarounds)
+ [warpSize](#warpsize)
+ [Textures and Cache Control](#textures-and-cache-control)
- [More Tips](#more-tips)
* [HIPTRACE Mode](#hiptrace-mode)
* [Environment Variables](#environment-variables)
* [Debugging hipcc](#debugging-hipcc)
* [What Does This Error Mean?](#what-does-this-error-mean)
+ [/usr/include/c++/v1/memory:5172:15: error: call to implicitly deleted default constructor of 'std::__1::bad_weak_ptr' throw bad_weak_ptr();](#usrincludecv1memory517215-error-call-to-implicitly-deleted-default-constructor-of-std__1bad_weak_ptr-throw-bad_weak_ptr)
+ [HIP Environment Variables](#hip-environment-variables)
+ [Editor Highlighting](#editor-highlighting)
<!-- tocstop -->
## Porting a New Cuda Project
@@ -463,17 +418,13 @@ void myFunc ()
```
Additionally, many of the Rodinia benchmarks demonstrate how to modify hipified programs so that textures are not required - search for USE_TEXTURES define in the rodinia source directory.
For example, [here
Cuda programs that employ sampler hardware must either wait for hcc texture support or use more-sophisticated workarounds.
## More Tips
### hcc CPU Mode
Recent hcc versions support CPU accelerator targets. This feature enables some interesting possibilities for HIP porting:
- hcc can run on any machine, including perhaps a cross-compiling environment on a machine also running nvcc
- Standard CPU debuggers can debug CPU code
- A single code path can run on an AMD or Nvidia GPU or CPU, but the CPU accelerator is a low-performance target---it’s just a single core and lacks SIMD acceleration
### HIPTRACE Mode
On an hcc/AMD platform, set the HIP_TRACE_API environment variable to see a textural API trace. Use the following bit mask:
@@ -490,7 +441,7 @@ On hcc/AMD platforms, set the HIP_PRINT_ENV environment variable to 1 and run an
- HIP_TRACE_API = 1: trace each HIP API call. Print the function name and return code to stderr as the program executes.
- HIP_LAUNCH_BLOCKING = 0: make HIP APIs “host-synchronous” so they are blocked until any kernel launches or data-copy commands are complete (an alias is CUDA_LAUNCH_BLOCKING)
- KMDUMPISA = 1 : Will dump the GCN ISA for all kernels into the local directory.
- KMDUMPISA = 1 : Will dump the GCN ISA for all kernels into the local directory. (This flag is provided by HCC).
### Debugging hipcc