hipify-clang#

hipify-clang is a clang-based tool for translating CUDA sources into HIP sources. It translates CUDA source into an abstract syntax tree, which is traversed by transformation matchers. After applying all the matchers, the output HIP source is produced.

Advantages:

  1. It is a translator; thus, any even very complicated constructs will be parsed successfully, or an error will be reported.

  2. It supports clang options like -I, -D, --cuda-path, etc.

  3. Seamless support of new CUDA versions as it is the clang’s responsibility.

  4. Ease of support.

Disadvantages:

  1. The main advantage is also the main disadvantage: the input CUDA code should be correct; incorrect code wouldn’t be translated to HIP.

  2. CUDA should be installed and provided in case of multiple installations by --cuda-path option.

  3. All the includes and defines should be provided to transform code successfully.

hipify-clang: dependencies#

hipify-clang requires:

  1. LLVM+CLANG of at least version 4.0.0; the latest stable and recommended release: 16.0.2.

  2. CUDA of at least version 8.0, the latest supported version is 12.1.1.

LLVM release version CUDA latest supported version Windows Linux
3.8.0*, 3.8.1*,
3.9.0*, 3.9.1*		 7.5 + +
4.0.0, 4.0.1,
5.0.0, 5.0.1, 5.0.2		 8.0 + +
6.0.0, 6.0.1 9.0 + +
7.0.0, 7.0.1, 7.1.0 9.2 works only with the patch
due to the clang's bug 38811
patch for 7.0.0**
patch for 7.0.1**
patch for 7.1.0**
 -
not working due to
the clang's bug 36384
8.0.0, 8.0.1 10.0 works only with the patch
due to the clang's bug 38811
patch for 8.0.0**
patch for 8.0.1**
 +
9.0.0, 9.0.1 10.1 + +
10.0.0, 10.0.1 11.0 + +
11.0.1, 11.1.0, 11.1.1 works only with the patch
due to the clang's bug 47332
patch for 10.0.0***
patch for 10.0.1***
11.0.0 11.0 + +
11.0.1, 11.1.0, 11.1.1 works only with the patch
due to the clang's bug 47332
patch for 11.0.0***
11.0.1, 11.1.0 11.2.2 + +
12.0.0, 12.0.1,
13.0.0, 13.0.1 		11.5.1 + +
14.0.0, 14.0.1,
14.0.2, 14.0.3, 14.0.4 		11.7.1 works only with the patch
due to the clang's bug 54609
patch for 14.0.0**
patch for 14.0.1**
patch for 14.0.2**
patch for 14.0.3**
patch for 14.0.4**
 +
14.0.5, 14.0.6,
15.0.0, 15.0.1, 15.0.2,
15.0.3, 15.0.4, 15.0.5,
15.0.6, 15.0.7		 11.8.0 + +
16.0.0, 16.0.1, 16.0.2 12.1.1 LATEST STABLE CONFIG
17.0.0git 12.1.1 + +

* LLVM 3.x is not supported anymore but might still work.

** Download the patch and unpack it into your LLVM distributive directory: a few header files will be overwritten; rebuilding of LLVM is not needed.

*** Download the patch and unpack it into your LLVM source directory: the file Cuda.cpp will be overwritten; needs further rebuilding of LLVM.

In most cases, you can get a suitable version of LLVM+CLANG with your package manager.

Failing that or having multiple versions of LLVM, you can download a release archive, build or install it, and set CMAKE_PREFIX_PATH so cmake can find it; for instance: -DCMAKE_PREFIX_PATH=d:\LLVM\16.0.2\dist

hipify-clang: usage#

To process a file, hipify-clang needs access to the same headers that would be required to compile it with clang.

For example:

./hipify-clang square.cu --cuda-path=/usr/local/cuda-12.1 -I /usr/local/cuda-12.1/samples/common/inc

Prerequisites#

The ROCm website describes how to set up the ROCm repositories and install the required platform dependencies.

./hipify-clang cpp17.cu --cuda-path=/usr/local/cuda-12.1 -- -std=c++17

The Clang manual for compiling CUDA may be useful.

For some hipification automation (starting from clang 8.0.0), it is also possible to provide a Compilation Database in JSON format in the compile_commands.json file:

-p <folder containing compile_commands.json> or
-p=<folder containing compile_commands.json>

The compilation database should be provided in the compile_commands.json file or generated by clang based on cmake; options separator '--' must not be used.

For a list of hipify-clang options, run hipify-clang --help.

