diff --git a/CMakeLists.txt b/CMakeLists.txt index ebdfc25..fb1c3eb 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -21,6 +21,8 @@ endif() project(OpenCL-ROCm) +include (GNUInstallDirs) + # Add path for custom modules set(CMAKE_MODULE_PATH ${CMAKE_MODULE_PATH} @@ -38,38 +40,28 @@ set(CLANG_ENABLE_STATIC_ANALYZER OFF CACHE BOOL "") # override default option value in library and driver set(GENERIC_IS_ZERO ON CACHE BOOL ON FORCE) -add_subdirectory(compiler/llvm EXCLUDE_FROM_ALL) - -find_package(LLVM REQUIRED CONFIG PATHS ${CMAKE_BINARY_DIR}/compiler/llvm NO_DEFAULT_PATH) +find_package(LLVM REQUIRED CONFIG PATHS ${LLVM_DIR} "/opt/rocm/llvm" NO_DEFAULT_PATH) list(APPEND CMAKE_MODULE_PATH "${LLVM_CMAKE_DIR}") include(AddLLVM) add_definitions(${LLVM_DEFINITIONS}) -# TODO: add find_package for Clang and lld, and also use LLVM/Clang variables got from their config -include_directories(${CMAKE_SOURCE_DIR}/compiler/llvm/tools/clang/include) -include_directories(${CMAKE_BINARY_DIR}/compiler/llvm/tools/clang/include) -include_directories(${CMAKE_SOURCE_DIR}/compiler/llvm/tools/lld/include) - -# TODO: move AMDGPU.h header to include folder -include_directories(${CMAKE_SOURCE_DIR}/compiler/llvm/lib/Target/AMDGPU) -include_directories(${CMAKE_BINARY_DIR}/compiler/llvm/lib/Target/AMDGPU) - -if(${USE_COMGR_LIBRARY} MATCHES "yes") - set(COMGR_DYN_DLL "yes") - add_definitions(-DCOMGR_DYN_DLL="yes") - include_directories(${CMAKE_SOURCE_DIR}/compiler/llvm/include) - add_definitions(-DUSE_COMGR_LIBRARY) -else() - add_subdirectory(compiler/driver EXCLUDE_FROM_ALL) -endif() +set(USE_COMGR_LIBRARY "yes") +find_package(amd_comgr REQUIRED CONFIG) +add_definitions(-DUSE_COMGR_LIBRARY) +FOREACH(DIR ${LLVM_INCLUDE_DIRS}) + include_directories("${DIR}") + include_directories("${DIR}/clang") + include_directories("${DIR}/lld") + # TODO: move AMDGPU.h header to include folder + include_directories("${DIR}/llvm/Target/AMDGPU") +ENDFOREACH() set(BUILD_HC_LIB OFF CACHE BOOL "") set(ROCM_DEVICELIB_INCLUDE_TESTS OFF CACHE BOOL "") set(AMDGCN_TARGETS_LIB_LIST "AMDGCN_LIB_TARGETS") set(AMDGCN_TARGETS_LIB_DEPS "AMDGCN_DEP_TARGETS") set(AMDGPU_TARGET_TRIPLE "amdgcn-amd-amdhsa-amdgizcl") -add_subdirectory(library/amdgcn EXCLUDE_FROM_ALL) add_subdirectory(compiler/lib/loaders/elf/utils/libelf) @@ -82,28 +74,6 @@ set(OPENCL_INCLUDE_DIRS ${CMAKE_SOURCE_DIR}/api/opencl/khronos/headers/opencl2.2) add_subdirectory(api/opencl/amdocl) -add_subdirectory(api/opencl/khronos/icd) - -add_subdirectory(tools/clinfo) - -install(PROGRAMS $ $ - DESTINATION bin/x86_64 - COMPONENT libraries) - -install(PROGRAMS $ - DESTINATION lib/x86_64 - COMPONENT applications) - -install(PROGRAMS $ - DESTINATION lib/x86_64 - COMPONENT libraries) - -install(DIRECTORY - "${CMAKE_CURRENT_SOURCE_DIR}/api/opencl/khronos/headers/opencl2.2/CL" - DESTINATION include - COMPONENT libraries - USE_SOURCE_PERMISSIONS - PATTERN cl_egl.h EXCLUDE) foreach(AMDGCN_LIB_TARGET ${AMDGCN_LIB_TARGETS}) get_target_property(lib_file_name ${AMDGCN_LIB_TARGET} ARCHIVE_OUTPUT_NAME) diff --git a/README.md b/README.md index a834965..e319be2 100644 --- a/README.md +++ b/README.md @@ -7,13 +7,6 @@ Developer preview Version 2 of the new * Supports offline ahead of time compilation today; during the Beta phase we will add in-process/in-memory compilation. -## GETTING REPO - -Repo is a git wrapper that manages a collection of git repositories. Install this tool and add it to the command search PATH: - - curl https://storage.googleapis.com/git-repo-downloads/repo > ~/bin/repo - chmod a+x ~/bin/repo - ## GETTING THE SOURCE CODE Main OpenCL™ Compatible Components: @@ -26,11 +19,6 @@ Main OpenCL™ Compatible Components: * https://github.com/RadeonOpenCompute/lld * https://github.com/KhronosGroup/OpenCL-ICD-Loader -Download the git projects with the following commands: - - ~/bin/repo init -u https://github.com/RadeonOpenCompute/ROCm-OpenCL-Runtime.git -b master -m opencl.xml - ~/bin/repo sync - ## INSTALL ROCm Follow the instructions at https://rocm.github.io/install.html to install ROCm. @@ -45,6 +33,11 @@ Copy the amdocl64.icd file to /etc/OpenCL/vendors To install additional dependencies: +* ROCm-OpenCL-Runtime +* ROCm-OpenCL-Driver +* ROC versions of LLVM, Clang, and lld +* ROCm-Device-Libs +* OpenCL-ICD-Loader * OCaml * findlib * A Python 2 environment or active virtualenv with the Microsoft Z3 package diff --git a/api/opencl/amdocl/CMakeLists.txt b/api/opencl/amdocl/CMakeLists.txt index 91bd42f..8606c98 100644 --- a/api/opencl/amdocl/CMakeLists.txt +++ b/api/opencl/amdocl/CMakeLists.txt @@ -21,7 +21,6 @@ include_directories(${CMAKE_SOURCE_DIR}/api/opencl) include_directories(${CMAKE_SOURCE_DIR}/api/opencl/khronos) include_directories(${CMAKE_SOURCE_DIR}/api/opencl/khronos/headers) include_directories(${CMAKE_SOURCE_DIR}/api/opencl/khronos/headers/opencl2.2) -include_directories(${CMAKE_SOURCE_DIR}/compiler/driver/src) include_directories(${CMAKE_SOURCE_DIR}/compiler/lib) include_directories(${CMAKE_SOURCE_DIR}/compiler/lib/include) include_directories(${CMAKE_SOURCE_DIR}/compiler/lib/backends/common) @@ -67,6 +66,9 @@ add_library(amdocl64 SHARED ${ADDITIONAL_LIBRARIES} ) -target_link_libraries(amdocl64 opencl_driver oclelf pthread dl ${ROCR_LIBRARIES}) +target_link_libraries(amdocl64 opencl_driver oclelf pthread dl ${ROCR_LIBRARIES} LLVMBitWriter LLVMIRReader LLVMLinker LLVMMCParser amd_comgr) -install(TARGETS amdocl64 LIBRARY DESTINATION lib/x86_64 COMPONENT applications) +install(TARGETS amdocl64 LIBRARY DESTINATION ${CMAKE_INSTALL_LIBDIR} COMPONENT applications) + +file(GENERATE OUTPUT "${CMAKE_BINARY_DIR}/amdocl64.icd" CONTENT "$") +install(FILES "${CMAKE_BINARY_DIR}/amdocl64.icd" DESTINATION "${CMAKE_INSTALL_FULL_SYSCONFDIR}/OpenCL/vendors/") diff --git a/api/opencl/amdocl/cl_execute.cpp b/api/opencl/amdocl/cl_execute.cpp index 0336353..ee799e6 100644 --- a/api/opencl/amdocl/cl_execute.cpp +++ b/api/opencl/amdocl/cl_execute.cpp @@ -10,7 +10,7 @@ #include "platform/program.hpp" #include "os/os.hpp" -#include +#include /*! \addtogroup API * @{ diff --git a/api/opencl/amdocl/cl_icd.cpp b/api/opencl/amdocl/cl_icd.cpp index 71b886c..7e4e317 100644 --- a/api/opencl/amdocl/cl_icd.cpp +++ b/api/opencl/amdocl/cl_icd.cpp @@ -10,7 +10,7 @@ #include "cl_d3d11_amd.hpp" #endif //_WIN32 -#include +#include #include diff --git a/opencl.xml b/opencl.xml deleted file mode 100644 index 4813651..0000000 --- a/opencl.xml +++ /dev/null @@ -1,17 +0,0 @@ - - - - - - - - - - - - - - - - - diff --git a/runtime/CMakeLists.txt b/runtime/CMakeLists.txt index 62e2bf7..5e4abae 100644 --- a/runtime/CMakeLists.txt +++ b/runtime/CMakeLists.txt @@ -14,8 +14,6 @@ include_directories(${CMAKE_SOURCE_DIR}/api/opencl) include_directories(${CMAKE_SOURCE_DIR}/api/opencl/khronos) include_directories(${CMAKE_SOURCE_DIR}/api/opencl/khronos/headers) include_directories(${CMAKE_SOURCE_DIR}/api/opencl/khronos/headers/opencl2.2) -include_directories(${CMAKE_SOURCE_DIR}/compiler/llvm/include) -include_directories(${CMAKE_SOURCE_DIR}/compiler/driver/src) include_directories(${CMAKE_SOURCE_DIR}/compiler/lib) include_directories(${CMAKE_SOURCE_DIR}/compiler/lib/include) include_directories(${CMAKE_SOURCE_DIR}/compiler/lib/backends/common) @@ -36,6 +34,7 @@ add_library(oclruntime OBJECT utils/flags.cpp utils/debug.cpp device/appprofile.cpp + device/comgrctx.cpp device/device.cpp device/hwdebug.cpp device/blitcl.cpp @@ -63,6 +62,8 @@ add_library(oclruntime OBJECT ${CMAKE_SOURCE_DIR}/compiler/tools/caching/cache.cpp ) set_target_properties(oclruntime PROPERTIES POSITION_INDEPENDENT_CODE ON) +add_dependencies(oclruntime opencl1.2-c.amdgcn.inc_target) +add_dependencies(oclruntime opencl2.0-c.amdgcn.inc_target) set(AMDGCN_DEP_LIST) if(AMDGCN_TARGETS_LIB_DEPS) diff --git a/runtime/device/device.hpp b/runtime/device/device.hpp index 450cd69..8134fa2 100644 --- a/runtime/device/device.hpp +++ b/runtime/device/device.hpp @@ -21,7 +21,7 @@ #if defined(WITH_LIGHTNING_COMPILER) && !defined(USE_COMGR_LIBRARY) #include "caching/cache.hpp" -#include "driver/AmdCompiler.h" +#include "AmdCompiler.h" #endif // defined(WITH_LIGHTNING_COMPILER) && ! defined(USE_COMGR_LIBRARY) #include "acl.h" diff --git a/runtime/device/devprogram.cpp b/runtime/device/devprogram.cpp index 22fb125..3d475b2 100644 --- a/runtime/device/devprogram.cpp +++ b/runtime/device/devprogram.cpp @@ -14,7 +14,7 @@ #if defined(WITH_LIGHTNING_COMPILER) || defined(USE_COMGR_LIBRARY) #ifndef USE_COMGR_LIBRARY -#include "driver/AmdCompiler.h" +#include "AmdCompiler.h" #include "libraries.amdgcn.inc" #include "opencl1.2-c.amdgcn.inc" #include "opencl2.0-c.amdgcn.inc" diff --git a/runtime/device/devprogram.hpp b/runtime/device/devprogram.hpp index 67af239..8390e7d 100644 --- a/runtime/device/devprogram.hpp +++ b/runtime/device/devprogram.hpp @@ -11,7 +11,7 @@ #if defined(WITH_LIGHTNING_COMPILER) || defined(USE_COMGR_LIBRARY) #ifndef USE_COMGR_LIBRARY -#include "driver/AmdCompiler.h" +#include "AmdCompiler.h" #else #include "amd_comgr.h" #endif diff --git a/runtime/device/rocm/CMakeLists.txt b/runtime/device/rocm/CMakeLists.txt index 4ce7c45..617bece 100644 --- a/runtime/device/rocm/CMakeLists.txt +++ b/runtime/device/rocm/CMakeLists.txt @@ -1,3 +1,21 @@ +find_package(Clang REQUIRED CONFIG) + +# FIXME: CLANG_CMAKE_DIR seems like the most stable way to find this, but +# really there is no way to reliably discover this header. +# +# We effectively back up to the Clang output directory (for the case of a build +# tree) or install prefix (for the case of an installed copy), and then search +# for a file named opencl-c.h anywhere below that. We take the first result in +# the case where there are multiple (e.g. if there is an installed copy nested +# in a build directory). This is a bit imprecise, but it covers cases like MSVC +# adding some additional configuration-specific subdirectories to the build +# tree but not to an installed copy. +file(GLOB_RECURSE OPENCL_C_H_LIST "${CLANG_CMAKE_DIR}/../../../*/opencl-c.h") +list(GET OPENCL_C_H_LIST 0 OPENCL_C_H) +if (NOT EXISTS "${OPENCL_C_H}" OR IS_DIRECTORY "${OPENCL_C_H}") + message(FATAL_ERROR "Unable to locate opencl-c.h from the supplied Clang. The path '${CLANG_CMAKE_DIR}/../../../*' was searched.") +endif() + include(bc2h) set(INC_SUFFIX "amdgcn.inc") @@ -74,8 +92,8 @@ endforeach() # generating opencl*.inc files add_custom_command(OUTPUT ${CMAKE_CURRENT_BINARY_DIR}/opencl1.2-c.amdgcn.pch - COMMAND clang -cc1 -x cl-header -triple amdgcn-amd-amdhsa -Werror -O3 -DNDEBUG -cl-std=CL1.2 -emit-pch -o ${CMAKE_CURRENT_BINARY_DIR}/opencl1.2-c.amdgcn.pch < ${CMAKE_SOURCE_DIR}/compiler/llvm/tools/clang/lib/Headers/opencl-c.h - DEPENDS clang ${CMAKE_SOURCE_DIR}/compiler/llvm/tools/clang/lib/Headers/opencl-c.h + COMMAND clang -cc1 -x cl-header -triple amdgcn-amd-amdhsa -Werror -O3 -DNDEBUG -cl-std=CL1.2 -emit-pch -o ${CMAKE_CURRENT_BINARY_DIR}/opencl1.2-c.amdgcn.pch < ${OPENCL_C_H} + DEPENDS clang ${OPENCL_C_H} COMMENT "Generating opencl1.2-c.amdgcn.pch" ) add_custom_command(OUTPUT ${CMAKE_CURRENT_BINARY_DIR}/opencl1.2-c.amdgcn.inc @@ -88,8 +106,8 @@ add_custom_target(opencl1.2-c.amdgcn.inc_target ALL DEPENDS ${CMAKE_CURRENT_BINA add_dependencies(oclrocm opencl1.2-c.amdgcn.inc_target) add_custom_command(OUTPUT ${CMAKE_CURRENT_BINARY_DIR}/opencl2.0-c.amdgcn.pch - COMMAND clang -cc1 -x cl-header -triple amdgcn-amd-amdhsa -Werror -O3 -DNDEBUG -cl-std=CL2.0 -emit-pch -o ${CMAKE_CURRENT_BINARY_DIR}/opencl2.0-c.amdgcn.pch < ${CMAKE_SOURCE_DIR}/compiler/llvm/tools/clang/lib/Headers/opencl-c.h - DEPENDS clang ${CMAKE_SOURCE_DIR}/compiler/llvm/tools/clang/lib/Headers/opencl-c.h + COMMAND clang -cc1 -x cl-header -triple amdgcn-amd-amdhsa -Werror -O3 -DNDEBUG -cl-std=CL2.0 -emit-pch -o ${CMAKE_CURRENT_BINARY_DIR}/opencl2.0-c.amdgcn.pch < ${OPENCL_C_H} + DEPENDS clang ${OPENCL_C_H} COMMENT "Generating opencl2.0-c.amdgcn.pch" ) add_custom_command(OUTPUT ${CMAKE_CURRENT_BINARY_DIR}/opencl2.0-c.amdgcn.inc diff --git a/runtime/device/rocm/rocdevice.cpp b/runtime/device/rocm/rocdevice.cpp index 8b9901b..2648459 100644 --- a/runtime/device/rocm/rocdevice.cpp +++ b/runtime/device/rocm/rocdevice.cpp @@ -20,7 +20,7 @@ #include "device/rocm/rocvirtual.hpp" #include "device/rocm/rocprogram.hpp" #if defined(WITH_LIGHTNING_COMPILER) && ! defined(USE_COMGR_LIBRARY) -#include "driver/AmdCompiler.h" +#include "AmdCompiler.h" #endif // defined(WITH_LIGHTNING_COMPILER) && ! defined(USE_COMGR_LIBRARY) #include "device/rocm/rocmemory.hpp" #include "device/rocm/rocglinterop.hpp" diff --git a/runtime/device/rocm/rockernel.cpp b/runtime/device/rocm/rockernel.cpp index 8a28acc..7d5b62d 100644 --- a/runtime/device/rocm/rockernel.cpp +++ b/runtime/device/rocm/rockernel.cpp @@ -11,7 +11,7 @@ #if defined(WITH_LIGHTNING_COMPILER) || defined(USE_COMGR_LIBRARY) #ifndef USE_COMGR_LIBRARY -#include "driver/AmdCompiler.h" +#include "AmdCompiler.h" #endif #include "llvm/Support/AMDGPUMetadata.h" diff --git a/runtime/device/rocm/rocprogram.cpp b/runtime/device/rocm/rocprogram.cpp index 191e58c..58d82e6 100644 --- a/runtime/device/rocm/rocprogram.cpp +++ b/runtime/device/rocm/rocprogram.cpp @@ -11,7 +11,7 @@ #include #include "libraries.amdgcn.inc" #ifndef USE_COMGR_LIBRARY -#include "driver/AmdCompiler.h" +#include "AmdCompiler.h" #endif #endif // defined(WITH_LIGHTNING_COMPILER) || defined(USE_COMGR_LIBRARY) diff --git a/runtime/platform/object.hpp b/runtime/platform/object.hpp index 8ab1b68..b33a9e3 100644 --- a/runtime/platform/object.hpp +++ b/runtime/platform/object.hpp @@ -9,7 +9,7 @@ #include "os/alloc.hpp" #include "thread/monitor.hpp" #include "utils/util.hpp" -#include +#include #define KHR_CL_TYPES_DO(F) \ diff --git a/tools/clinfo/CMakeLists.txt b/tools/clinfo/CMakeLists.txt deleted file mode 100644 index 974a46f..0000000 --- a/tools/clinfo/CMakeLists.txt +++ /dev/null @@ -1,13 +0,0 @@ -set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -std=c++11") - -set (CMAKE_RUNTIME_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR}/bin) - -include_directories(${OPENCL_INCLUDE_DIRS}) - -add_definitions(-DHAVE_CL2_HPP) - -add_executable(clinfo clinfo.cpp) - -target_link_libraries(clinfo OpenCL) - -install(TARGETS clinfo DESTINATION bin) diff --git a/tools/clinfo/clinfo.cpp b/tools/clinfo/clinfo.cpp deleted file mode 100644 index c5afb7f..0000000 --- a/tools/clinfo/clinfo.cpp +++ /dev/null @@ -1,832 +0,0 @@ -/* ============================================================ - -Copyright (c) 2010 Advanced Micro Devices, Inc. All rights reserved. - -Redistribution and use of this material is permitted under the following -conditions: - -Redistributions must retain the above copyright notice and all terms of this -license. - -In no event shall anyone redistributing or accessing or using this material -commence or participate in any arbitration or legal action relating to this -material against Advanced Micro Devices, Inc. or any copyright holders or -contributors. The foregoing shall survive any expiration or termination of -this license or any agreement or access or use related to this material. - -ANY BREACH OF ANY TERM OF THIS LICENSE SHALL RESULT IN THE IMMEDIATE REVOCATION -OF ALL RIGHTS TO REDISTRIBUTE, ACCESS OR USE THIS MATERIAL. - -THIS MATERIAL IS PROVIDED BY ADVANCED MICRO DEVICES, INC. AND ANY COPYRIGHT -HOLDERS AND CONTRIBUTORS "AS IS" IN ITS CURRENT CONDITION AND WITHOUT ANY -REPRESENTATIONS, GUARANTEE, OR WARRANTY OF ANY KIND OR IN ANY WAY RELATED TO -SUPPORT, INDEMNITY, ERROR FREE OR UNINTERRUPTED OPERA TION, OR THAT IT IS FREE -FROM DEFECTS OR VIRUSES. ALL OBLIGATIONS ARE HEREBY DISCLAIMED - WHETHER -EXPRESS, IMPLIED, OR STATUTORY - INCLUDING, BUT NOT LIMITED TO, ANY IMPLIED -WARRANTIES OF TITLE, MERCHANTABILITY, FITNESS FOR A PARTICULAR PURPOSE, -ACCURACY, COMPLETENESS, OPERABILITY, QUALITY OF SERVICE, OR NON-INFRINGEMENT. -IN NO EVENT SHALL ADVANCED MICRO DEVICES, INC. OR ANY COPYRIGHT HOLDERS OR -CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, PUNITIVE, -EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT -OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, REVENUE, DATA, OR PROFITS; OR -BUSINESS INTERRUPTION) HOWEVER CAUSED OR BASED ON ANY THEORY OF LIABILITY -ARISING IN ANY WAY RELATED TO THIS MATERIAL, EVEN IF ADVISED OF THE POSSIBILITY -OF SUCH DAMAGE. THE ENTIRE AND AGGREGATE LIABILITY OF ADVANCED MICRO DEVICES, -INC. AND ANY COPYRIGHT HOLDERS AND CONTRIBUTORS SHALL NOT EXCEED TEN DOLLARS -(US $10.00). ANYONE REDISTRIBUTING OR ACCESSING OR USING THIS MATERIAL ACCEPTS -THIS ALLOCATION OF RISK AND AGREES TO RELEASE ADVANCED MICRO DEVICES, INC. AND -ANY COPYRIGHT HOLDERS AND CONTRIBUTORS FROM ANY AND ALL LIABILITIES, -OBLIGATIONS, CLAIMS, OR DEMANDS IN EXCESS OF TEN DOLLARS (US $10.00). THE -FOREGOING ARE ESSENTIAL TERMS OF THIS LICENSE AND, IF ANY OF THESE TERMS ARE -CONSTRUED AS UNENFORCEABLE, FAIL IN ESSENTIAL PURPOSE, OR BECOME VOID OR -DETRIMENTAL TO ADVANCED MICRO DEVICES, INC. OR ANY COPYRIGHT HOLDERS OR -CONTRIBUTORS FOR ANY REASON, THEN ALL RIGHTS TO REDISTRIBUTE, ACCESS OR USE -THIS MATERIAL SHALL TERMINATE IMMEDIATELY. MOREOVER, THE FOREGOING SHALL -SURVIVE ANY EXPIRATION OR TERMINATION OF THIS LICENSE OR ANY AGREEMENT OR -ACCESS OR USE RELATED TO THIS MATERIAL. - -NOTICE IS HEREBY PROVIDED, AND BY REDISTRIBUTING OR ACCESSING OR USING THIS -MATERIAL SUCH NOTICE IS ACKNOWLEDGED, THAT THIS MATERIAL MAY BE SUBJECT TO -RESTRICTIONS UNDER THE LAWS AND REGULATIONS OF THE UNITED STATES OR OTHER -COUNTRIES, WHICH INCLUDE BUT ARE NOT LIMITED TO, U.S. EXPORT CONTROL LAWS SUCH -AS THE EXPORT ADMINISTRATION REGULATIONS AND NATIONAL SECURITY CONTROLS AS -DEFINED THEREUNDER, AS WELL AS STATE DEPARTMENT CONTROLS UNDER THE U.S. -MUNITIONS LIST. THIS MATERIAL MAY NOT BE USED, RELEASED, TRANSFERRED, IMPORTED, -EXPORTED AND/OR RE-EXPORTED IN ANY MANNER PROHIBITED UNDER ANY APPLICABLE LAWS, -INCLUDING U.S. EXPORT CONTROL LAWS REGARDING SPECIFICALLY DESIGNATED PERSONS, -COUNTRIES AND NATIONALS OF COUNTRIES SUBJECT TO NATIONAL SECURITY CONTROLS. -MOREOVER, THE FOREGOING SHALL SURVIVE ANY EXPIRATION OR TERMINATION OF ANY -LICENSE OR AGREEMENT OR ACCESS OR USE RELATED TO THIS MATERIAL. - -NOTICE REGARDING THE U.S. GOVERNMENT AND DOD AGENCIES: This material is -provided with "RESTRICTED RIGHTS" and/or "LIMITED RIGHTS" as applicable to -computer software and technical data, respectively. Use, duplication, -distribution or disclosure by the U.S. Government and/or DOD agencies is -subject to the full extent of restrictions in all applicable regulations, -including those found at FAR52.227 and DFARS252.227 et seq. and any successor -regulations thereof. Use of this material by the U.S. Government and/or DOD -agencies is acknowledgment of the proprietary rights of any copyright holders -and contributors, including those of Advanced Micro Devices, Inc., as well as -the provisions of FAR52.227-14 through 23 regarding privately developed and/or -commercial computer software. - -This license forms the entire agreement regarding the subject matter hereof and -supersedes all proposals and prior discussions and writings between the parties -with respect thereto. This license does not affect any ownership, rights, title, -or interest in, or relating to, this material. No terms of this license can be -modified or waived, and no breach of this license can be excused, unless done -so in a writing signed by all affected parties. Each term of this license is -separately enforceable. If any term of this license is determined to be or -becomes unenforceable or illegal, such term shall be reformed to the minimum -extent necessary in order for this license to remain in effect in accordance -with its terms as modified by such reformation. This license shall be governed -by and construed in accordance with the laws of the State of Texas without -regard to rules on conflicts of law of any state or jurisdiction or the United -Nations Convention on the International Sale of Goods. All disputes arising out -of this license shall be subject to the jurisdiction of the federal and state -courts in Austin, Texas, and all defenses are hereby waived concerning personal -jurisdiction and venue of these courts. - -============================================================ */ - -#include -#include -#include -#include -#include -#include -#include -#if !defined(_WIN32) -#include -#endif - -#ifdef _MSC_VER -#pragma warning(disable: 4290) -#endif - -#if defined(HAVE_CL2_HPP) -#define CL_HPP_ENABLE_EXCEPTIONS -#define CL_HPP_MINIMUM_OPENCL_VERSION 120 -#define CL_HPP_TARGET_OPENCL_VERSION 200 -#define CL_HPP_ENABLE_PROGRAM_CONSTRUCTION_FROM_ARRAY_COMPATIBILITY -#include "CL/cl2.hpp" -#else // !HAVE_CL2_HPP -#define __CL_ENABLE_EXCEPTIONS -#define __MAX_DEFAULT_VECTOR_SIZE 50 -#define CL_USE_DEPRECATED_OPENCL_1_1_APIS -#define CL_USE_DEPRECATED_OPENCL_2_0_APIS -#include "cl.hpp" -#endif // !HAVE_CL2_HPP - -bool verbose = false; - -/// Returns EXIT_SUCCESS on success, EXIT_FAILURE on failure. -int -main(int argc, char** argv) -{ - /* Error flag */ - cl_int err; - - //parse input - for(int i = 1; i < argc; i++){ - if ((strcmp(argv[i], "-v") == 0) || - (strcmp(argv[i], "--verbose") == 0)){ - verbose = true; - } else if ((strcmp(argv[i], "-h") == 0) || - (strcmp(argv[i], "--help") == 0)){ - std::cout << "Usage is: " << argv[0] << " [-v|--verbose]" << std::endl; - return EXIT_FAILURE; - } - } - - // Platform info - std::vector platforms; - - try { - err = cl::Platform::get(&platforms); - - // Iteratate over platforms - std::cout << "Number of platforms:\t\t\t\t " - << platforms.size() - << std::endl; - for (std::vector::iterator i = platforms.begin(); - i != platforms.end(); - ++i) { - const cl::Platform& platform = *i; - - std::cout << " Platform Profile:\t\t\t\t " - << platform.getInfo().c_str() - << std::endl; - std::cout << " Platform Version:\t\t\t\t " - << platform.getInfo().c_str() - << std::endl; - std::cout << " Platform Name:\t\t\t\t " - << platform.getInfo().c_str() - << std::endl; - std::cout << " Platform Vendor:\t\t\t\t " - << platform.getInfo().c_str() << std::endl; - if (platform.getInfo().size() > 0) { - std::cout << " Platform Extensions:\t\t\t\t " - << platform.getInfo().c_str() - << std::endl; - } - } - - std::cout << std::endl << std:: endl; - // Now Iteratate over each platform and its devices - for (std::vector::iterator p = platforms.begin(); - p != platforms.end(); - ++p) { - const cl::Platform& platform = *p; - std::cout << " Platform Name:\t\t\t\t " - << platform.getInfo().c_str() - << std::endl; - - std::vector devices; - platform.getDevices(CL_DEVICE_TYPE_ALL, &devices); - - // Get OpenCL version - std::string platformVersionStr = platform.getInfo(); - std::string openclVerstionStr(platformVersionStr.c_str()); - size_t vStart = openclVerstionStr.find(" ", 0); - size_t vEnd = openclVerstionStr.find(" ", vStart + 1); - std::string vStrVal = openclVerstionStr.substr(vStart + 1, vEnd - vStart - 1); - - std::cout << "Number of devices:\t\t\t\t " << devices.size() << std::endl; - for (std::vector::iterator i = devices.begin(); - i != devices.end(); - ++i) { - const cl::Device& device = *i; - /* Get device name */ - std::string deviceName = device.getInfo(); - cl_device_type dtype = device.getInfo(); - - /* Get CAL driver version in int */ - std::string driverVersion = device.getInfo(); - std::string calVersion(driverVersion.c_str()); - calVersion = calVersion.substr(calVersion.find_last_of(".") + 1); - int version = atoi(calVersion.c_str()); - - std::cout << " Device Type:\t\t\t\t\t " ; - switch (dtype) { - case CL_DEVICE_TYPE_ACCELERATOR: - std::cout << "CL_DEVICE_TYPE_ACCRLERATOR" << std::endl; - break; - case CL_DEVICE_TYPE_CPU: - std::cout << "CL_DEVICE_TYPE_CPU" << std::endl; - break; - case CL_DEVICE_TYPE_DEFAULT: - std::cout << "CL_DEVICE_TYPE_DEFAULT" << std::endl; - break; - case CL_DEVICE_TYPE_GPU: - std::cout << "CL_DEVICE_TYPE_GPU" << std::endl; - break; - } - - std::cout << " Vendor ID:\t\t\t\t\t " - << std::hex - << device.getInfo() - << "h" - << std::dec - << std::endl; - - bool isAMDPlatform = (strcmp(platform.getInfo().c_str(), "AMD Accelerated Parallel Processing") == 0) ? true : false; - if (isAMDPlatform) - { - std::string boardName; - device.getInfo(CL_DEVICE_BOARD_NAME_AMD, &boardName); - std::cout << " Board name:\t\t\t\t\t " - << boardName.c_str() - << std::endl; - - cl_device_topology_amd topology; - err = device.getInfo(CL_DEVICE_TOPOLOGY_AMD, &topology); - if (topology.raw.type == CL_DEVICE_TOPOLOGY_TYPE_PCIE_AMD) { - std::cout << " Device Topology:\t\t\t\t " - << "PCI[ B#" << (int)topology.pcie.bus - << ", D#" << (int)topology.pcie.device - << ", F#" << (int)topology.pcie.function - << " ]" << std::endl; - } - } - - std::cout << " Max compute units:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max work items dimensions:\t\t\t " - << device.getInfo() - << std::endl; - - std::vector< ::size_t> witems = - device.getInfo(); - for (unsigned int x = 0; - x < device.getInfo(); - x++) { - std::cout << " Max work items[" - << x << "]:\t\t\t\t " - << witems[x] - << std::endl; - } - - std::cout << " Max work group size:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Preferred vector width char:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Preferred vector width short:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Preferred vector width int:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Preferred vector width long:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Preferred vector width float:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Preferred vector width double:\t\t " - << device.getInfo() - << std::endl; - -#ifdef CL_VERSION_1_1 - if(vStrVal.compare("1.0") > 0) - { - std::cout << " Native vector width char:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Native vector width short:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Native vector width int:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Native vector width long:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Native vector width float:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Native vector width double:\t\t\t " - << device.getInfo() - << std::endl; - } -#endif // CL_VERSION_1_1 - std::cout << " Max clock frequency:\t\t\t\t " - << device.getInfo() - << "Mhz" - << std::endl; - - std::cout << " Address bits:\t\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max memory allocation:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Image support:\t\t\t\t " - << (device.getInfo() ? "Yes" : "No") - << std::endl; - - if (device.getInfo()) - { - - std::cout << " Max number of images read arguments:\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max number of images write arguments:\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max image 2D width:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max image 2D height:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max image 3D width:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max image 3D height:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max image 3D depth:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max samplers within kernel:\t\t\t " - << device.getInfo() - << std::endl; - - if (verbose) - { - std::cout << " Image formats supported:" << std::endl; - std::vector formats; - - cl_context_properties cps[3] = { CL_CONTEXT_PLATFORM, (cl_context_properties)(*p)(), 0 }; - std::vector device; - device.push_back(*i); - cl::Context context(device, cps, NULL, NULL, &err); - - std::map channelOrder; - channelOrder[CL_R] = "CL_R"; - channelOrder[CL_A] = "CL_A"; - channelOrder[CL_RG] = "CL_RG"; - channelOrder[CL_RA] = "CL_RA"; - channelOrder[CL_RGB] = "CL_RGB"; - channelOrder[CL_RGBA] = "CL_RGBA"; - channelOrder[CL_BGRA] = "CL_BGRA"; - channelOrder[CL_ARGB] = "CL_ARGB"; - channelOrder[CL_INTENSITY] = "CL_INTENSITY"; - channelOrder[CL_LUMINANCE] = "CL_LUMINANCE"; - channelOrder[CL_Rx] = "CL_Rx"; - channelOrder[CL_RGx] = "CL_RGx"; - channelOrder[CL_RGBx] = "CL_RGBx"; - - std::map > channelType; - channelType[CL_SNORM_INT8] = std::make_pair("snorm", "int8"); - channelType[CL_SNORM_INT16] = std::make_pair("snorm", "int16"); - channelType[CL_UNORM_INT8] = std::make_pair("unorm", "int8"); - channelType[CL_UNORM_INT16] = std::make_pair("unorm", "int16"); - channelType[CL_UNORM_SHORT_565] = std::make_pair("unorm", "short_565"); - channelType[CL_UNORM_SHORT_555] = std::make_pair("unorm", "short_555"); - channelType[CL_UNORM_INT_101010] = std::make_pair("unorm", "int_101010"); - channelType[CL_SIGNED_INT8] = std::make_pair("signed", "int8"); - channelType[CL_SIGNED_INT16] = std::make_pair("signed", "int16"); - channelType[CL_SIGNED_INT32] = std::make_pair("signed", "int32"); - channelType[CL_UNSIGNED_INT8] = std::make_pair("unsigned", "int8"); - channelType[CL_UNSIGNED_INT16] = std::make_pair("unsigned", "int16"); - channelType[CL_UNSIGNED_INT32] = std::make_pair("unsigned", "int32"); - channelType[CL_HALF_FLOAT] = std::make_pair("half_float", ""); - channelType[CL_FLOAT] = std::make_pair("float", ""); - - std::vector > imageDimensions; - imageDimensions.push_back(std::make_pair(CL_MEM_OBJECT_IMAGE2D, std::string("2D "))); - imageDimensions.push_back(std::make_pair(CL_MEM_OBJECT_IMAGE3D, std::string("3D "))); - for(std::vector >::iterator id = imageDimensions.begin(); - id != imageDimensions.end(); - id++){ - - struct imageAccessStruct { - std::string name; - int access; - std::vector formats; - } imageAccess[] = {{std::string("Read-Write/Read-Only/Write-Only"), CL_MEM_READ_WRITE, std::vector()}, - {std::string("Read-Only"), CL_MEM_READ_ONLY, std::vector()}, - {std::string("Write-Only"), CL_MEM_WRITE_ONLY, std::vector()}}; - - for(size_t ia=0; ia < sizeof(imageAccess)/sizeof(imageAccessStruct); ia++){ - context.getSupportedImageFormats(imageAccess[ia].access, (*id).first, &(imageAccess[ia].formats)); - bool printTopHeader = true; - for (std::map::iterator o = channelOrder.begin(); - o != channelOrder.end(); - o++) - { - bool printHeader = true; - - for (std::vector::iterator it = imageAccess[ia].formats.begin(); - it != imageAccess[ia].formats.end(); - ++it) - { - if ( (*o).first == (int)(*it).image_channel_order) - { - bool printedAlready = false; - //see if this was already print in RW/RO/WO - if (ia !=0) - { - for (std::vector::iterator searchIt = imageAccess[0].formats.begin(); - searchIt != imageAccess[0].formats.end(); - searchIt++) - { - if ( ((*searchIt).image_channel_data_type == (*it).image_channel_data_type) && - ((*searchIt).image_channel_order == (*it).image_channel_order)) - { - printedAlready = true; - break; - } - } - } - if (printedAlready) - { - continue; - } - if (printTopHeader) - { - std::cout << " " << (*id).second << imageAccess[ia].name << std::endl; - printTopHeader = false; - } - if (printHeader) - { - std::cout << " " << (*o).second << ": "; - printHeader = false; - } - std::cout << channelType[(*it).image_channel_data_type].first; - if (channelType[(*it).image_channel_data_type].second != "") - { - std::cout << "-" - << channelType[(*it).image_channel_data_type].second; - } - if (it != (imageAccess[ia].formats.end() - 1)) - { - std::cout << " "; - } - } - } - if (printHeader == false) - { - std::cout << std::endl; - } - } - } - } - } - } - - std::cout << " Max size of kernel argument:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Alignment (bits) of base address:\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Minimum alignment (bytes) for any datatype:\t " - << device.getInfo() - << std::endl; - - std::cout << " Single precision floating point capability" << std::endl; - std::cout << " Denorms:\t\t\t\t\t " - << (device.getInfo() & - CL_FP_DENORM ? "Yes" : "No") - << std::endl; - std::cout << " Quiet NaNs:\t\t\t\t\t " - << (device.getInfo() & - CL_FP_INF_NAN ? "Yes" : "No") - << std::endl; - std::cout << " Round to nearest even:\t\t\t " - << (device.getInfo() & - CL_FP_ROUND_TO_NEAREST ? "Yes" : "No") - << std::endl; - std::cout << " Round to zero:\t\t\t\t " - << (device.getInfo() & - CL_FP_ROUND_TO_ZERO ? "Yes" : "No") - << std::endl; - std::cout << " Round to +ve and infinity:\t\t\t " - << (device.getInfo() & - CL_FP_ROUND_TO_INF ? "Yes" : "No") - << std::endl; - std::cout << " IEEE754-2008 fused multiply-add:\t\t " - << (device.getInfo() & - CL_FP_FMA ? "Yes" : "No") - << std::endl; - - std::cout << " Cache type:\t\t\t\t\t " ; - switch (device.getInfo()) { - case CL_NONE: - std::cout << "None" << std::endl; - break; - case CL_READ_ONLY_CACHE: - std::cout << "Read only" << std::endl; - break; - case CL_READ_WRITE_CACHE: - std::cout << "Read/Write" << std::endl; - break; - } - - std::cout << " Cache line size:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Cache size:\t\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Global memory size:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Constant buffer size:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max number of constant args:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Local memory type:\t\t\t\t " ; - switch (device.getInfo()) { - case CL_LOCAL: - std::cout << "Scratchpad" << std::endl; - break; - case CL_GLOBAL: - std::cout << "Global" << std::endl; - break; - } - - - std::cout << " Local memory size:\t\t\t\t " - << device.getInfo() - << std::endl; - -#if defined(CL_VERSION_2_0) - if(vStrVal.compare("2") > 0) - { - std::cout << " Max pipe arguments:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max pipe active reservations:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max pipe packet size:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max global variable size:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max global variable preferred total size:\t " - << device.getInfo() - << std::endl; - - std::cout << " Max read/write image args:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max on device events:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Queue on device max size:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Max on device queues:\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Queue on device preferred size:\t\t " - << device.getInfo() - << std::endl; - - std::cout << " SVM capabilities:\t\t\t\t " << std::endl; - std::cout << " Coarse grain buffer:\t\t\t " - << (device.getInfo() & - CL_DEVICE_SVM_COARSE_GRAIN_BUFFER ? "Yes" : "No") - << std::endl; - std::cout << " Fine grain buffer:\t\t\t\t " - << (device.getInfo() & - CL_DEVICE_SVM_FINE_GRAIN_BUFFER ? "Yes" : "No") - << std::endl; - std::cout << " Fine grain system:\t\t\t\t " - << (device.getInfo() & - CL_DEVICE_SVM_FINE_GRAIN_SYSTEM ? "Yes" : "No") - << std::endl; - std::cout << " Atomics:\t\t\t\t\t " - << (device.getInfo() & - CL_DEVICE_SVM_ATOMICS ? "Yes" : "No") - << std::endl; - - std::cout << " Preferred platform atomic alignment:\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Preferred global atomic alignment:\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Preferred local atomic alignment:\t\t " - << device.getInfo() - << std::endl; - } -#endif // CL_VERSION_2_0 - -#if defined(CL_VERSION_1_1) && !defined(ATI_ARCH_ARM) - if(vStrVal.compare("1.0") > 0) - { - cl_context_properties cps[3] = { CL_CONTEXT_PLATFORM, (cl_context_properties)(*p)(), 0 }; - - std::vector device; - device.push_back(*i); - - cl::Context context(device, cps, NULL, NULL, &err); - if (err != CL_SUCCESS) { - std::cerr << "Context::Context() failed (" << err << ")\n"; - return EXIT_FAILURE; - } - std::string kernelStr("__kernel void hello(){ size_t i = get_global_id(0); size_t j = get_global_id(1);}"); - cl::Program::Sources sources(1, std::make_pair(kernelStr.data(), kernelStr.size())); - - cl::Program program = cl::Program(context, sources, &err); - if (err != CL_SUCCESS) { - std::cerr << "Program::Program() failed (" << err << ")\n"; - return EXIT_FAILURE; - } - - err = program.build(device); - if (err != CL_SUCCESS) { - - if(err == CL_BUILD_PROGRAM_FAILURE) - { - std::string str = program.getBuildInfo((*i)); - - std::cout << " \n\t\t\tBUILD LOG\n"; - std::cout << " ************************************************\n"; - std::cout << str.c_str() << std::endl; - std::cout << " ************************************************\n"; - } - - std::cerr << "Program::build() failed (" << err << ")\n"; - return EXIT_FAILURE; - } - - cl::Kernel kernel(program, "hello", &err); - if (err != CL_SUCCESS) { - std::cerr << "Kernel::Kernel() failed (" << err << ")\n"; - return EXIT_FAILURE; - } - - std::cout << " Kernel Preferred work group size multiple:\t " - << kernel.getWorkGroupInfo((*i), &err) - << std::endl; - } - -#endif // CL_VERSION_1_1 - - std::cout << " Error correction support:\t\t\t " - << device.getInfo() - << std::endl; -#ifdef CL_VERSION_1_1 - if(vStrVal.compare("1.0") > 0) - { - std::cout << " Unified memory for Host and Device:\t\t " - << device.getInfo() - << std::endl; - } -#endif // CL_VERSION_1_1 - std::cout << " Profiling timer resolution:\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Device endianess:\t\t\t\t " - << (device.getInfo() ? "Little" : "Big") - << std::endl; - - std::cout << " Available:\t\t\t\t\t " - << (device.getInfo() ? "Yes" : "No") - << std::endl; - - std::cout << " Compiler available:\t\t\t\t " - << (device.getInfo() ? "Yes" : "No") - << std::endl; - - std::cout << " Execution capabilities:\t\t\t\t " << std::endl; - std::cout << " Execute OpenCL kernels:\t\t\t " - << (device.getInfo() & - CL_EXEC_KERNEL ? "Yes" : "No") - << std::endl; - std::cout << " Execute native function:\t\t\t " - << (device.getInfo() & - CL_EXEC_NATIVE_KERNEL ? "Yes" : "No") - << std::endl; - - std::cout << " Queue on Host properties:\t\t\t\t " << std::endl; - std::cout << " Out-of-Order:\t\t\t\t " - << (device.getInfo() & - CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE ? "Yes" : "No") - << std::endl; - std::cout << " Profiling :\t\t\t\t\t " - << (device.getInfo() & - CL_QUEUE_PROFILING_ENABLE ? "Yes" : "No") - << std::endl; - -#ifdef CL_VERSION_2_0 - if(vStrVal.compare("2") > 0) - { - std::cout << " Queue on Device properties:\t\t\t\t " << std::endl; - std::cout << " Out-of-Order:\t\t\t\t " - << (device.getInfo() & - CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE ? "Yes" : "No") - << std::endl; - std::cout << " Profiling :\t\t\t\t\t " - << (device.getInfo() & - CL_QUEUE_PROFILING_ENABLE ? "Yes" : "No") - << std::endl; - } -#endif - - std::cout << " Platform ID:\t\t\t\t\t " - << device.getInfo() - << std::endl; - - std::cout << " Name:\t\t\t\t\t\t " - << device.getInfo().c_str() - << std::endl; - - std::cout << " Vendor:\t\t\t\t\t " - << device.getInfo().c_str() - << std::endl; -#ifdef CL_VERSION_1_1 - if(vStrVal.compare("1.0") > 0) - { - std::cout << " Device OpenCL C version:\t\t\t " - << device.getInfo().c_str() - << std::endl; - } -#endif // CL_VERSION_1_1 - std::cout << " Driver version:\t\t\t\t " - << device.getInfo().c_str() - << std::endl; - - std::cout << " Profile:\t\t\t\t\t " - << device.getInfo().c_str() - << std::endl; - - std::cout << " Version:\t\t\t\t\t " - << device.getInfo().c_str() - << std::endl; - - - std::cout << " Extensions:\t\t\t\t\t " - << device.getInfo().c_str() - << std::endl; - - std::cout << std::endl << std::endl; - } - } - } - catch (cl::Error err) - { - std::cerr - << "ERROR: " - << err.what() - << "(" - << err.err() - << ")" - << std::endl; - return EXIT_FAILURE; - } - - return EXIT_SUCCESS; -}