From e78fd3fdcfcdcfbc76324cad8a5eccb83082d706 Mon Sep 17 00:00:00 2001 From: PBerger Date: Sun, 2 Mar 2014 01:26:48 +0100 Subject: [PATCH] Added a function header for a Value Iteration Kernel. Removed the intermediate project from CMake Former-commit-id: 8b49570eb0facad002c116b505784f680795ef29 --- resources/cudaForStorm/CMakeLists.txt | 190 ++++-------------- .../cudaForStorm/srcCuda/allCudaKernels.h | 3 +- .../srcCuda/basicValueIteration.cu | 50 +++++ .../srcCuda/basicValueIteration.h | 4 + 4 files changed, 93 insertions(+), 154 deletions(-) create mode 100644 resources/cudaForStorm/srcCuda/basicValueIteration.cu create mode 100644 resources/cudaForStorm/srcCuda/basicValueIteration.h diff --git a/resources/cudaForStorm/CMakeLists.txt b/resources/cudaForStorm/CMakeLists.txt index ab2c2c8c7..1ddfe128f 100644 --- a/resources/cudaForStorm/CMakeLists.txt +++ b/resources/cudaForStorm/CMakeLists.txt @@ -11,20 +11,17 @@ set (STORM_CPP_VERSION_MINOR 0) include_directories("${PROJECT_SOURCE_DIR}") include_directories("${PROJECT_SOURCE_DIR}/src") -message(STATUS "CUDA_PATH is ${CUDA_PATH} or $ENV{CUDA_PATH}") +message(STATUS "StoRM (CudaPlugin) - CUDA_PATH is ${CUDA_PATH} or $ENV{CUDA_PATH}") ############################################################# ## ## CMake options of StoRM ## ############################################################# -option(DEBUG "Sets whether the DEBUG mode is used" ON) -option(USE_POPCNT "Sets whether the popcnt instruction is going to be used." ON) +option(CUDAFORSTORM_DEBUG "Sets whether the DEBUG mode is used" ON) option(LINK_LIBCXXABI "Sets whether libc++abi should be linked." OFF) option(USE_LIBCXX "Sets whether the standard library is libc++." OFF) -option(ENABLE_GLPK "Sets whether StoRM is built with support for glpk." OFF) -set(GUROBI_ROOT "" CACHE STRING "The root directory of Gurobi (if available).") -set(Z3_ROOT "" CACHE STRING "The root directory of Z3 (if available).") + set(ADDITIONAL_INCLUDE_DIRS "" CACHE STRING "Additional directories added to the include directories.") set(ADDITIONAL_LINK_DIRS "" CACHE STRING "Additional directories added to the link directories.") set(STORM_LIB_INSTALL_DIR "${PROJECT_SOURCE_DIR}/../../build/cudaForStorm" CACHE STRING "The Build directory of storm, where the library files should be installed to (if available).") @@ -43,28 +40,15 @@ find_package(Doxygen REQUIRED) find_package(Threads REQUIRED) # If the DEBUG option was turned on, we will target a debug version and a release version otherwise -if (DEBUG) +if (CUDAFORSTORM_DEBUG) set (CMAKE_BUILD_TYPE "DEBUG") else() set (CMAKE_BUILD_TYPE "RELEASE") endif() -message(STATUS "StoRM - Building ${CMAKE_BUILD_TYPE} version.") - -if ("${GUROBI_ROOT}" STREQUAL "") - set(ENABLE_GUROBI OFF) -else() - set(ENABLE_GUROBI ON) -endif() - -if ("${Z3_ROOT}" STREQUAL "") - set(ENABLE_Z3 OFF) -else() - set(ENABLE_Z3 ON) - set(Z3_LIB_NAME "z3") -endif() +message(STATUS "StoRM (CudaPlugin) - Building ${CMAKE_BUILD_TYPE} version.") -message(STATUS "StoRM - CMAKE_BUILD_TYPE: ${CMAKE_BUILD_TYPE}") -message(STATUS "StoRM - CMAKE_BUILD_TYPE (ENV): $ENV{CMAKE_BUILD_TYPE}") +message(STATUS "StoRM (CudaPlugin) - CMAKE_BUILD_TYPE: ${CMAKE_BUILD_TYPE}") +message(STATUS "StoRM (CudaPlugin) - CMAKE_BUILD_TYPE (ENV): $ENV{CMAKE_BUILD_TYPE}") ############################################################# ## @@ -78,8 +62,8 @@ set(CUDA_SEPARABLE_COMPILATION ON) #set(CUDA_NVCC_FLAGS "-arch=sm_30") # Because the FindCUDA.cmake file has a path related bug, two folders have to be present -file(MAKE_DIRECTORY "${PROJECT_BINARY_DIR}/CMakeFiles/cudaLibrary.dir/Debug") -file(MAKE_DIRECTORY "${PROJECT_BINARY_DIR}/CMakeFiles/cudaLibrary.dir/Release") +file(MAKE_DIRECTORY "${PROJECT_BINARY_DIR}/CMakeFiles/cudaForStorm.dir/Debug") +file(MAKE_DIRECTORY "${PROJECT_BINARY_DIR}/CMakeFiles/cudaForStorm.dir/Release") ############################################################# @@ -87,28 +71,13 @@ file(MAKE_DIRECTORY "${PROJECT_BINARY_DIR}/CMakeFiles/cudaLibrary.dir/Release") ## Compiler specific settings and definitions ## ############################################################# - -# Path to the no-strict-aliasing target -set(CONVERSIONHELPER_TARGET "${PROJECT_SOURCE_DIR}/src/utility/ConversionHelper.cpp") - if(CMAKE_COMPILER_IS_GNUCC) - message(STATUS "StoRM - Using Compiler Configuration: GCC") + message(STATUS "StoRM (CudaPlugin) - Using Compiler Configuration: GCC") # Set standard flags for GCC set(CMAKE_CXX_FLAGS_RELEASE "${CMAKE_CXX_FLAGS_RELEASE} -funroll-loops") set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -std=c++11 -Wall -pedantic") - # -Werror is atm removed as this gave some problems with existing code - # May be re-set later - # (Thomas Heinemann, 2012-12-21) - - # Turn on popcnt instruction if desired (yes by default) - if (USE_POPCNT) - set (CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -mpopcnt") - endif(USE_POPCNT) - - # Set the no-strict-aliasing target for GCC - set_source_files_properties(${CONVERSIONHELPER_TARGET} PROPERTIES COMPILE_FLAGS " -fno-strict-aliasing ") elseif(MSVC) - message(STATUS "StoRM - Using Compiler Configuration: MSVC") + message(STATUS "StoRM (CudaPlugin) - Using Compiler Configuration: MSVC") # required for GMM to compile, ugly error directive in their code add_definitions(/D_SCL_SECURE_NO_DEPRECATE /D_CRT_SECURE_NO_WARNINGS) # required as the PRCTL Parser bloats object files (COFF) beyond their maximum size (see http://msdn.microsoft.com/en-us/library/8578y171(v=vs.110).aspx) @@ -117,26 +86,18 @@ elseif(MSVC) add_definitions(/D_VARIADIC_MAX=10) # Windows.h breaks GMM in gmm_except.h because of its macro definition for min and max add_definitions(/DNOMINMAX) - - if(ENABLE_Z3) - set(Z3_LIB_NAME "libz3") - endif() - - # MSVC does not do strict-aliasing, so no option needed else(CLANG) - message(STATUS "StoRM - Using Compiler Configuration: Clang (LLVM)") + message(STATUS "StoRM (CudaPlugin) - Using Compiler Configuration: Clang (LLVM)") # As CLANG is not set as a variable, we need to set it in case we have not matched another compiler. set (CLANG ON) # Set standard flags for clang set (CMAKE_CXX_FLAGS_RELEASE "${CMAKE_CXX_FLAGS_RELEASE} -funroll-loops -O3") if(UNIX AND NOT APPLE AND NOT USE_LIBCXX) set(CLANG_STDLIB libstdc++) - message(STATUS "StoRM - Linking against libstdc++") + message(STATUS "StoRM (CudaPlugin) - Linking against libstdc++") else() set(CLANG_STDLIB libc++) - message(STATUS "StoRM - Linking against libc++") - # Disable Cotire - set(STORM_USE_COTIRE OFF) + message(STATUS "StoRM (CudaPlugin) - Linking against libc++") # Set up some Xcode specific settings set(CMAKE_XCODE_ATTRIBUTE_CLANG_CXX_LANGUAGE_STANDARD "c++11") set(CMAKE_XCODE_ATTRIBUTE_CLANG_CXX_LIBRARY "libc++") @@ -145,14 +106,6 @@ else(CLANG) set (CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -std=c++11 -stdlib=${CLANG_STDLIB} -Wall -pedantic -Wno-unused-variable -DBOOST_RESULT_OF_USE_TR1 -DBOOST_NO_DECLTYPE -ftemplate-depth=1024") set (CMAKE_CXX_FLAGS_DEBUG "${CMAKE_CXX_FLAGS_DEBUG} -g") - - # Turn on popcnt instruction if desired (yes by default) - if (USE_POPCNT) - set (CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -mpopcnt") - endif(USE_POPCNT) - - # Set the no-strict-aliasing target for Clang - set_source_files_properties(${CONVERSIONHELPER_TARGET} PROPERTIES COMPILE_FLAGS " -fno-strict-aliasing ") endif() ############################################################# @@ -160,35 +113,6 @@ endif() ## CMake-generated Config File for StoRM ## ############################################################# -# Base path for test files -set(STORM_CPP_TESTS_BASE_PATH "${PROJECT_SOURCE_DIR}/test") -# Gurobi Defines -if (ENABLE_GUROBI) - set(STORM_CPP_GUROBI_DEF "define") -else() - set(STORM_CPP_GUROBI_DEF "undef") -endif() - -# glpk defines -if (ENABLE_GLPK) - set(STORM_CPP_GLPK_DEF "define") -else() - set(STORM_CPP_GLPK_DEF "undef") -endif() - -# Z3 Defines -if (ENABLE_Z3) - set(STORM_CPP_Z3_DEF "define") -else() - set(STORM_CPP_Z3_DEF "undef") -endif() - -# Intel TBB Defines -if (TBB_FOUND AND ENABLE_INTELTBB) - set(STORM_CPP_INTELTBB_DEF "define") -else() - set(STORM_CPP_INTELTBB_DEF "undef") -endif() # Configure a header file to pass some of the CMake settings to the source code configure_file ( @@ -213,7 +137,7 @@ file(GLOB_RECURSE CUDAFORSTORM_CUDA_SOURCES "${PROJECT_SOURCE_DIR}/srcCuda/*.cu" file(GLOB_RECURSE CUDAFORSTORM_CUDA_HEADERS "${PROJECT_SOURCE_DIR}/srcCuda/*.h") # Additional include files like the storm-config.h -file(GLOB_RECURSE STORM_BUILD_HEADERS ${PROJECT_BINARY_DIR}/include/*.h) +file(GLOB_RECURSE CUDAFORSTORM_BUILD_HEADERS ${PROJECT_BINARY_DIR}/include/*.h) # Group the headers and sources source_group(main FILES ${CUDAFORSTORM_HEADERS} ${CUDAFORSTORM_SOURCES}) @@ -221,11 +145,11 @@ source_group(cuda FILES ${CUDAFORSTORM_CUDA_SOURCES} ${CUDAFORSTORM_CUDA_HEADERS # Add custom additional include or link directories if (ADDITIONAL_INCLUDE_DIRS) - message(STATUS "StoRM - Using additional include directories ${ADDITIONAL_INCLUDE_DIRS}") + message(STATUS "StoRM (CudaPlugin) - Using additional include directories ${ADDITIONAL_INCLUDE_DIRS}") include_directories(${ADDITIONAL_INCLUDE_DIRS}) endif(ADDITIONAL_INCLUDE_DIRS) if (ADDITIONAL_LINK_DIRS) - message(STATUS "StoRM - Using additional link directories ${ADDITIONAL_LINK_DIRS}") + message(STATUS "StoRM (CudaPlugin) - Using additional link directories ${ADDITIONAL_LINK_DIRS}") link_directories(${ADDITIONAL_LINK_DIRS}) endif(ADDITIONAL_LINK_DIRS) @@ -234,19 +158,7 @@ endif(ADDITIONAL_LINK_DIRS) ## Pre executable-creation link_directories setup ## ############################################################# -if (ENABLE_GUROBI) - link_directories("${GUROBI_ROOT}/lib") -endif() -if (ENABLE_Z3) - link_directories("${Z3_ROOT}/bin") -endif() -if ((NOT Boost_LIBRARY_DIRS) OR ("${Boost_LIBRARY_DIRS}" STREQUAL "")) - set(Boost_LIBRARY_DIRS "${Boost_INCLUDE_DIRS}/stage/lib") -endif () -link_directories(${Boost_LIBRARY_DIRS}) -if (TBB_FOUND AND ENABLE_INTELTBB) - link_directories(${TBB_LIBRARY_DIRS}) -endif() + ############################################################################### ## # @@ -255,17 +167,15 @@ endif() ## All link_directories() calls MUST be made before this point # ## # ############################################################################### - -# Since this will be a library include (GenerateExportHeader) -add_library(cudaForStorm SHARED ${CUDAFORSTORM_HEADERS} ${CUDAFORSTORM_SOURCES}) -GENERATE_EXPORT_HEADER( cudaForStorm - BASE_NAME cudaForStorm - EXPORT_MACRO_NAME cudaForStorm_EXPORT - EXPORT_FILE_NAME include/cudaForStorm_Export.h - STATIC_DEFINE cudaForStorm_BUILT_AS_STATIC -) +#add_library(cudaForStorm SHARED ${CUDAFORSTORM_HEADERS} ${CUDAFORSTORM_SOURCES}) +#GENERATE_EXPORT_HEADER( cudaForStorm +# BASE_NAME cudaForStorm +# EXPORT_MACRO_NAME cudaForStorm_EXPORT +# EXPORT_FILE_NAME include/cudaForStorm_Export.h +# STATIC_DEFINE cudaForStorm_BUILT_AS_STATIC +#) ############################################################# ## @@ -273,50 +183,24 @@ GENERATE_EXPORT_HEADER( cudaForStorm ## ############################################################# #set(CUDA_NVCC_FLAGS ${CUDA_NVCC_FLAGS} --gpu-architecture sm_30) -cuda_add_library(cudaLibrary +cuda_add_library(cudaForStorm ${CUDAFORSTORM_CUDA_SOURCES} ${CUDAFORSTORM_CUDA_HEADERS} OPTIONS -DSTUFF="" -arch=sm_30 RELEASE -DNDEBUG DEBUG -g -DDEBUG ) -target_link_libraries(cudaLibrary ${CUDA_cusparse_LIBRARY}) -ADD_DEPENDENCIES(cudaForStorm cudaLibrary) -target_link_libraries(cudaForStorm cudaLibrary) -message(STATUS "Found CUDA SDK in Version ${CUDA_VERSION_STRING}, sparse lib is ${CUDA_cusparse_LIBRARY}") +GENERATE_EXPORT_HEADER( cudaForStorm + BASE_NAME cudaForStorm + EXPORT_MACRO_NAME cudaForStorm_EXPORT + EXPORT_FILE_NAME include/cudaForStorm_Export.h + STATIC_DEFINE cudaForStorm_BUILT_AS_STATIC +) +#target_link_libraries(cudaLibrary ${CUDA_cusparse_LIBRARY}) +#ADD_DEPENDENCIES(cudaForStorm cudaLibrary) +#target_link_libraries(cudaForStorm cudaLibrary) +message(STATUS "StoRM (CudaPlugin) - Found CUDA SDK in Version ${CUDA_VERSION_STRING}, sparse lib is ${CUDA_cusparse_LIBRARY}") include_directories(${CUDA_INCLUDE_DIRS}) -############################################################# -## -## Gurobi (optional) -## -############################################################# -if (ENABLE_GUROBI) - message (STATUS "StoRM - Linking with Gurobi") - include_directories("${GUROBI_ROOT}/include") - target_link_libraries(cudaForStorm "gurobi56") -endif(ENABLE_GUROBI) - -############################################################# -## -## glpk (optional) -## -############################################################# -if (ENABLE_GLPK) - message (STATUS "StoRM - Linking with glpk") - target_link_libraries(cudaForStorm "glpk") -endif(ENABLE_GLPK) - -############################################################# -## -## Z3 (optional) -## -############################################################# -if (ENABLE_Z3) - message (STATUS "StoRM - Linking with Z3") - include_directories("${Z3_ROOT}/include") - target_link_libraries(cudaForStorm ${Z3_LIB_NAME}) -endif(ENABLE_Z3) - ############################################################# ## ## Threads @@ -333,7 +217,7 @@ endif(MSVC) # Link against libc++abi if requested. May be needed to build on Linux systems using clang. if (LINK_LIBCXXABI) - message (STATUS "StoRM - Linking against libc++abi.") + message (STATUS "StoRM (CudaPlugin) - Linking against libc++abi.") target_link_libraries(cudaForStorm "c++abi") endif(LINK_LIBCXXABI) diff --git a/resources/cudaForStorm/srcCuda/allCudaKernels.h b/resources/cudaForStorm/srcCuda/allCudaKernels.h index 1631b9104..182f1b770 100644 --- a/resources/cudaForStorm/srcCuda/allCudaKernels.h +++ b/resources/cudaForStorm/srcCuda/allCudaKernels.h @@ -1,4 +1,5 @@ #include "utility.h" #include "bandWidth.h" #include "basicAdd.h" -#include "kernelSwitchTest.h" \ No newline at end of file +#include "kernelSwitchTest.h" +#include "basicValueIteration.h" \ No newline at end of file diff --git a/resources/cudaForStorm/srcCuda/basicValueIteration.cu b/resources/cudaForStorm/srcCuda/basicValueIteration.cu new file mode 100644 index 000000000..d22d289ec --- /dev/null +++ b/resources/cudaForStorm/srcCuda/basicValueIteration.cu @@ -0,0 +1,50 @@ +#include "basicValueIteration.h" + +#include +#include + +#include +#include "cusparse_v2.h" + + +__global__ void cuda_kernel_basicValueIteration_mvReduce(int const * const A, int * const B) { + *B = *A; +} + +void basicValueIteration_mvReduce(uint_fast64_t const maxIterationCount, std::vector const& matrixRowIndices, std::vector const& matrixColumnIndices, std::vector const& matrixValues, std::vector& x, std::vector const& b, std::vector const& nondeterministicChoiceIndices) { + std::cout << "basicValueIteration_mvReduce is implemented for ValueType == double :)" << std::endl; +} + +/* +void kernelSwitchTest(size_t N) { + int* deviceIntA; + int* deviceIntB; + + if (cudaMalloc((void**)&deviceIntA, sizeof(int)) != cudaSuccess) { + std::cout << "Error in cudaMalloc while allocating " << sizeof(int) << " Bytes!" << std::endl; + return; + } + if (cudaMalloc((void**)&deviceIntB, sizeof(int)) != cudaSuccess) { + std::cout << "Error in cudaMalloc while allocating " << sizeof(int) << " Bytes!" << std::endl; + return; + } + + // Allocate space on the device + auto start_time = std::chrono::high_resolution_clock::now(); + for (int i = 0; i < N; ++i) { + cuda_kernel_kernelSwitchTest<<<1,1>>>(deviceIntA, deviceIntB); + } + auto end_time = std::chrono::high_resolution_clock::now(); + std::cout << "Switching the Kernel " << N << " times took " << std::chrono::duration_cast(end_time - start_time).count() << "micros" << std::endl; + std::cout << "Resulting in " << (std::chrono::duration_cast(end_time - start_time).count() / ((double)(N))) << "Microseconds per Kernel Switch" << std::endl; + + // Free memory on device + if (cudaFree(deviceIntA) != cudaSuccess) { + std::cout << "Error in cudaFree!" << std::endl; + return; + } + if (cudaFree(deviceIntB) != cudaSuccess) { + std::cout << "Error in cudaFree!" << std::endl; + return; + } +}*/ \ No newline at end of file diff --git a/resources/cudaForStorm/srcCuda/basicValueIteration.h b/resources/cudaForStorm/srcCuda/basicValueIteration.h new file mode 100644 index 000000000..92bb44270 --- /dev/null +++ b/resources/cudaForStorm/srcCuda/basicValueIteration.h @@ -0,0 +1,4 @@ +#include +#include + +void basicValueIteration_mvReduce(uint_fast64_t const maxIterationCount, std::vector const& matrixRowIndices, std::vector const& matrixColumnIndices, std::vector const& matrixValues, std::vector& x, std::vector const& b, std::vector const& nondeterministicChoiceIndices); \ No newline at end of file