hipify-clang: building#

cd .. \
mkdir build dist \
cd build

cmake \
 -DCMAKE_INSTALL_PREFIX=../dist \
 -DCMAKE_BUILD_TYPE=Release \
 ../hipify

make -j install

On Windows, the following option should be specified for cmake in the first place: -G "Visual Studio 17 2022"; the generated hipify-clang.sln should be built by Visual Studio 17 2022 instead of make. Please, see hipify-clang: Windows for the supported tools for building.

Debug build type -DCMAKE_BUILD_TYPE=Debug is also supported and tested; LLVM+CLANG should be built in Debug mode as well. 64-bit build mode (-Thost=x64 on Windows) is also supported; LLVM+CLANG should be built in 64-bit mode as well.

The binary can then be found at ./dist/hipify-clang or at the folder specified by the -DCMAKE_INSTALL_PREFIX option.

hipify-clang: testing#

hipify-clang has unit tests using LLVM lit/FileCheck.

LLVM+CLANG should be built from sources, pre-built binaries are not exhaustive for testing. Before building ensure that the software required for building is of an appropriate version.

LLVM <= 9.0.1:

  1. download LLVM+CLANG sources;

  2. build LLVM+CLANG:

     cd .. \
     mkdir build dist \
     cd build

Linux:

     cmake \
      -DCMAKE_INSTALL_PREFIX=../dist \
      -DLLVM_SOURCE_DIR=../llvm \
      -DLLVM_TARGETS_TO_BUILD="X86;NVPTX" \
      -DCMAKE_BUILD_TYPE=Release \
      ../llvm
     make -j install

Windows:

     cmake \
      -G "Visual Studio 16 2019" \
      -A x64 \
      -Thost=x64 \
      -DCMAKE_INSTALL_PREFIX=../dist \
      -DLLVM_SOURCE_DIR=../llvm \
      -DLLVM_TARGETS_TO_BUILD="NVPTX" \
      -DCMAKE_BUILD_TYPE=Release \
      ../llvm

Run Visual Studio 16 2019, open the generated LLVM.sln, build all, and build the INSTALL project.

LLVM >= 10.0.0:

  1. download LLVM project sources;

  2. build LLVM project:

      cd .. \
      mkdir build dist \
      cd build

Linux:

     cmake \
      -DCMAKE_INSTALL_PREFIX=../dist \
      -DLLVM_TARGETS_TO_BUILD="X86;NVPTX" \
      -DLLVM_ENABLE_PROJECTS="clang" \
      -DCMAKE_BUILD_TYPE=Release \
      ../llvm-project/llvm
     make -j install

Windows:

     cmake \
      -G "Visual Studio 17 2022" \
      -A x64 \
      -Thost=x64 \
      -DCMAKE_INSTALL_PREFIX=../dist \
      -DLLVM_TARGETS_TO_BUILD="NVPTX" \
      -DLLVM_ENABLE_PROJECTS="clang" \
      -DCMAKE_BUILD_TYPE=Release \
      ../llvm-project/llvm

Run Visual Studio 17 2022, open the generated LLVM.sln, build all, build project INSTALL.

  1. Ensure CUDA of minimum version 7.0 is installed.

    • Having multiple CUDA installations to choose a particular version the DCUDA_TOOLKIT_ROOT_DIR option should be specified:

      • Linux: -DCUDA_TOOLKIT_ROOT_DIR=/usr/include

      • Windows: -DCUDA_TOOLKIT_ROOT_DIR="c:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v12.1"

        -DCUDA_SDK_ROOT_DIR="c:/ProgramData/NVIDIA Corporation/CUDA Samples/v12.1"

  2. Ensure cuDNN of the version corresponding to CUDA’s version is installed.

    • Path to cuDNN should be specified by the CUDA_DNN_ROOT_DIR option:

      • Linux: -DCUDA_DNN_ROOT_DIR=/usr/include

      • Windows: -DCUDA_DNN_ROOT_DIR=d:/CUDNN/cudnn-12.1-windows-x64-v8.9.0

  3. Ensure CUB of the version corresponding to CUDA’s version is installed.

    • Path to CUB should be specified by the CUDA_CUB_ROOT_DIR option:

      • Linux: -DCUDA_CUB_ROOT_DIR=/srv/git/CUB

      • Windows: -DCUDA_CUB_ROOT_DIR=d:/GIT/cub

  4. Ensure python of minimum required version 2.7 is installed.

  5. Ensure lit and FileCheck are installed - these are distributed with LLVM.

    • Install lit into python:

      • Linux: python /usr/llvm/16.0.2/llvm-project/llvm/utils/lit/setup.py install

      • Windows: python d:/LLVM/16.0.2/llvm-project/llvm/utils/lit/setup.py install

    • Starting with LLVM 6.0.1 path to llvm-lit python script should be specified by the LLVM_EXTERNAL_LIT option:

      • Linux: -DLLVM_EXTERNAL_LIT=/usr/llvm/16.0.2/build/bin/llvm-lit

      • Windows: -DLLVM_EXTERNAL_LIT=d:/LLVM/16.0.2/build/Release/bin/llvm-lit.py

    • FileCheck:

      • Linux: copy from /usr/llvm/16.0.2/build/bin/ to CMAKE_INSTALL_PREFIX/dist/bin

      • Windows: copy from d:/LLVM/16.0.2/build/Release/bin to CMAKE_INSTALL_PREFIX/dist/bin

      • Or specify the path to FileCheck in CMAKE_INSTALL_PREFIX option

  6. To run OpenGL tests successfully on:

     - ***Linux***: install at least essential GL headers (on Ubuntu by `sudo apt-get install mesa-common-dev`)

 - ***Windows***: nothing to do: all the required headers are shipped with Windows SDK

  7. Set HIPIFY_CLANG_TESTS option turned on: -DHIPIFY_CLANG_TESTS=1.

  8. Build and run tests:

