Enable ocltst execution on Rock

build_tools/github_actions/fetch_test_configurations.py

@@ -71,6 +71,20 @@ def _get_script_path(script_name: str) -> str:
             "windows": 4,
         },
     },
+    # ocltest
+    "ocltst": {
+        "job_name": "ocltst",
+        "fetch_artifact_args": "--ocltst --tests",
+        "timeout_minutes": 30,
+        "test_script": f"python {_get_script_path('test_ocltst.py')}",
+        "platform": ["linux", "windows"],
+        "total_shards_dict": {
+            "linux": 1,
+            "windows": 1,
+        },
+        "container_image": "ghcr.io/rocm/no_rocm_image_ubuntu24_04_ocl_rt@sha256:f059191fdd5cf3bf8c5a47cb017a7be6e9bded0d52be594feb96f724f6a838eb",
+        "container_options": "--cap-add=SYS_PTRACE --cap-add=SYS_NICE",
+    },
     # BLAS tests
     "rocblas": {
         "job_name": "rocblas",

build_tools/github_actions/test_executable_scripts/test_ocltst.py

@@ -0,0 +1,90 @@
+import logging
+import os
+import shlex
+import subprocess
+from pathlib import Path
+import sys
+import platform
+import shutil
+
+logging.basicConfig(level=logging.INFO)
+THEROCK_BIN_DIR_STR = os.getenv("THEROCK_BIN_DIR")
+if THEROCK_BIN_DIR_STR is None:
+    logging.info(
+        "++ Error: env(THEROCK_BIN_DIR) is not set. Please set it before executing tests."
+    )
+    sys.exit(1)
+THEROCK_BIN_DIR = Path(THEROCK_BIN_DIR_STR).resolve()
+THEROCK_DIR = Path(THEROCK_BIN_DIR).parent
+env = os.environ.copy()
+is_windows = platform.system() == "Windows"
+
+
+# copies the dlls to local ocltst path.
+# to overwrite the registry entries
+def copy_dlls_exe_path(ocltst_path):
+    if platform.system() == "Windows":
+        # hip and comgr dlls need to be copied to the same folder as exectuable
+        dlls_pattern = ["amdocl*.dll", "amd_comgr*.dll", "OpenCL.dll"]
+        dlls_to_copy = []
+        for pattern in dlls_pattern:
+            dlls_to_copy.extend(THEROCK_BIN_DIR.glob(pattern))
+        for dll in dlls_to_copy:
+            try:
+                shutil.copy(dll, ocltst_path)
+                logging.info(f"++ Copied: {dll} to {ocltst_path}")
+            except Exception as e:
+                logging.info(f"++ Error copying {dll}: {e}")
+
+
+# returns ocltst path
+def setup_env(env):
+    ROCM_PATH = Path(THEROCK_DIR)
+    env["ROCM_PATH"] = str(ROCM_PATH)
+    if not is_windows:
+        ROCK_LIB_PATH = Path(THEROCK_DIR) / "lib"
+        OCL_LIB = Path(ROCK_LIB_PATH) / "opencl"
+        LLVM_LIB = Path(ROCK_LIB_PATH) / "llvm" / "lib"
+        ROCM_SYSDEPS_LIB = Path(ROCK_LIB_PATH) / "rocm_sysdeps" / "lib"
+        OCL_ICD_VENDORS = Path(THEROCK_DIR) / "etc" / "OpenCL" / "vendors"
+        OCLTST_PATH = Path(THEROCK_DIR) / "share" / "opencl" / "ocltst"
+        LD_LIBRARY_PATH = os.getenv("LD_LIBRARY_PATH")
+        if LD_LIBRARY_PATH is not None:
+            LD_LIBRARY_PATH = Path(LD_LIBRARY_PATH)
+        env["LD_LIBRARY_PATH"] = (
+            f"{ROCK_LIB_PATH}:{OCL_LIB}:{LLVM_LIB}:{ROCM_SYSDEPS_LIB}:{LD_LIBRARY_PATH}:{OCLTST_PATH}"
+        )
+        env["OCL_ICD_VENDORS"] = f"{OCL_ICD_VENDORS}/"
+        logging.info(f"++ Setting LD_LIBRARY_PATH={env['LD_LIBRARY_PATH']}")
+        logging.info(f"++ Setting OCL_ICD_VENDORS={env['OCL_ICD_VENDORS']}")
+    else:
+
+        OCLTST_PATH = Path(THEROCK_DIR) / "tests" / "ocltst"
+        copy_dlls_exe_path(OCLTST_PATH)
+        OCL_DLL_FILE = Path(OCLTST_PATH) / "amdocl64.dll"
+        OCL_ICD_DLL = Path(THEROCK_BIN_DIR) / "OpenCL.dll"
+        env["OCL_ICD_FILENAMES"] = str(OCL_DLL_FILE)
+        logging.info(f"++ Setting OCL_ICD_FILENAMES={env['OCL_ICD_FILENAMES']}")
+    return OCLTST_PATH
+
+
+def execute_tests(env):
+    OCLTST_PATH = setup_env(env)
+    OCLTST = Path(OCLTST_PATH) / "ocltst"
+    module = "liboclruntime.so" if not is_windows else "oclruntime.dll"
+    # command to execute ocltst tests
+    cmd = [
+        f"{OCLTST}",
+        "-m",  # module to test
+        f"{module}",
+        "-s",  # threads to spawn/use
+        f"16",
+        "-A",  # exclude tests in the file
+        "oclruntime.exclude",  # perf related tests skipped
+    ]
+    logging.info(f"++ Exec [{OCLTST_PATH}]$ {shlex.join(cmd)}")
+    subprocess.run(cmd, cwd=OCLTST_PATH, check=True, env=env, shell=False)
+
+
+if __name__ == "__main__":
+    execute_tests(env)

build_tools/install_rocm_from_artifacts.py

@@ -38,6 +38,7 @@
     [--rocprofiler-sdk | --no-rocprofiler-sdk ]
     [--rocprofiler-systems | --no-rocprofiler-systems]
     [--rocrtst | --no-rocrtst]
+    [--ocltst | --no-ocltst]
     [--rocwmma | --no-rocwmma]
     [--libhipcxx | --no-libhipcxx]
     [--tests | --no-tests]
@@ -360,6 +361,7 @@ def retrieve_artifacts_by_run_id(args):
             args.rocprofiler_sdk,
             args.rocprofiler_systems,
             args.rocrtst,
+            args.ocltst,
             args.rocwmma,
             args.libhipcxx,
         ]
