diff --git a/hpvm/include/hpvm.h b/hpvm/include/hpvm.h index f76dde028e61d2c940a15fc34e1dda10431a06b8..393c3726c57dfef549c79a8074cb5a7a70f66f40 100644 --- a/hpvm/include/hpvm.h +++ b/hpvm/include/hpvm.h @@ -6,106 +6,110 @@ *cr ***************************************************************************/ -// This changes hpvm::DEVICE (backward-compatible hint name) to hpvm::CUDNN_TARGET. -// hpvm::DEVICE is deprecated; do not use. +// This changes hpvm::DEVICE (backward-compatible hint name) to +// hpvm::CUDNN_TARGET. hpvm::DEVICE is deprecated; do not use. #ifndef DEVICE #define DEVICE CUDNN_TARGET #endif #include "SupportHPVM/HPVMHint.h" +#ifdef __cplusplus +#define CXX_NOEXCEPT noexcept +#else +#define CXX_NOEXCEPT +#endif + #ifdef __cplusplus extern "C" { -void __hpvm__hint(hpvm::Target); +void __hpvm__hint(hpvm::Target) CXX_NOEXCEPT; #else void __hpvm__hint(enum Target); #endif -void *__hpvm__createNodeND(unsigned, ...); -void __hpvm__return(unsigned, ...); - -void __hpvm__attributes(unsigned, ...); -void __hpvm__init(); -void __hpvm__cleanup(); - -void __hpvm__bindIn(void *, unsigned, unsigned, unsigned); -void __hpvm__bindOut(void *, unsigned, unsigned, unsigned); -void *__hpvm__edge(void *, void *, unsigned, unsigned, unsigned, unsigned); -void __hpvm__push(void *, void *); -void *__hpvm__pop(void *); -void *__hpvm__launch(unsigned, ...); -void __hpvm__wait(void *); - -void *__hpvm__getNode(); -void *__hpvm__getParentNode(void *); -void __hpvm__barrier(); -void *__hpvm__malloc(long); -long __hpvm__getNodeInstanceID_x(void *); -long __hpvm__getNodeInstanceID_y(void *); -long __hpvm__getNodeInstanceID_z(void *); -long __hpvm__getNumNodeInstances_x(void *); -long __hpvm__getNumNodeInstances_y(void *); -long __hpvm__getNumNodeInstances_z(void *); +void *__hpvm__createNodeND(unsigned, ...) CXX_NOEXCEPT; +void __hpvm__return(unsigned, ...) CXX_NOEXCEPT; + +void __hpvm__attributes(unsigned, ...) CXX_NOEXCEPT; +void __hpvm__init() CXX_NOEXCEPT; +void __hpvm__cleanup() CXX_NOEXCEPT; + +void __hpvm__bindIn(void *, unsigned, unsigned, unsigned) CXX_NOEXCEPT; +void __hpvm__bindOut(void *, unsigned, unsigned, unsigned) CXX_NOEXCEPT; +void *__hpvm__edge(void *, void *, unsigned, unsigned, unsigned, unsigned) + CXX_NOEXCEPT; +void __hpvm__push(void *, void *) CXX_NOEXCEPT; +void *__hpvm__pop(void *) CXX_NOEXCEPT; +void *__hpvm__launch(unsigned, ...) CXX_NOEXCEPT; +void __hpvm__wait(void *) CXX_NOEXCEPT; + +void *__hpvm__getNode() CXX_NOEXCEPT; +void *__hpvm__getParentNode(void *) CXX_NOEXCEPT; +void __hpvm__barrier() CXX_NOEXCEPT; +void *__hpvm__malloc(long) CXX_NOEXCEPT; +long __hpvm__getNodeInstanceID_x(void *) CXX_NOEXCEPT; +long __hpvm__getNodeInstanceID_y(void *) CXX_NOEXCEPT; +long __hpvm__getNodeInstanceID_z(void *) CXX_NOEXCEPT; +long __hpvm__getNumNodeInstances_x(void *) CXX_NOEXCEPT; +long __hpvm__getNumNodeInstances_y(void *) CXX_NOEXCEPT; +long __hpvm__getNumNodeInstances_z(void *) CXX_NOEXCEPT; // Atomic // signed int -int __hpvm__atomic_cmpxchg(int *, int, int); -int __hpvm__atomic_add(int *, int); -int __hpvm__atomic_sub(int *, int); -int __hpvm__atomic_xchg(int *, int); -int __hpvm__atomic_inc(int *); -int __hpvm__atomic_dec(int *); -int __hpvm__atomic_min(int *, int); -int __hpvm__atomic_max(int *, int); -int __hpvm__atomic_umax(int *, int); -int __hpvm__atomic_umin(int *, int); -int __hpvm__atomic_and(int *, int); -int __hpvm__atomic_or(int *, int); -int __hpvm__atomic_xor(int *, int); +int __hpvm__atomic_cmpxchg(int *, int, int) CXX_NOEXCEPT; +int __hpvm__atomic_add(int *, int) CXX_NOEXCEPT; +int __hpvm__atomic_sub(int *, int) CXX_NOEXCEPT; +int __hpvm__atomic_xchg(int *, int) CXX_NOEXCEPT; +int __hpvm__atomic_inc(int *) CXX_NOEXCEPT; +int __hpvm__atomic_dec(int *) CXX_NOEXCEPT; +int __hpvm__atomic_min(int *, int) CXX_NOEXCEPT; +int __hpvm__atomic_max(int *, int) CXX_NOEXCEPT; +int __hpvm__atomic_umax(int *, int) CXX_NOEXCEPT; +int __hpvm__atomic_umin(int *, int) CXX_NOEXCEPT; +int __hpvm__atomic_and(int *, int) CXX_NOEXCEPT; +int __hpvm__atomic_or(int *, int) CXX_NOEXCEPT; +int __hpvm__atomic_xor(int *, int) CXX_NOEXCEPT; // Special Func -float __hpvm__floor(float); -float __hpvm__rsqrt(float); -float __hpvm__sqrt(float); -float __hpvm__sin(float); -float __hpvm__cos(float); +float __hpvm__floor(float) CXX_NOEXCEPT; +float __hpvm__rsqrt(float) CXX_NOEXCEPT; +float __hpvm__sqrt(float) CXX_NOEXCEPT; +float __hpvm__sin(float) CXX_NOEXCEPT; +float __hpvm__cos(float) CXX_NOEXCEPT; /* * ApproxHPVM specific function calls */ -void *__hpvm__tensor_add(void *, void *); -void *__hpvm__tensor_mul(void *, void *); -void *__hpvm__tensor_convolution(void *, void *, int, int, int, int); -void *__hpvm__tensor_group_convolution(void *, void *, int, int, int, int, int, - int); -void *__hpvm__tensor_batchnorm(void *, void *, void *, void *, void *, double); -void *__hpvm__tensor_pool_max(void *, int, int, int, int, int, int); -void *__hpvm__tensor_pool_mean(void *, int, int, int, int, int, int); -void *__hpvm__tensor_relu(void *); -void *__hpvm__tensor_tanh(void *); -void *__hpvm__tensor_softmax(void *); +void *__hpvm__tensor_add(void *, void *) CXX_NOEXCEPT; +void *__hpvm__tensor_mul(void *, void *) CXX_NOEXCEPT; +void * +__hpvm__tensor_convolution(void *, void *, int, int, int, int) CXX_NOEXCEPT; +void *__hpvm__tensor_group_convolution( + void *, void *, int, int, int, int, int, int) CXX_NOEXCEPT; +void *__hpvm__tensor_batchnorm(void *, void *, void *, void *, void *, double) + CXX_NOEXCEPT; +void * +__hpvm__tensor_pool_max(void *, int, int, int, int, int, int) CXX_NOEXCEPT; +void * +__hpvm__tensor_pool_mean(void *, int, int, int, int, int, int) CXX_NOEXCEPT; +void *__hpvm__tensor_relu(void *) CXX_NOEXCEPT; +void *__hpvm__tensor_tanh(void *) CXX_NOEXCEPT; +void *__hpvm__tensor_softmax(void *) CXX_NOEXCEPT; // New HPVM intrinsic for Setting Node ID -void *__hpvm__node_id(int); +void *__hpvm__node_id(int) CXX_NOEXCEPT; #include <unistd.h> -long get_global_id(int); -long get_group_id(int); -long get_local_id(int); -long get_local_size(int); +long get_global_id(int) CXX_NOEXCEPT; +long get_group_id(int) CXX_NOEXCEPT; +long get_local_id(int) CXX_NOEXCEPT; +long get_local_size(int) CXX_NOEXCEPT; -void llvm_hpvm_track_mem(void *, size_t); -void llvm_hpvm_untrack_mem(void *); -void llvm_hpvm_request_mem(void *, size_t); - -/*#ifdef __cplusplus -void hpvm_request_tensor(void*, hpvm::Target); -#else -void hpvm_request_tensor(void*, enum Target); -#endif -*/ +void llvm_hpvm_track_mem(void *, size_t) CXX_NOEXCEPT; +void llvm_hpvm_untrack_mem(void *) CXX_NOEXCEPT; +void llvm_hpvm_request_mem(void *, size_t) CXX_NOEXCEPT; #ifdef __cplusplus } diff --git a/hpvm/projects/torch2hpvm/torch2hpvm/compile.py b/hpvm/projects/torch2hpvm/torch2hpvm/compile.py index a3865a007c86605ba044f91f6fa3d6202bd2110c..12d1c544d0dbd876961bf6567e24657d578b2910 100644 --- a/hpvm/projects/torch2hpvm/torch2hpvm/compile.py +++ b/hpvm/projects/torch2hpvm/torch2hpvm/compile.py @@ -165,7 +165,6 @@ class ModelExporter: args = [ "hpvm-clang", "-O3", - "-fno-exceptions", str(self.codefile), str(output_binary), *self.compile_args, diff --git a/hpvm/test/dnn_benchmarks/hpvm-c/CMakeLists.txt b/hpvm/test/dnn_benchmarks/hpvm-c/CMakeLists.txt index aedf0640025703b62ed5e9a810f5c3d68e800f6f..6d0c480bac5c70d90e96e00667375d0e69f2729a 100644 --- a/hpvm/test/dnn_benchmarks/hpvm-c/CMakeLists.txt +++ b/hpvm/test/dnn_benchmarks/hpvm-c/CMakeLists.txt @@ -9,7 +9,7 @@ function(compile_hpvm_c target_name src_filepath codegen_target) OUTPUT ${output_bin_path} DEPENDS ${src_filepath} hpvm-clang COMMAND hpvm-clang - ${src_filepath} ${output_bin_path} -O3 -fno-exceptions + ${src_filepath} ${output_bin_path} -O3 "-DMODEL_PARAMS_DIR=${MODEL_PARAMS_DIR}" -t ${codegen_target} ${ARGN} ) diff --git a/hpvm/test/dnn_benchmarks/keras/Benchmark.py b/hpvm/test/dnn_benchmarks/keras/Benchmark.py index ae0adc1a51d07788976d7b2d43a5e890ecb70adc..5b25db407f771f7a302a467f46add6227bab552e 100644 --- a/hpvm/test/dnn_benchmarks/keras/Benchmark.py +++ b/hpvm/test/dnn_benchmarks/keras/Benchmark.py @@ -69,8 +69,7 @@ class Benchmark: try: subprocess.run([ "hpvm-clang", src_file, target_binary, - "-t", "tensor", "--conf-file", approx_conf_file, - "-fno-exceptions" + "-t", "tensor", "--conf-file", approx_conf_file ], check=True) except: print ("\n\n ERROR: HPVM Compilation Failed!! \n\n") diff --git a/hpvm/tools/hpvm-clang/main.py.in b/hpvm/tools/hpvm-clang/main.py.in index 0f240ad04f4f8797144bde484b841c60af91cb52..9d1fe40a78da4932d694ede8f3fc159b45930e6a 100644 --- a/hpvm/tools/hpvm-clang/main.py.in +++ b/hpvm/tools/hpvm-clang/main.py.in @@ -41,9 +41,7 @@ def compile_hpvm_c( ): from subprocess import check_output - # FIXME: Added this because Exceptions are not supported in HPVM. - flags = (flags or []) + ["no-exceptions"] - + flags = (flags or []) passes = ["LLVMBuildDFG"] pass_flags = ["buildDFG"] if tensor_target == "tensor":