hipify-clang: Linux#

On Linux the following configurations are tested:

Ubuntu 14: LLVM 4.0.0 - 7.1.0, CUDA 7.0 - 9.0, cuDNN 5.0.5 - 7.6.5

Ubuntu 16-18: LLVM 8.0.0 - 14.0.6, CUDA 8.0 - 10.2, cuDNN 5.1.10 - 8.0.5

Ubuntu 20-21: LLVM 9.0.0 - 16.0.2, CUDA 8.0 - 12.1.1, cuDNN 5.1.10 - 8.9.0

Minimum build system requirements for the above configurations:

Python 2.7, cmake 3.16.8, GNU C/C++ 7.5.

Recommended build system requirements:

Python 3.9.7, cmake 3.22.0, GNU C/C++ 11.2.

Here is an example of building hipify-clang with testing support on Ubuntu 21.10.0:

cmake
 -DHIPIFY_CLANG_TESTS=1 \
 -DCMAKE_BUILD_TYPE=Release \
 -DCMAKE_INSTALL_PREFIX=../dist \
 -DCMAKE_PREFIX_PATH=/usr/llvm/16.0.2/dist \
 -DCUDA_TOOLKIT_ROOT_DIR=/usr/local/cuda \
 -DCUDA_DNN_ROOT_DIR=/usr/local/cuda \
 -DCUDA_CUB_ROOT_DIR=/usr/CUB \
 -DLLVM_EXTERNAL_LIT=/usr/llvm/16.0.2/build/bin/llvm-lit \
 ../hipify

A corresponding successful output:

-- The C compiler identification is GNU 11.2.0
-- The CXX compiler identification is GNU 11.2.0
-- Detecting C compiler ABI info
-- Detecting C compiler ABI info - done
-- Check for working C compiler: /usr/bin/cc - skipped
-- Detecting C compile features
-- Detecting C compile features - done
-- Detecting CXX compiler ABI info
-- Detecting CXX compiler ABI info - done
-- Check for working CXX compiler: /usr/bin/c++ - skipped
-- Detecting CXX compile features
-- Detecting CXX compile features - done
-- Found ZLIB: /usr/lib/x86_64-linux-gnu/libz.so (found version "1.2.11")
-- Found LLVM 16.0.2:
--    - CMake module path: /usr/llvm/16.0.2/dist/lib/cmake/llvm
--    - Include path     : /usr/llvm/16.0.2/dist/include
--    - Binary path      : /usr/llvm/16.0.2/dist/bin
-- Linker detection: GNU ld
-- Found PythonInterp: /usr/bin/python (found suitable version "3.9.7", minimum required is "2.7")
-- Found lit: /usr/local/bin/lit
-- Found FileCheck: /usr/llvm/16.0.2/dist/bin/FileCheck
-- Looking for pthread.h
-- Looking for pthread.h - found
-- Performing Test CMAKE_HAVE_LIBC_PTHREAD
-- Performing Test CMAKE_HAVE_LIBC_PTHREAD - Success
-- Found Threads: TRUE
-- Found CUDA: /usr/local/cuda (found version "12.1")
-- Configuring done
-- Generating done
-- Build files have been written to: /usr/hipify/build

