Merge branch 'master' into feature_purge_genco
[ROCm/hip commit: 72d3e679e4]
Cette révision appartient à :
Maneesh Gupta
GitHub
Externe
+4
-3
@@ -151,7 +151,7 @@ def docker_build_inside_image( def build_image, String inside_args, String platf
|
||||
// The rm command needs to run as sudo because the test steps below create files owned by root
|
||||
sh """#!/usr/bin/env bash
|
||||
set -x
|
||||
sudo rm -rf ${build_dir_rel}
|
||||
rm -rf ${build_dir_rel}
|
||||
mkdir -p ${build_dir_rel}
|
||||
cd ${build_dir_rel}
|
||||
cmake -DCMAKE_BUILD_TYPE=${build_config} -DCMAKE_INSTALL_PREFIX=staging ${optional_configure} ${source_hip_abs}
|
||||
@@ -160,6 +160,7 @@ def docker_build_inside_image( def build_image, String inside_args, String platf
|
||||
}
|
||||
|
||||
// Cap the maximum amount of testing, in case of hangs
|
||||
// Excluding hipPrintfKernel test from automation; variable fails on CI test machines
|
||||
timeout(time: 1, unit: 'HOURS')
|
||||
{
|
||||
stage("${platform} unit testing")
|
||||
@@ -169,7 +170,7 @@ def docker_build_inside_image( def build_image, String inside_args, String platf
|
||||
cd ${build_dir_rel}
|
||||
make install -j\$(nproc)
|
||||
make build_tests -i -j\$(nproc)
|
||||
make test
|
||||
ctest -E hipPrintfKernel
|
||||
"""
|
||||
// If unit tests output a junit or xunit file in the future, jenkins can parse that file
|
||||
// to display test results on the dashboard
|
||||
@@ -193,7 +194,7 @@ def docker_build_inside_image( def build_image, String inside_args, String platf
|
||||
if( platform.toLowerCase( ).startsWith( 'hcc-ctu' ) )
|
||||
{
|
||||
archiveArtifacts artifacts: "${build_dir_rel}/*.deb", fingerprint: true
|
||||
archiveArtifacts artifacts: "${build_dir_rel}/*.rpm", fingerprint: true
|
||||
// archiveArtifacts artifacts: "${build_dir_rel}/*.rpm", fingerprint: true
|
||||
}
|
||||
}
|
||||
}
|
||||
|
||||
@@ -74,7 +74,7 @@ HIP provides 3 environment variables in the HIP_*_BLOCKING family. These introd
|
||||
- HIP_API_BLOCKING : Forces hipMemcpyAsync and hipMemsetAsync to be host-synchronous, meaning they will wait for the requested operation to complete before returning to the caller.
|
||||
|
||||
These options cause HCC to serialize. Useful if you have libraries or code which is calling HCC kernels directly rather than using HIP.
|
||||
- HCC_SERIALZIE_KERNELS : 0x1=pre-serialize before each kernel launch, 0x2=post-serialize after each kernel launch., 0x3= pre- and post- serialize.
|
||||
- HCC_SERIALIZE_KERNEL : 0x1=pre-serialize before each kernel launch, 0x2=post-serialize after each kernel launch., 0x3= pre- and post- serialize.
|
||||
- HCC_SERIALIZE_COPY : 0x1=pre-serialize before each async copy, 0x2=post-serialize after each async copy., 0x3= pre- and post- serialize.
|
||||
|
||||
- HSA_ENABLE_SDMA=0 : Causes host-to-device and device-to-host copies to use compute shader blit kernels rather than the dedicated DMA copy engines. Compute shader copies have low latency (typically < 5us) and can achieve approximately 80% of the bandwidth of the DMA copy engine. This flag is useful to isolate issues with the hardware copy engines.
|
||||
|
||||
@@ -379,12 +379,12 @@ const std::map <llvm::StringRef, hipCounter> CUDA_INCLUDE_MAP{
|
||||
{"vector_types.h", {"hip/hip_vector_types.h", CONV_INCLUDE, API_RUNTIME}},
|
||||
|
||||
// CUBLAS includes
|
||||
{"cublas.h", {"hipblas.h", CONV_INCLUDE, API_BLAS}},
|
||||
{"cublas_v2.h", {"hipblas.h", CONV_INCLUDE, API_BLAS}},
|
||||
{"cublas.h", {"hipblas.h", CONV_INCLUDE_CUDA_MAIN_H, API_BLAS}},
|
||||
{"cublas_v2.h", {"hipblas.h", CONV_INCLUDE_CUDA_MAIN_H, API_BLAS}},
|
||||
|
||||
// CURAND includes
|
||||
{"curand.h", {"hiprand.h", CONV_INCLUDE, API_RAND}},
|
||||
{"curand_kernel.h", {"hiprand_kernel.h", CONV_INCLUDE, API_RAND}},
|
||||
{"curand.h", {"hiprand.h", CONV_INCLUDE, API_RAND}},
|
||||
{"curand_kernel.h", {"hiprand_kernel.h", CONV_INCLUDE, API_RAND}},
|
||||
|
||||
// HIP includes
|
||||
// TODO: uncomment this when hip/cudacommon.h will be renamed to hip/hipcommon.h
|
||||
|
||||
@@ -149,23 +149,38 @@ void HipifyAction::InclusionDirective(clang::SourceLocation hash_loc,
|
||||
if (!SM.isWrittenInMainFile(hash_loc)) {
|
||||
return;
|
||||
}
|
||||
if (!firstHeader) {
|
||||
firstHeader = true;
|
||||
firstHeaderLoc = hash_loc;
|
||||
}
|
||||
|
||||
const auto found = CUDA_INCLUDE_MAP.find(file_name);
|
||||
if (found == CUDA_INCLUDE_MAP.end()) {
|
||||
if (!firstNotMainHeader) {
|
||||
firstNotMainHeader = true;
|
||||
firstNotMainHeaderLoc = hash_loc;
|
||||
}
|
||||
return;
|
||||
}
|
||||
|
||||
// Special-casing to avoid duplication of the hip_runtime include.
|
||||
bool secondMainInclude = false;
|
||||
if (found->second.hipName == "hip/hip_runtime.h") {
|
||||
if (insertedRuntimeHeader) {
|
||||
secondMainInclude = true;
|
||||
if (found->second.countType == CONV_INCLUDE_CUDA_MAIN_H) {
|
||||
switch (found->second.countApiType) {
|
||||
case API_DRIVER:
|
||||
case API_RUNTIME:
|
||||
if (insertedRuntimeHeader) {
|
||||
secondMainInclude = true;
|
||||
break;
|
||||
}
|
||||
insertedRuntimeHeader = true;
|
||||
break;
|
||||
case API_BLAS:
|
||||
if (insertedBLASHeader) {
|
||||
secondMainInclude = true;
|
||||
break;
|
||||
}
|
||||
insertedBLASHeader = true;
|
||||
break;
|
||||
default:
|
||||
break;
|
||||
}
|
||||
insertedRuntimeHeader = true;
|
||||
}
|
||||
|
||||
Statistics::current().incrementCounter(found->second, file_name.str());
|
||||
@@ -198,8 +213,13 @@ void HipifyAction::InclusionDirective(clang::SourceLocation hash_loc,
|
||||
}
|
||||
|
||||
void HipifyAction::PragmaDirective(clang::SourceLocation Loc, clang::PragmaIntroducerKind Introducer) {
|
||||
if (pragmaOnce) { return; }
|
||||
if (pragmaOnce) {
|
||||
return;
|
||||
}
|
||||
clang::SourceManager& SM = getCompilerInstance().getSourceManager();
|
||||
if (!SM.isWrittenInMainFile(Loc)) {
|
||||
return;
|
||||
}
|
||||
clang::Preprocessor& PP = getCompilerInstance().getPreprocessor();
|
||||
const clang::Token tok = PP.LookAhead(0);
|
||||
StringRef Text(SM.getCharacterData(tok.getLocation()), tok.getLength());
|
||||
@@ -356,8 +376,8 @@ void HipifyAction::EndSourceFileAction() {
|
||||
clang::SourceLocation sl;
|
||||
if (pragmaOnce) {
|
||||
sl = pragmaOnceLoc;
|
||||
} else if (firstNotMainHeader) {
|
||||
sl = firstNotMainHeaderLoc;
|
||||
} else if (firstHeader) {
|
||||
sl = firstHeaderLoc;
|
||||
} else {
|
||||
sl = SM.getLocForStartOfFile(SM.getMainFileID());
|
||||
}
|
||||
|
||||
@@ -23,9 +23,10 @@ private:
|
||||
// not, we insert it at the top of the file when we finish processing it.
|
||||
// This approach means we do the best it's possible to do w.r.t preserving the user's include order.
|
||||
bool insertedRuntimeHeader = false;
|
||||
bool firstNotMainHeader = false;
|
||||
bool insertedBLASHeader = false;
|
||||
bool firstHeader = false;
|
||||
bool pragmaOnce = false;
|
||||
clang::SourceLocation firstNotMainHeaderLoc;
|
||||
clang::SourceLocation firstHeaderLoc;
|
||||
clang::SourceLocation pragmaOnceLoc;
|
||||
|
||||
/**
|
||||
|
||||
@@ -152,9 +152,9 @@ namespace hip_impl
|
||||
};
|
||||
|
||||
// CREATORS
|
||||
template<typename I>
|
||||
Bundled_code_header::Bundled_code_header(I f, I l) : Bundled_code_header{}
|
||||
template<typename RandomAccessIterator>
|
||||
Bundled_code_header::Bundled_code_header(RandomAccessIterator f, RandomAccessIterator l) : Bundled_code_header{}
|
||||
{
|
||||
read(f, l, *this);
|
||||
}
|
||||
} // Namespace hip_impl.
|
||||
} // Namespace hip_impl.
|
||||
|
||||
@@ -8,12 +8,12 @@ if(DOXYGEN_EXE)
|
||||
install(DIRECTORY RuntimeAPI/html DESTINATION docs/docs/RuntimeAPI)
|
||||
endif()
|
||||
|
||||
find_program(GRIP_EXE grip)
|
||||
if(GRIP_EXE)
|
||||
add_custom_target(convert_md_to_html ALL
|
||||
COMMAND @hip_SOURCE_DIR@/packaging/convert_md_to_html.sh @hip_SOURCE_DIR@ ${PROJECT_BINARY_DIR}/md2html)
|
||||
install(DIRECTORY md2html/ DESTINATION docs)
|
||||
endif()
|
||||
#find_program(GRIP_EXE grip)
|
||||
#if(GRIP_EXE)
|
||||
# add_custom_target(convert_md_to_html ALL
|
||||
# COMMAND @hip_SOURCE_DIR@/packaging/convert_md_to_html.sh @hip_SOURCE_DIR@ ${PROJECT_BINARY_DIR}/md2html)
|
||||
# install(DIRECTORY md2html/ DESTINATION docs)
|
||||
#endif()
|
||||
|
||||
#############################
|
||||
# Packaging steps
|
||||
|
||||
@@ -22,7 +22,12 @@ THE SOFTWARE.
|
||||
#include<cuda.h>
|
||||
#include<cuda_runtime.h>
|
||||
#include<iostream>
|
||||
#include<unistd.h>
|
||||
#ifdef _WIN32
|
||||
#include <windows.h>
|
||||
#define sleep(x) Sleep(x)
|
||||
#else
|
||||
#include <unistd.h>
|
||||
#endif
|
||||
#include<stdio.h>
|
||||
#include<malloc.h>
|
||||
|
||||
@@ -33,7 +38,7 @@ THE SOFTWARE.
|
||||
// CHECK: if(status != hipSuccess) {
|
||||
#define check(msg, status){ \
|
||||
if(status != cudaSuccess) { \
|
||||
printf("%s failed. \n", #msg); \
|
||||
printf("%s failed. \n", #msg); \
|
||||
} \
|
||||
}
|
||||
|
||||
|
||||
@@ -1,6 +1,8 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
// CHECK-NOT: #include <cuda_runtime.h>
|
||||
// CHECK: #include <stdio.h>
|
||||
#include <cuda.h>
|
||||
// CHECK-NOT: #include<cuda_runtime.h>
|
||||
#include <cuda_runtime.h>
|
||||
#include <stdio.h>
|
||||
|
||||
@@ -1,6 +1,8 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
#include <cuda_runtime.h>
|
||||
// CHECK-NOT: #include<cuda.h>
|
||||
#include <cuda.h>
|
||||
// CHECK: #include "hip/hip_runtime.h"
|
||||
// CHECK-NOT: #include "cuda_runtime.h"
|
||||
// CHECK: #include <stdio.h>
|
||||
#include "cuda.h"
|
||||
#include "cuda_runtime.h"
|
||||
#include <stdio.h>
|
||||
|
||||
@@ -0,0 +1,8 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hipblas.h>
|
||||
// CHECK-NOT: #include <cublas_v2.h>
|
||||
// CHECK: #include <stdio.h>
|
||||
#include <cublas.h>
|
||||
#include <cublas_v2.h>
|
||||
#include <stdio.h>
|
||||
@@ -0,0 +1,8 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include "hipblas.h"
|
||||
// CHECK-NOT: #include "cublas.h"
|
||||
// CHECK: #include <stdio.h>
|
||||
#include "cublas_v2.h"
|
||||
#include "cublas.h"
|
||||
#include <stdio.h>
|
||||
@@ -0,0 +1,14 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
// CHECK-NOT: #include <cuda_runtime.h>
|
||||
// CHECK: #include <iostream>
|
||||
// CHECK: #include "hipblas.h"
|
||||
// CHECK-NOT: #include "cublas.h"
|
||||
// CHECK: #include <stdio.h>
|
||||
#include <cuda.h>
|
||||
#include <cuda_runtime.h>
|
||||
#include <iostream>
|
||||
#include "cublas_v2.h"
|
||||
#include "cublas.h"
|
||||
#include <stdio.h>
|
||||
@@ -14,5 +14,6 @@ set clang_args=%4%clang_args%
|
||||
|
||||
%HIPIFY% -o=%TMP_FILE% %IN_FILE% -- %clang_args%
|
||||
if errorlevel 1 (echo Error: hipify-clang.exe failed with exit code: %errorlevel% && exit /b %errorlevel%)
|
||||
%FILE_CHECK% %IN_FILE% -input-file=%TMP_FILE%
|
||||
|
||||
findstr /v /r /c:"[ ]*//[ ]*[CHECK*|RUN]" %TMP_FILE% | %FILE_CHECK% %IN_FILE%
|
||||
if errorlevel 1 (echo Error: FileCheck.exe failed with exit code: %errorlevel% && exit /b %errorlevel%)
|
||||
|
||||
Référencer dans un nouveau ticket
Bloquer un utilisateur