From 5127770544f60e372bf79e4af549a07fa13e0eb7 Mon Sep 17 00:00:00 2001
From: foreman <dl.constructicon@amd.com>
Date: Thu, 27 Apr 2017 00:42:38 -0400
Subject: [PATCH] P4 to Git Change 1403108 by lmoriche@lmoriche_palamida on
 2017/04/27 00:33:06

	SWDEV-102733 - [OCL-LC-ROCm] Cmake build Write CMakeLists.txt to enable building with and without the DK environment
	- Remove the dependency on libHSAIL

Affected files ...

... //depot/stg/opencl/drivers/opencl/CMakeLists.txt#6 edit
... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/CMakeLists.txt#2 edit
... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/rocprogram.hpp#24 edit
... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/rocvirtual.cpp#37 edit
---
 rocclr/CMakeLists.txt                     | 4 ----
 rocclr/runtime/device/rocm/CMakeLists.txt | 3 ---
 rocclr/runtime/device/rocm/rocprogram.hpp | 2 --
 rocclr/runtime/device/rocm/rocvirtual.cpp | 6 ++++--
 4 files changed, 4 insertions(+), 11 deletions(-)

diff --git a/rocclr/CMakeLists.txt b/rocclr/CMakeLists.txt
index 1ba68e03d3..9396d5fd4b 100644
--- a/rocclr/CMakeLists.txt
+++ b/rocclr/CMakeLists.txt
@@ -30,10 +30,6 @@ include_directories(${CMAKE_BINARY_DIR}/compiler/llvm/tools/clang/include)
 include_directories(${CMAKE_SOURCE_DIR}/compiler/llvm/lib/Target/AMDGPU)
 include_directories(${CMAKE_BINARY_DIR}/compiler/llvm/lib/Target/AMDGPU)
 
-set(AMD_EXTENSIONS CACHE BOOL OFF)
-set(BUILD_HSAILASM CACHE BOOL OFF)
-add_subdirectory(compiler/sc/HSAIL/hsail-tools)
-
 add_subdirectory(compiler/driver)
 
 set(CMAKE_C_COMPILER "${PROJECT_BINARY_DIR}/compiler/llvm/bin/clang")
diff --git a/rocclr/runtime/device/rocm/CMakeLists.txt b/rocclr/runtime/device/rocm/CMakeLists.txt
index 9579b3ea97..0075b636d9 100644
--- a/rocclr/runtime/device/rocm/CMakeLists.txt
+++ b/rocclr/runtime/device/rocm/CMakeLists.txt
@@ -26,9 +26,6 @@ set(headers
 
 include_directories(${CMAKE_SOURCE_DIR}/compiler/lib/loaders/elf/utils/common)
 include_directories(${CMAKE_SOURCE_DIR}/compiler/lib/loaders/elf/utils/libelf)
-include_directories(${CMAKE_SOURCE_DIR}/compiler/sc/HSAIL/include)
-include_directories(${CMAKE_SOURCE_DIR}/compiler/sc/HSAIL/hsail-tools/libHSAIL)
-include_directories(${CMAKE_BINARY_DIR}/compiler/sc/HSAIL/hsail-tools/libHSAIL/generated)
 include_directories(${CMAKE_CURRENT_BINARY_DIR})
 
 
diff --git a/rocclr/runtime/device/rocm/rocprogram.hpp b/rocclr/runtime/device/rocm/rocprogram.hpp
index b9909bd0d3..6bde68594f 100644
--- a/rocclr/runtime/device/rocm/rocprogram.hpp
+++ b/rocclr/runtime/device/rocm/rocprogram.hpp
@@ -15,7 +15,6 @@
 #include <fstream>
 #include <iostream>
 #include "rocdevice.hpp"
-#include "HSAILItems.h"
 
 #if defined(WITH_LIGHTNING_COMPILER)
 #include "AMDGPUCodeObjectMetadata.h"
@@ -26,7 +25,6 @@ typedef llvm::AMDGPU::CodeObject::Kernel::Metadata KernelMD;
 typedef llvm::AMDGPU::CodeObject::Kernel::Arg::Metadata KernelArgMD;
 #endif  // defined(WITH_LIGHTNING_COMPILER)
 
-using namespace HSAIL_ASM;
 //! \namespace roc HSA Device Implementation
 namespace roc {
 
diff --git a/rocclr/runtime/device/rocm/rocvirtual.cpp b/rocclr/runtime/device/rocm/rocvirtual.cpp
index 70cc58f6fa..ee802c0581 100644
--- a/rocclr/runtime/device/rocm/rocvirtual.cpp
+++ b/rocclr/runtime/device/rocm/rocvirtual.cpp
@@ -19,6 +19,7 @@
 #include <fstream>
 #include <vector>
 #include <string>
+#include <limits>
 
 /**
 * HSA image object size in bytes (see HSAIL spec)
@@ -487,8 +488,9 @@ bool VirtualGPU::create(bool profilingEna) {
   // Pick a reasonable queue size
   uint32_t queue_size = 1024;
   queue_size = (queue_max_packets < queue_size) ? queue_max_packets : queue_size;
-  while (hsa_queue_create(gpu_device_, queue_size, HSA_QUEUE_TYPE_MULTI, nullptr, nullptr, UINT_MAX,
-                          UINT_MAX, &gpu_queue_) != HSA_STATUS_SUCCESS) {
+  while (hsa_queue_create(gpu_device_, queue_size, HSA_QUEUE_TYPE_MULTI, nullptr, nullptr,
+                          std::numeric_limits<uint>::max(), std::numeric_limits<uint>::max(),
+                          &gpu_queue_) != HSA_STATUS_SUCCESS) {
     queue_size >>= 1;
     if (queue_size < 64) {
       return false;