make test-hipify

A corresponding successful output:

Running HIPify regression tests
========================================
CUDA 12.1 - will be used for testing
LLVM 16.0.2 - will be used for testing
x86_64 - Platform architecture
Linux 5.13.0-21-generic - Platform OS
64 - hipify-clang binary bitness
64 - python 3.9.7 binary bitness
========================================
-- Testing: 82 tests, 12 threads --
PASS: hipify :: unit_tests/casts/reinterpret_cast.cu (1 of 82)
PASS: hipify :: unit_tests/device/atomics.cu (2 of 82)
PASS: hipify :: unit_tests/compilation_database/cd_intro.cu (3 of 82)
PASS: hipify :: unit_tests/device/device_symbols.cu (4 of 82)
PASS: hipify :: unit_tests/device/math_functions.cu (5 of 82)
PASS: hipify :: unit_tests/headers/headers_test_01.cu (6 of 82)
PASS: hipify :: unit_tests/headers/headers_test_02.cu (7 of 82)
PASS: hipify :: unit_tests/headers/headers_test_03.cu (8 of 82)
PASS: hipify :: unit_tests/headers/headers_test_05.cu (9 of 82)
PASS: hipify :: unit_tests/headers/headers_test_06_12000.cu (10 of 82)
PASS: hipify :: unit_tests/headers/headers_test_04.cu (11 of 82)
PASS: hipify :: unit_tests/headers/headers_test_07_12000.cu (12 of 82)
PASS: hipify :: unit_tests/headers/headers_test_10.cu (13 of 82)
PASS: hipify :: unit_tests/headers/headers_test_11.cu (14 of 82)
PASS: hipify :: unit_tests/headers/headers_test_08_12000.cu (15 of 82)
PASS: hipify :: unit_tests/kernel_launch/kernel_launch_01.cu (16 of 82)
PASS: hipify :: unit_tests/headers/headers_test_09_12000.cu (17 of 82)
PASS: hipify :: unit_tests/libraries/CAFFE2/caffe2_02.cu (18 of 82)
PASS: hipify :: unit_tests/libraries/CAFFE2/caffe2_01.cu (19 of 82)
PASS: hipify :: unit_tests/libraries/cuBLAS/cublas_sgemm_matrix_multiplication.cu (20 of 82)
PASS: hipify :: unit_tests/libraries/cuBLAS/rocBLAS/cublas_1_based_indexing_rocblas.cu (21 of 82)
PASS: hipify :: unit_tests/libraries/cuBLAS/cublas_1_based_indexing.cu (22 of 82)
PASS: hipify :: unit_tests/libraries/cuBLAS/cublas_v1.cu (23 of 82)
PASS: hipify :: unit_tests/libraries/cuComplex/cuComplex_Julia.cu (24 of 82)
PASS: hipify :: unit_tests/libraries/cuDNN/cudnn_softmax.cu (25 of 82)
PASS: hipify :: unit_tests/libraries/cuFFT/simple_cufft.cu (26 of 82)
PASS: hipify :: unit_tests/libraries/cuBLAS/rocBLAS/cublas_sgemm_matrix_multiplication_rocblas.cu (27 of 82)
PASS: hipify :: unit_tests/libraries/cuRAND/poisson_api_example.cu (28 of 82)
PASS: hipify :: unit_tests/libraries/cuSPARSE/cuSPARSE_03.cu (29 of 82)
PASS: hipify :: unit_tests/libraries/cuRAND/benchmark_curand_generate.cpp (30 of 82)
PASS: hipify :: unit_tests/libraries/cuRAND/benchmark_curand_kernel.cpp (31 of 82)
PASS: hipify :: unit_tests/libraries/cuSPARSE/cuSPARSE_04.cu (32 of 82)
PASS: hipify :: unit_tests/libraries/cuSPARSE/cuSPARSE_05.cu (33 of 82)
PASS: hipify :: unit_tests/libraries/cuSPARSE/cuSPARSE_06.cu (34 of 82)
PASS: hipify :: unit_tests/libraries/cuSPARSE/cuSPARSE_07.cu (35 of 82)
PASS: hipify :: unit_tests/libraries/cuSPARSE/cuSPARSE_08.cu (36 of 82)
PASS: hipify :: unit_tests/libraries/cuSPARSE/cuSPARSE_09.cu (37 of 82)
PASS: hipify :: unit_tests/namespace/ns_kernel_launch.cu (38 of 82)
PASS: hipify :: unit_tests/libraries/cuSPARSE/cuSPARSE_10.cu (39 of 82)
PASS: hipify :: unit_tests/pp/pp_if_else_conditionals.cu (40 of 82)
PASS: hipify :: unit_tests/pp/pp_if_else_conditionals_01.cu (41 of 82)
PASS: hipify :: unit_tests/pp/pp_if_else_conditionals_01_LLVM_10.cu (42 of 82)
PASS: hipify :: unit_tests/pp/pp_if_else_conditionals_LLVM_10.cu (43 of 82)
PASS: hipify :: unit_tests/samples/2_Cookbook/0_MatrixTranspose/MatrixTranspose.cpp (44 of 82)
PASS: hipify :: unit_tests/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp (45 of 82)
PASS: hipify :: unit_tests/samples/2_Cookbook/13_occupancy/occupancy.cpp (46 of 82)
PASS: hipify :: unit_tests/samples/2_Cookbook/1_hipEvent/hipEvent.cpp (47 of 82)
PASS: hipify :: unit_tests/samples/2_Cookbook/2_Profiler/Profiler.cpp (48 of 82)
PASS: hipify :: unit_tests/samples/2_Cookbook/7_streams/stream.cpp (49 of 82)
PASS: hipify :: unit_tests/samples/MallocManaged.cpp (50 of 82)
PASS: hipify :: unit_tests/samples/2_Cookbook/8_peer2peer/peer2peer.cpp (51 of 82)
PASS: hipify :: unit_tests/samples/allocators.cu (52 of 82)
PASS: hipify :: unit_tests/samples/half2_allocators.cu (53 of 82)
PASS: hipify :: unit_tests/samples/coalescing.cu (54 of 82)
PASS: hipify :: unit_tests/samples/dynamic_shared_memory.cu (55 of 82)
PASS: hipify :: unit_tests/samples/axpy.cu (56 of 82)
PASS: hipify :: unit_tests/samples/cudaRegister.cu (57 of 82)
PASS: hipify :: unit_tests/samples/intro.cu (58 of 82)
PASS: hipify :: unit_tests/samples/square.cu (59 of 82)
PASS: hipify :: unit_tests/samples/static_shared_memory.cu (60 of 82)
PASS: hipify :: unit_tests/samples/vec_add.cu (61 of 82)
PASS: hipify :: unit_tests/kernel_launch/kernel_launch_syntax.cu (62 of 82)
PASS: hipify :: unit_tests/synthetic/driver_structs.cu (63 of 82)
PASS: hipify :: unit_tests/synthetic/driver_enums.cu (64 of 82)
PASS: hipify :: unit_tests/synthetic/driver_defines.cu (65 of 82)
PASS: hipify :: unit_tests/synthetic/driver_typedefs.cu (66 of 82)
PASS: hipify :: unit_tests/synthetic/driver_functions.cu (67 of 82)
PASS: hipify :: unit_tests/synthetic/driver_functions_internal.cu (68 of 82)
PASS: hipify :: unit_tests/synthetic/driver_unions.cu (69 of 82)
PASS: hipify :: unit_tests/synthetic/runtime_defines.cu (70 of 82)
PASS: hipify :: unit_tests/synthetic/runtime_enums.cu (71 of 82)
PASS: hipify :: unit_tests/synthetic/runtime_unions.cu (72 of 82)
PASS: hipify :: unit_tests/synthetic/runtime_structs.cu (73 of 82)
PASS: hipify :: unit_tests/synthetic/runtime_typedefs.cu (74 of 82)
PASS: hipify :: unit_tests/synthetic/runtime_functions.cu (75 of 82)
PASS: hipify :: unit_tests/synthetic/runtime_functions_11010.cu (76 of 82)
PASS: hipify :: unit_tests/synthetic/libraries/cudnn2miopen.cu (77 of 82)
PASS: hipify :: unit_tests/graph/simple_mechs.cu (78 of 82)
PASS: hipify :: unit_tests/options/kernel-execution-syntax/both-kernel-execution-syntax.cu (79 of 82)
PASS: hipify :: unit_tests/options/kernel-execution-syntax/cuda-kernel-execution-syntax.cu (80 of 82)
PASS: hipify :: unit_tests/options/kernel-execution-syntax/hip-kernel-execution-syntax.cu (81 of 82)
PASS: hipify :: unit_tests/options/kernel-execution-syntax/none-kernel-execution-syntax.cu (82 of 82)
Testing Time: 5.78s
  Expected Passes    : 82