@@ -443,6 +445,10 @@ def retrieve_artifacts_by_run_id(args):
             # rocrtst depends on sysdeps-hwloc (which depends on sysdeps-libpciaccess)
             extra_artifacts.append("sysdeps-hwloc")
             extra_artifacts.append("sysdeps-libpciaccess")
+        if args.ocltst:
+            # ocltst binaries ship in the core-ocl test artifact (base already includes _lib/_dev).
+            if args.tests:
+                argv.append("core-ocl_test")
         if args.rocwmma:
             extra_artifacts.append("rocwmma")
             argv.append("rocwmma_dev")
@@ -781,6 +787,13 @@ def main(argv):
         action=argparse.BooleanOptionalAction,
     )
 
+    artifacts_group.add_argument(
+        "--ocltst",
+        default=False,
+        help="Include ocltst from core-ocl",
+        action=argparse.BooleanOptionalAction,
+    )
+
     artifacts_group.add_argument(
         "--rocwmma",
         default=False,

core/CMakeLists.txt

@@ -374,6 +374,7 @@ if(THEROCK_ENABLE_OCL_RUNTIME)
 
   set(OCL_CLR_CMAKE_ARGS)
   set(OCL_CLR_RUNTIME_DEPS)
+
   if(WIN32)
     # Windows CLR options
     set(_compute_pal_dir "${THEROCK_ROCM_SYSTEMS_SOURCE_DIR}/shared/amdgpu-windows-interop")
@@ -394,10 +395,11 @@ if(THEROCK_ENABLE_OCL_RUNTIME)
       rocprofiler-register
       ROCR-Runtime
     )
-    list(APPEND OCL_CLR_CMAKE_ARGS
-      "-DBUILD_TESTS=${THEROCK_BUILD_TESTING}"
-    )
   endif()
+  # ocltst build flag
+  list(APPEND OCL_CLR_CMAKE_ARGS
+  "-DBUILD_TESTS=${THEROCK_BUILD_TESTING}"
+  )
 
   therock_cmake_subproject_declare(ocl-clr
     USE_DIST_AMDGPU_TARGETS