fi
if [[ "${DOCKER_IMAGE_NAME}" == *"gpu"* ]]; then
- if ! type "nvidia-docker" > /dev/null
+ if ! type "nvidia-docker" 1> /dev/null 2> /dev/null
then
DOCKER_BINARY="docker"
CUDA_ENV=" --gpus all "${CUDA_ENV}
# By default we cleanup - remove the container once it finish running (--rm)
# and share the PID namespace (--pid=host) so the process inside does not have
# pid 1 and SIGKILL is propagated to the process inside (jenkins can kill it).
-echo ${DOCKER_BINARY}
${DOCKER_BINARY} run --rm --pid=host\
-v ${WORKSPACE}:/workspace \
-v ${SCRIPT_DIR}:/docker \
${DOCKER_IMAGE_NAME}\
bash --login /docker/with_the_same_user \
${COMMAND[@]}
+
std::string CodeGenCUDA::Finish() {
if (enable_fp16_) {
decl_stream << "#include <cuda_fp16.h>\n";
- decl_stream << "__device__ half max" \
- "(const half a, const half b)\n"
- "{\n return __hgt(__half(a), __half(b)) ? a : b;\n}\n";
- decl_stream << "__device__ half min(const half a, const half b)\n"
- "{\n return __hlt(__half(a), __half(b)) ? a : b;\n}\n";
- decl_stream << "__device__ half operator+" \
- "(const volatile __half &a, const volatile __half &b)\n"
- "{\n return __hadd(a, b);\n}\n";
- decl_stream << "__device__ half operator<=" \
- "(const volatile __half &a, const volatile __half &b)\n"
- "{\n return __hlt(a, b);\n}\n";
- decl_stream << "__device__ half operator*" \
- "(const volatile __half &a, const volatile __half &b)\n"
- "{\n return __hmul(a, b);\n}\n";
+ decl_stream << "__device__ half max"
+ << "(half a, half b)\n"
+ << "{\n return __hgt(__half(a), __half(b)) ? a : b;\n}\n";
+ decl_stream << "__device__ half min(half a, half b)\n"
+ << "{\n return __hlt(__half(a), __half(b)) ? a : b;\n}\n";
+ decl_stream << "__device__ half operator<="
+ << "(__half a, __half b)\n"
+ << "{\n return __hlt(a, b);\n}\n";
+ decl_stream << "__device__ half operator+"
+ << "(__half a, __half &b)\n"
+ <<"{\n return __hadd(a, b);\n}\n";
+ decl_stream << "__device__ half operator*"
+ << "(__half a, __half b)\n"
+ << "{\n return __hmul(a, b);\n}\n";
}
if (enable_int8_) {