Windows variant of Linux installer without MSys2

.gitignore

@@ -1,6 +1,6 @@
 levelZeroLib/.idea/
 levelZeroLib/build/
 levelzero.iml
-opencl-copy.bc
+*.bc
+*.log
 target/
-

CHANGELOG.md

@@ -1,3 +1,17 @@
+## Beehive LevelZero-JNI 0.1.3
+14/03/2024
+
+- Add Windwos installer support
+
+## Beehive LevelZero-JNI 0.1.2
+12/12/23
+
+- Multiple SPIR-V Devices Fixed: [commit](https://github.com/beehive-lab/levelzero-jni/commit/fe20b18c9623b4d0533ee50d878b266ecdce46dc)
+- Support for data send/receive using Panama off-heap buffers 
+- Use `-O2` optimization: [commit](https://github.com/beehive-lab/levelzero-jni/commit/721b8aed7ac4e419843b3029be99c11267eeb32c)
+- Fix `string` release in JNI code: [commit](https://github.com/beehive-lab/levelzero-jni/commit/3c6d463ebafbf9d2de7be128f79483ff28c5ace6)
+- Fix Device properties: [commit](https://github.com/beehive-lab/levelzero-jni/commit/83c2e032197e2f8a13d895d2b75f72693424bd7b)
+
 ## LevelZero-JNI 0.1.1
 10/03/2022
 
@@ -11,8 +25,9 @@
 ## LevelZero-JNI 0.1.0
 03/12/2021
 
-- Initial prototype of LevelZero JNI.
+- Initial prototype of LevelZero JNI
 - It covers a subset of the Intel Level-Zero 1.1.2 Spec (Feb 2021)
 - C++ Wrapper for JNI provided
 - Java Interface provided
-- Set of examples and documentation
+- Set of examples and documentation
+

README.md

@@ -1,67 +1,108 @@
-# LevelZero JNI 
+# Beehive LevelZero JNI
 
-Baremetal GPU and FPGA programming for Java using the [LevelZero API](https://spec.oneapi.io/level-zero/latest/index.html). 
+Baremetal GPU and FPGA programming for Java using Intel's [Level Zero API](https://spec.oneapi.io/level-zero/latest/index.html). This project is a Java Native Interface (JNI) binding for Intel's Level Zero. This library is designed to be as closed as possible to the Level Zero API for C++. Subset of Level Zero 1.4.0 supported (Level Zero May 2022 version)
 
 
-This project is a Java Native Interface (JNI) binding for Intel's Level Zero. This library is as designed to be as closed as possible to the LevelZero API for C++. 
+## Compilation & configuration of the JNI Level Zero API
 
+### 1) Compile Level Zero API
 
-Subset of LevelZero 1.4.0 supported (Level Zero May 2022 version)
+#### Linux
 
+Note: Using tag `v1.4.1` from `level-zero` which implements Level Zero Specification 1.2.
 
-## Compilation & Configuration of the JNI Level-Zero API 
+```bash
+git clone https://github.com/oneapi-src/level-zero.git
+cd level-zero
+git checkout tags/v1.4.1
+mkdir build
+cd build
+cmake ..
+cmake --build . --config Release
+cmake --build . --config Release --target package
+```
 
-### 1) Install Level-Zero
+#### Windows
 
+Configuration:
+- Lenovo IdeaPad Gaming 3 15IHU6
+- Windows 11
+- VS Community 2022
+  + components C++, Git, Spectre mitigated libraries
+- CMake 3.26.3, Maven 3.9.1, JDK 21
 
-Note: Using commit `551dd5810a3cea7a7e26ac4441da31878e804b53` from `level-zero` 
+Run commands in _x64 Native Tools Command Prompt for VS 2022_.
 
+```cmd
+git clone https://github.com/oneapi-src/level-zero
+cd level-zero
+md build
+cd build
+cmake ..
+cmake --build . --config Release
 
-```bash
-$ git clone https://github.com/oneapi-src/level-zero
-$ mkdir build
-$ cd build
-$ cmake ..
-$ cmake --build . --config Release
-$ cmake --build . --config Release --target package
+rem check
+.\bin\Release\zello_world.exe
 ```
 
+Note: If `zello_world.exe` fails, search for existing Level Zero API DLLs (file names start with `ze_`, e.g. `ze_tracing_layer.dll`) in `c:\windows\system32` and move them to another folder.
 
-### 2) Compile JNI Native Code 
+### 2) Compile Level Zero JNI native code
 
-Set the paths to the directory of Level-Zero installation. Here's an example:
+Set the paths to the directory of Level Zero installation. Here are examples:
+
+#### Linux
 
 ```bash
-$ scl enable devtoolset-9 bash # << Only for CentOS
-$ export CPLUS_INCLUDE_PATH=<path-to-levelzero>/include:$CPLUS_INCLUDE_PATH
-$ export LD_LIBRARY_PATH=<path-to-levelzero>/build/lib:$LD_LIBRARY_PATH 
-$ export ZE_SHARED_LOADER="<path-to-levelzero>/build/lib/libze_loader.so"
-$ cd levelZeroLib
-$ mkdir build
-$ cd build
-$ cmake .. 
-$ make 
+scl enable devtoolset-9 bash # << Only for CentOS
+git clone https://github.com/beehive-lab/levelzero-jni
+export ZE_SHARED_LOADER="<path-to-levelzero>/build/lib/libze_loader.so"
+export CPLUS_INCLUDE_PATH=<path-to-levelzero>/include:$CPLUS_INCLUDE_PATH
+export C_INCLUDE_PATH=<path-to-levelzero>/include:$CPLUS_INCLUDE_PATH
+cd levelzero-jni/levelZeroLib
+mkdir build
+cd build
+cmake ..
+make
 ```
 
-##### 2.1 Obtain the LLVM-SPIRV Compiler
+#### Windows
+
+Note: Run commands in _x64 Native Tools Command Prompt for VS 2022_.
 
-In case you want to compile kernels from OpenCL C to SPIR-V and use the Level Zero library, you need to download the `llvm-spirv` compiler.
-The implementation we are currently using is the `intel/llvm`: [https://github.com/intel/llvm](https://github.com/intel/llvm).
+```cmd
+git clone https://github.com/otabuzzman/levelzero-jni
+set ZE_SHARED_LOADER=%USERPROFILE%\lab\level-zero\build\lib\release\ze_loader.lib
+set CPLUS_INCLUDE_PATH=%USERPROFILE%\lab\level-zero\include
+set C_INCLUDE_PATH=%USERPROFILE%\lab\level-zero\include
 
+rem add the folder with Intel's Level Zero API DLLs to PATH
+set PATH=%USERPROFILE%\lab\level-zero\build\bin\release;%PATH%
 
-### 3) Compile & Run a Java test
+cd levelzero-jni\levelZeroLib
+md build
+cd build
+cmake ..
+cmake --build . --config Release
+```
+
+#### Obtain a SPIR-V compiler (Linux and Windwos)
+
+In case you want to compile kernels from OpenCL C to SPIR-V and use the Level Zero API, you need to download the `llvm-spirv` compiler. The implementation we are currently using is [Intel LLVM](https://github.com/intel/llvm).
 
+### 3) Compile and run a Java test
+
+#### Linux
 
 ```bash
-$ mvn clean package
-$ ./scripts/run.sh   ## < This script compiles an OpenCL C program to SPIR-V using the llvm-spirv compiler (see 2.1)
+mvn clean package
+./scripts/run.sh   ## < This script compiles an OpenCL C program to SPIR-V using the llvm-spirv compiler (see 2.1)
 ```
 
 The OpenCL C kernel provided for this example is as follows:
 
-
 ```c
-__kernel void copydata(__global int* input, __global int* output) {
+__kernel void copyData(__global int* input, __global int* output) {
 	uint idx = get_global_id(0);
 	output[idx] = input[idx];
 }
@@ -70,17 +111,38 @@ __kernel void copydata(__global int* input, __global int* output) {
 To compile from OpenCL C to SPIR-V:
 
 ```bash
-$ clang -cc1 -triple spir copy_data.cl -O0 -finclude-default-header -emit-llvm-bc -o opencl-copy.bc
-$ llvm-spirv opencl-copy.bc -o opencl-copy.spv
-$ mv opencl-copy.spv /tmp/opencl-copy.spv
+clang -cc1 -triple spir copyData.cl -O0 -finclude-default-header -emit-llvm-bc -o copyData.bc
+llvm-spirv copyData.bc -o copyData.spv
 ```
 
+#### Windows
+
+Note: Java programs that use levelzero-jni are based on DLLs, which are provided by Intel's Level Zero API. For these programs to find these DLLs, the PATH environment variable must contain the folder that contains the DLLs.
+
+```cmd
+mvn clean package
+
+rem add the folder with Intel's Level Zero API DLLs to PATH
+set PATH=%USERPROFILE%\lab\level-zero\build\bin\release;%PATH%
+
+rem copyData.spv file expected in CWD
+.\scripts\run.cmd
+
+# more tests
+.\scripts\copies.cmd
+.\scripts\events.cmd
+.\scripts\fences.cmd
+.\scripts\kernelTimers.cmd
+.\scripts\transferTimers.cmd
+.\scripts\largeBuffers.cmd
+```
+
+
 ## License
 
 This project is developed at [The University of Manchester](https://www.manchester.ac.uk/), and it is fully open source under the [MIT](https://github.com/beehive-lab/levelzero-jni/blob/master/LICENSE) license.
 
 
 ## Acknowledgments
 
-The work was partially funded by the EU Horizon 2020 [Elegant 957286](https://www.elegant-h2020.eu/) project, and Intel Coorporation (https://www.intel.it/content/www/it/it/homepage.html).
-
+The work was partially funded by the EU Horizon 2020 [Elegant 957286](https://www.elegant-h2020.eu/) project, and [Intel Coorporation](https://www.intel.it/content/www/it/it/homepage.html).

copyData.cl

@@ -0,0 +1,4 @@
+__kernel void copyData(__global int* input, __global int* output) {
+	uint idx = get_global_id(0);
+	output[idx] = input[idx];
+}

copyLong.cl

@@ -0,0 +1,4 @@
+__kernel void copyData(__global long* input, __global long* output) {
+	uint idx = get_global_id(0);
+	output[idx] = input[idx];
+}

copyTest.cl

@@ -0,0 +1,17 @@
+__kernel void copyTest(__global uchar *_heap_base)
+{
+	int i_8, i_7, i_1, i_2;
+	ulong ul_0, ul_6;
+	long l_3, l_5, l_4;
+
+	__global ulong *_frame = (__global ulong *) &_heap_base[0];
+
+	ul_0  =  (ulong) _frame[3];
+	i_1  =  get_global_id(0);
+	i_2  =  i_1;
+	l_3  =  (long) i_2;
+	l_4  =  l_3 << 3;  // Long buffer
+	l_5  =  l_4 + 16L; // Randomly starting in position 16
+	ul_6  =  ul_0 + l_5;
+	*((__global int *) ul_6)  =  555;
+}

levelZeroLib/CMakeLists.txt

@@ -7,24 +7,68 @@ find_package(JNI REQUIRED)
 #	message("- ZE_SHARED_LOADER found")
 #else()
 #	message("ERROR: ZE_SHARED_LOADER is not set")
+#if(MSVC)
+#	message("Set this export to PATH-TO\level-zero\build\lib\release\ze_loader.lib")
+#else()
 #	message("Set this export to PATH-TO/level-zero/build/lib/libze_loader.so")
+#endif()
 #	return()
 #endif()
 
 # set the C++14 standard
 set(CMAKE_CXX_STANDARD 14)
-set(GCC_INTEL_LEVEL0 "-std=c++14 -O2 -fpermissive -rdynamic -fPIC -lstdc++  -fno-stack-protector")
 
+if(MSVC)
+	string(REPLACE "/MDd" "/MTd" CMAKE_CXX_FLAGS_DEBUG "${CMAKE_CXX_FLAGS_DEBUG}")
+	string(REPLACE "/MD" "/MT" CMAKE_CXX_FLAGS_RELEASE "${CMAKE_CXX_FLAGS_RELEASE}")
 
-# This should point to: "PATH-TO/level-zero/build/lib/libze_loader.so"
-set(ZE_LOADER $ENV{ZE_SHARED_LOADER})
+	# enable multi-process compilation, not supported by clang-cl
+	if(NOT CMAKE_CXX_COMPILER_ID STREQUAL Clang)
+		set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} /MP")
+	endif()
+
+	# enable exceptions handling
+	set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} /EHsc")
+
+	# enable creation of PDB files for Release Builds
+	set(CMAKE_CXX_FLAGS_RELEASE "${CMAKE_CXX_FLAGS_RELEASE} /Zi")
+	set(CMAKE_SHARED_LINKER_FLAGS_RELEASE "${CMAKE_SHARED_LINKER_FLAGS_RELEASE} /DEBUG /OPT:REF /OPT:ICF")
 
-add_definitions(${GCC_INTEL_LEVEL0})
+	# enable CET shadow stack
+	set(CMAKE_SHARED_LINKER_FLAGS "${CMAKE_SHARED_LINKER_FLAGS} /CETCOMPAT")
+
+	#Use of sccache with MSVC requires workaround of replacing /Zi with /Z7
+	#https://github.com/mozilla/sccache
+	if(USE_Z7) #sccache
+		string(REPLACE "/Zi" "/Z7" CMAKE_CXX_FLAGS_DEBUG "${CMAKE_CXX_FLAGS_DEBUG}")
+		string(REPLACE "/Zi" "/Z7" CMAKE_C_FLAGS_DEBUG "${CMAKE_C_FLAGS_DEBUG}")
+		string(REPLACE "/Zi" "/Z7" CMAKE_CXX_FLAGS_RELEASE "${CMAKE_CXX_FLAGS_RELEASE}")
+		string(REPLACE "/Zi" "/Z7" CMAKE_C_FLAGS_RELEASE "${CMAKE_C_FLAGS_RELEASE}")
+		string(REPLACE "/Zi" "/Z7" CMAKE_CXX_FLAGS_RELWITHDEBINFO "${CMAKE_CXX_FLAGS_RELWITHDEBINFO}")
+		string(REPLACE "/Zi" "/Z7" CMAKE_C_FLAGS_RELWITHDEBINFO "${CMAKE_C_FLAGS_RELWITHDEBINFO}")
+	endif()
+else()
+	set(GCC_INTEL_LEVEL0 "-std=c++14")
+	set(GCC_INTEL_LEVEL0 "${GCC_INTEL_LEVEL0} -O2")
+	set(GCC_INTEL_LEVEL0 "${GCC_INTEL_LEVEL0} -fpermissive")
+	set(GCC_INTEL_LEVEL0 "${GCC_INTEL_LEVEL0} -fno-stack-protector")
+	set(GCC_INTEL_LEVEL0 "${GCC_INTEL_LEVEL0} -fPIC")
+	set(GCC_INTEL_LEVEL0 "${GCC_INTEL_LEVEL0} -rdynamic")
+	set(GCC_INTEL_LEVEL0 "${GCC_INTEL_LEVEL0} -lstdc++")
+	add_definitions(${GCC_INTEL_LEVEL0})
+endif()
+
+# This should point to:
+#  Linos: "PATH-TO/level-zero/build/lib/libze_loader.so"
+#  Winos: "PATH-TO\level-zero\build\lib\release\ze_loader.lib"
+set(ZE_LOADER $ENV{ZE_SHARED_LOADER})
 
 include_directories(
 	./
 	${JNI_INCLUDE_DIRS}
 	${OPENCL_INCLUDE_DIRS}
+	$ENV{CPLUS_INCLUDE_PATH}
+	$ENV{C_INCLUDE_PATH}
 	)
 
 file(GLOB_RECURSE "src/*.cpp")
@@ -36,10 +80,11 @@ add_library(tornado-levelzero SHARED
 	src/levelZeroModule.cpp
 	src/levelZeroKernel.cpp
 	src/levelZeroCommandList.cpp
-    src/levelZeroCommandQueue.cpp
+	src/levelZeroCommandQueue.cpp
 	src/levelZeroFence.cpp
 	src/levelZeroTimeStampKernel.cpp
-		src/levelZeroDescriptors.cpp)
+	src/levelZeroDescriptors.cpp
+	)
 
 target_link_libraries(tornado-levelzero ${OpenCL_LIBRARIES} ${JNI_LIB_DIRS})
 target_link_libraries(tornado-levelzero ${ZE_LOADER})

levelZeroLib/src/levelZeroBuffer.cpp

@@ -140,7 +140,11 @@ JNIEXPORT void JNICALL Java_uk_ac_manchester_tornado_drivers_spirv_levelzero_Lev
     jbyte *offHeapByteArray = static_cast<jbyte *>(buffer);
     jbyte *arrayByte = env->GetByteArrayElements(javaArray, 0);
     int size = env->GetArrayLength(javaArray);
+#ifdef _WIN32
+    _memccpy(offHeapByteArray, arrayByte, 0, size);
+#else
     memccpy(offHeapByteArray, arrayByte, 0, size);
+#endif
 }
 
 /*
@@ -169,10 +173,17 @@ JNIEXPORT void JNICALL Java_uk_ac_manchester_tornado_drivers_spirv_levelzero_Lev
     jfieldID fieldPointer = env->GetFieldID(klass, "ptrBuffer", "J");
     jlong ptr = env->GetLongField(javaBufferObject, fieldPointer);
 
+#ifdef _WIN32
+    int64_t *buffer = nullptr;
+    if (ptr != -1) {
+        buffer = reinterpret_cast<int64_t *>(ptr);
+    }
+#else
     long *buffer = nullptr;
     if (ptr != -1) {
         buffer = reinterpret_cast<long *>(ptr);
     }
+#endif
     for (int i = 0; i < size; i++) {
         buffer[i] = value;
     }
@@ -217,7 +228,11 @@ JNIEXPORT void JNICALL Java_uk_ac_manchester_tornado_drivers_spirv_levelzero_Lev
     jlong *offHeapByteArray = static_cast<jlong *>(buffer);
     jlong *arrayByte = env->GetLongArrayElements(array, 0);
     int size = env->GetArrayLength(array);
+#ifdef _WIN32
+    _memccpy(offHeapByteArray, arrayByte, 0, size);
+#else
     memccpy(offHeapByteArray, arrayByte, 0, size);
+#endif
 }
 
 /*