[100%] Built target test-hipify

hipify-clang: Windows#

Tested configurations:

LLVM

CUDA

cuDNN

Visual Studio (latest)

cmake

Python

4.0.0 - 5.0.2

7.0 - 8.0

5.1.10 - 7.1.4

2015.14.0, 2017.15.5.2

3.5.1 - 3.18.0

3.6.4 - 3.8.5

6.0.0 - 6.0.1

7.0 - 9.0

7.0.5 - 7.6.5

2015.14.0, 2017.15.5.5

3.6.0 - 3.18.0

3.7.2 - 3.8.5

7.0.0 - 7.1.0

7.0 - 9.2

7.6.5

2017.15.9.11

3.13.3 - 3.18.0

3.7.3 - 3.8.5

8.0.0 - 8.0.1

7.0 - 10.0

7.6.5

2017.15.9.15

3.14.2 - 3.18.0

3.7.4 - 3.8.5

9.0.0 - 9.0.1

7.0 - 10.1

7.6.5

2017.15.9.20, 2019.16.4.5

3.16.4 - 3.18.0

3.8.0 - 3.8.5

10.0.0 - 11.0.0

7.0 - 11.1

7.6.5 - 8.0.5

2017.15.9.30, 2019.16.8.3

3.19.2

3.9.1

