diff --git a/CMakeLists.txt b/CMakeLists.txt index 5aedd12d77..1011afd7a0 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -2,32 +2,34 @@ cmake_minimum_required(VERSION 2.6) project(hip_hcc) # preserve HCC_HOME env in the generated -set (HCC_HOME $ENV{HCC_HOME}) -if (NOT DEFINED HCC_HOME) - set (HCC_HOME /opt/hcc) + +if (NOT DEFINED HCC_DIR) + set (HCC_DIR "/opt/hcc") endif() +message(STATUS ${HCC_DIR}) -MESSAGE ("HCC_HOME=" ${HCC_HOME}) - -set(HSA_PATH "/opt/hsa") - - -set (HIP_INSTALL_PATH $ENV{HIP_PATH}) -if (NOT DEFINED HIP_INSTALL_PATH) - set (HIP_INSTALLPATH /opt/hip) +if(NOT DEFINED HIP_INSTALL_DIR) + set(HIP_INSTALL_DIR "/opt/hip") endif() +message(STATUS ${HIP_INSTALL_DIR}) -if (CMAKE_INSTALL_PREFIX_INITIALIZED_TO_DEFAULT) - set(CMAKE_INSTALL_PREFIX "${HIP_INSTALL_PATH}" CACHE PATH "Default installation path of hip" FORCE) -endif () +if(NOT DEFINED HSA_DIR) + set(HSA_DIR "/opt/hsa") +endif() +message(STATUS ${HSA_DIR}) + + +#if (CMAKE_INSTALL_PREFIX_INITIALIZED_TO_DEFAULT) +# set(CMAKE_INSTALL_PREFIX "${HIP_INSTALL_PATH}" CACHE PATH "Default installation path of hip" FORCE) +#endif () include_directories(${PROJECT_SOURCE_DIR}/include) -set(CMAKE_CXX_COMPILER "${HCC_HOME}/bin/hcc") -set(CMAKE_C_COMPILER "${HCC_HOME}/bin/hcc") +set(CMAKE_CXX_COMPILER "${HCC_DIR}/bin/hcc") +set(CMAKE_C_COMPILER "${HCC_DIR}/bin/hcc") -set(CMAKE_CXX_FLAGS " -hc -I${HCC_HOME}/include -I${HSA_PATH}/include -I${HIP_PATH}/include -stdlib=libc++ ") -set(CMAKE_C_FLAGS " -hc -I${HCC_HOME}/include -I${HSA_PATH}/include -I${HIP_PATH}/include -stdlib=libc++ ") +set(CMAKE_CXX_FLAGS " -hc -I${HCC_DIR}/include -I${HSA_DIR}/include -stdlib=libc++ ") +set(CMAKE_C_FLAGS " -hc -I${HCC_DIR}/include -I${HSA_DIR}/include -stdlib=libc++ ") set(SOURCE_FILES src/hip_hcc.cpp src/hip_device.cpp @@ -40,9 +42,9 @@ src/staging_buffer.cpp) add_library(hip_hcc STATIC ${SOURCE_FILES}) -install(TARGETS hip_hcc DESTINATION lib) -install(DIRECTORY src DESTINATION .) -install(DIRECTORY bin DESTINATION . USE_SOURCE_PERMISSIONS) -install(DIRECTORY include DESTINATION .) +install(TARGETS hip_hcc DESTINATION ${HIP_INSTALL_DIR}/lib) +install(DIRECTORY src DESTINATION ${HIP_INSTALL_DIR}) +install(DIRECTORY bin DESTINATION ${HIP_INSTALL_DIR} USE_SOURCE_PERMISSIONS) +install(DIRECTORY include DESTINATION ${HIP_INSTALL_DIR}) diff --git a/include/hcc_detail/hip_runtime_api.h b/include/hcc_detail/hip_runtime_api.h index 0ced74a084..92aef3d213 100644 --- a/include/hcc_detail/hip_runtime_api.h +++ b/include/hcc_detail/hip_runtime_api.h @@ -47,28 +47,36 @@ extern "C" { * @{ */ //! Flags that can be used with hipStreamCreateWithFlags -#define hipStreamDefault 0x00 ///< Default stream creation flags. These are used with hipStreamCreate(). -#define hipStreamNonBlocking 0x01 ///< Stream does not implicitly synchronize with null stream +#define hipStreamDefault 0x00 ///< Default stream creation flags. These are used with hipStreamCreate(). +#define hipStreamNonBlocking 0x01 ///< Stream does not implicitly synchronize with null stream //! Flags that can be used with hipEventCreateWithFlags: -#define hipEventDefault 0x0 ///< Default flags -#define hipEventBlockingSync 0x1 ///< Waiting will yield CPU. Power-friendly and usage-friendly but may increase latency. -#define hipEventDisableTiming 0x2 ///< Disable event's capability to record timing information. May improve performance. -#define hipEventInterprocess 0x4 ///< Event can support IPC. @warning - not supported in HIP. +#define hipEventDefault 0x0 ///< Default flags +#define hipEventBlockingSync 0x1 ///< Waiting will yield CPU. Power-friendly and usage-friendly but may increase latency. +#define hipEventDisableTiming 0x2 ///< Disable event's capability to record timing information. May improve performance. +#define hipEventInterprocess 0x4 ///< Event can support IPC. @warning - not supported in HIP. //! Flags that can be used with hipHostMalloc -#define hipHostMallocDefault 0x0 -#define hipHostMallocPortable 0x1 -#define hipHostMallocMapped 0x2 -#define hipHostMallocWriteCombined 0x4 +#define hipHostMallocDefault 0x0 +#define hipHostMallocPortable 0x1 +#define hipHostMallocMapped 0x2 +#define hipHostMallocWriteCombined 0x4 //! Flags that can be used with hipHostRegister -#define hipHostRegisterDefault 0x0 ///< Memory is Mapped and Portable -#define hipHostRegisterPortable 0x1 ///< Memory is considered registered by all contexts. HIP only supports one context so this is always assumed true. -#define hipHostRegisterMapped 0x2 ///< Map the allocation into the address space for the current device. The device pointer can be obtained with #hipHostGetDevicePointer. -#define hipHostRegisterIoMemory 0x4 ///< Not supported. +#define hipHostRegisterDefault 0x0 ///< Memory is Mapped and Portable +#define hipHostRegisterPortable 0x1 ///< Memory is considered registered by all contexts. HIP only supports one context so this is always assumed true. +#define hipHostRegisterMapped 0x2 ///< Map the allocation into the address space for the current device. The device pointer can be obtained with #hipHostGetDevicePointer. +#define hipHostRegisterIoMemory 0x4 ///< Not supported. + + +#define hipDeviceScheduleAuto 0x0 +#define hipDeviceScheduleSpin 0x1 +#define hipDeviceScheduleYield 0x2 +#define hipDeviceBlockingSync 0x4 +#define hipDeviceMapHost 0x8 +#define hipDeviceLmemResizeToMax 0x16 /** * @warning On AMD devices and recent Nvidia devices, these hints and controls are ignored. diff --git a/src/hip_hcc.cpp b/src/hip_hcc.cpp index ff2876ffb3..0631cc6814 100644 --- a/src/hip_hcc.cpp +++ b/src/hip_hcc.cpp @@ -658,6 +658,11 @@ hipError_t ihipDevice_t::getProperties(hipDeviceProp_t* prop) prop->arch.hasDynamicParallelism = 0; prop->concurrentKernels = 1; // All ROCR hardware supports executing multiple kernels concurrently + if ( _device_flags | hipDeviceMapHost) { + prop->canMapHostMemory = 1; + } else { + prop->canMapHostMemory = 0; + } return e; } @@ -859,7 +864,7 @@ void ihipInit() //If device is not in visible devices list, ignore continue; } - g_devices[g_deviceCnt].init(g_deviceCnt, accs[i], 0); + g_devices[g_deviceCnt].init(g_deviceCnt, accs[i], hipDeviceMapHost); g_deviceCnt++; } } diff --git a/tests/src/CMakeLists.txt b/tests/src/CMakeLists.txt index b470956254..3846c1c287 100644 --- a/tests/src/CMakeLists.txt +++ b/tests/src/CMakeLists.txt @@ -8,7 +8,8 @@ include_directories( ${PROJECT_SOURCE_DIR}/include ) set (HIP_Unit_Test_VERSION_MAJOR 1) set (HIP_Unit_Test_VERSION_MINOR 0) -set (HIP_PATH $ENV{HIP_PATH}) +set(HIP_PATH $ENV{HIP_PATH}) +MESSAGE("HIP_PATH=" ${HIP_PATH}) if (NOT DEFINED HIP_PATH) set (HIP_PATH ../..) endif() @@ -44,9 +45,7 @@ if (${HIP_PLATFORM} STREQUAL "hcc") # This will create a subdir "hip_hcc" in the test build directory # Any changes to hip_hcc source will be detected and force the library and then the tests to be rebuilt. - add_subdirectory(${HIP_PATH} hip_hcc) - link_directories(${CMAKE_CURRENT_BINARY_DIR}/hip_hcc) # search the local hip_hcc for libhip_hcc.a - + elseif (${HIP_PLATFORM} STREQUAL "nvcc") MESSAGE ("HIP_PLATFORM=nvcc") diff --git a/tests/src/hipHostAlloc.cpp b/tests/src/hipHostAlloc.cpp index 2c72c3e33a..01ca04b311 100644 --- a/tests/src/hipHostAlloc.cpp +++ b/tests/src/hipHostAlloc.cpp @@ -39,7 +39,8 @@ int device; HIPCHECK(hipGetDevice(&device)); HIPCHECK(hipGetDeviceProperties(&prop, device)); if(prop.canMapHostMemory != 1){ -//std::cout<<"Exiting..."< #include