11.0.1 - 11.1.0

7.0 - 11.2.2

7.6.5 - 8.0.5

2017.15.9.31, 2019.16.8.4

3.19.3

3.9.2

12.0.0 - 13.0.1

7.0 - 11.5.1

7.6.5 - 8.3.2

2017.15.9.43, 2019.16.11.9

3.22.2

3.10.2

14.0.0 - 14.0.6

7.0 - 11.7.1

8.0.5 - 8.4.1

2017.15.9.49, 2019.16.11.17, 2022.17.2.6

3.24.0

3.10.6

15.0.0 - 15.0.7

7.0 - 11.8.0

8.0.5 - 8.8.1

2017.15.9.53, 2019.16.11.25, 2022.17.5.2

3.26.0

3.11.2

16.0.0 - 16.0.2

7.0 - 12.1.1

8.0.5 - 8.9.0

2017.15.9.54, 2019.16.11.26, 2022.17.5.4

3.26.3

3.11.3

17.0.0git

7.0 - 12.1.1

8.0.5 - 8.9.0

2017.15.9.54, 2019.16.11.26, 2022.17.5.4

3.26.3

3.11.3

Building with testing support by Visual Studio 17 2022 on Windows 10:

cmake
 -G "Visual Studio 17 2022" \
 -A x64 \
 -Thost=x64 \
 -DHIPIFY_CLANG_TESTS=1 \
 -DCMAKE_BUILD_TYPE=Release \
 -DCMAKE_INSTALL_PREFIX=../dist \
 -DCMAKE_PREFIX_PATH=d:/LLVM/16.0.2/dist \
 -DCUDA_TOOLKIT_ROOT_DIR="c:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v12.1" \
 -DCUDA_SDK_ROOT_DIR="c:/ProgramData/NVIDIA Corporation/CUDA Samples/v12.1" \
 -DCUDA_DNN_ROOT_DIR=d:/CUDNN/cudnn-12.1-windows-x64-v8.9.0 \
 -DCUDA_CUB_ROOT_DIR=d:/GIT/cub \
 -DLLVM_EXTERNAL_LIT=d:/LLVM/16.0.2/build/Release/bin/llvm-lit.py \
 ../hipify

A corresponding successful output:

-- Found LLVM 16.0.2:
--    - CMake module path: d:/LLVM/16.0.2/dist/lib/cmake/llvm
--    - Include path     : d:/LLVM/16.0.2/dist/include
--    - Binary path      : d:/LLVM/16.0.2/dist/bin
-- Found PythonInterp: c:/Program Files/Python311/python.exe (found suitable version "3.11.3", minimum required is "3.6")
-- Found lit: c:/Program Files/Python311/Scripts/lit.exe
-- Found FileCheck: d:/LLVM/16.0.2/dist/bin/FileCheck.exe
-- Found CUDA: c:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v12.1 (found version "12.1")
-- Configuring done
-- Generating done
-- Build files have been written to: d:/hipify/build

Run Visual Studio 17 2022, open the generated hipify-clang.sln, build project test-hipify.