Add tests and enable testing in pre-commit and scheduled CI

.github/actions/build-opencl-clang/action.yml

@@ -17,6 +17,10 @@ inputs:
     description: 'Build type to pass to CMake'
     required: false
     default: Release
+  run_tests:
+    description: 'Run ninja check-opencl-clang after build'
+    required: false
+    default: 'false'
 
 runs:
   using: 'composite'
@@ -59,10 +63,17 @@ runs:
           -DLLVM_EXTERNAL_OPENCL_CLANG_SOURCE_DIR=${{ github.workspace }}/llvm-project/opencl-clang        \
           -DCMAKE_BUILD_TYPE=${{ inputs.build_type }}  \
           -DLLVM_TARGETS_TO_BUILD="X86" \
-          -G "Unix Makefiles"
+          -G "Ninja"
 
     - name: Build
       shell: bash
       run: |
         cd build
-        make opencl-clang -j8
+        ninja opencl-clang
+
+    - name: Test
+      if: ${{ inputs.run_tests == 'true' }}
+      shell: bash
+      run: |
+        cd build
+        ninja check-opencl-clang

.github/workflows/on-push-verification-in-tree.yml → .github/workflows/on-push-verification-in-tree-test.yml

@@ -1,10 +1,10 @@
 # ===---
 # Running on push & pull_request.
-#   This workflow parses the destination branch
-#   to choose correct dependencies revisions
+#   Builds opencl-clang in-tree and runs ninja check-opencl-clang.
+#   Parses the destination branch to choose correct dependency revisions.
 # ===---
 
-name: In-tree build
+name: In-tree Build and Test
 run-name: '${{ github.event_name }}: ${{ github.base_ref }} ${{ github.ref_name }}' # github.base_ref null for 'on: push'
 
 permissions:
@@ -43,3 +43,4 @@ jobs:
           ref_llvm: main
           ref_translator: main
           ref_opencl-clang: ${{ github.ref }}
+          run_tests: 'true'

.github/workflows/on-push-verification-out-of-tree.yml

@@ -2,7 +2,7 @@
 # Running on push & pull_request.
 # ===---
 
-name: Out-of-tree build
+name: Build
 run-name: '${{ github.event_name }}: ${{ github.base_ref }} ${{ github.ref_name }}' # github.base_ref null for 'on: push'
 
 permissions:
@@ -28,7 +28,7 @@ on:
 jobs:
 
   verify_default_branch:
-    name: linux
+    name: Out-of-tree Linux
     # ref_name for 'on: push'
     # base_ref for 'on: pull_request'
     runs-on: ubuntu-24.04

.github/workflows/scheduled-verification.yml

@@ -29,6 +29,7 @@ jobs:
           ref_llvm: main
           ref_translator: main
           ref_opencl-clang: main
+          run_tests: 'true'
 
   verify-release-branches:
     name: Verify `ocl-open-${{ matrix.llvm_version }}` release branch
@@ -47,3 +48,4 @@ jobs:
           ref_llvm: release/${{ matrix.llvm_version }}.x
           ref_translator: llvm_release_${{ matrix.llvm_version }}0
           ref_opencl-clang: ocl-open-${{ matrix.llvm_version }}0
+          run_tests: 'true'

CMakeLists.txt

@@ -340,3 +340,8 @@ if (WIN32)
 else (WIN32)
     SET_LINUX_EXPORTS_FILE( ${TARGET_NAME} opencl_clang.map )
 endif(WIN32)
+
+if(LLVM_INCLUDE_TESTS)
+  set(OPENCL_CLANG_SOURCE_DIR ${CMAKE_CURRENT_SOURCE_DIR})
+  add_subdirectory(tests)
+endif()

opencl_clang_options.td

@@ -44,3 +44,6 @@ def target_triple : Separate<["-"], "target-triple">,  HelpText<"Specify target
 def spir_std_1_0: Flag<["-"], "spir-std=1.0">;
 def spir_std_1_2: Flag<["-"], "spir-std=1.2">;
 def x : Separate<["-"], "x">;
+def Xclang : Separate<["-"], "Xclang">, HelpText<"Pass argument to clang -cc1">;
+def W_Joined : Joined<["-"], "W">, HelpText<"Enable/disable warning flags">;
+def W_Separate : Separate<["-"], "W">;

options_compile.cpp

@@ -146,6 +146,14 @@ std::string EffectiveOptionsFilter::processOptions(const OpenCLArgList &args,
     case OPT_COMPILE_D:
     case OPT_COMPILE_I:
     case OPT_COMPILE_Werror:
+    case OPT_COMPILE_W_Joined:
+    case OPT_COMPILE_W_Separate:
+      effectiveArgs.push_back((*it)->getAsString(args));
+      break;
+    case OPT_COMPILE_Xclang:
+      effectiveArgs.push_back("-Xclang");
+      effectiveArgs.push_back((*it)->getValue());
+      break;
     case OPT_COMPILE_cl_single_precision_constant:
     case OPT_COMPILE_cl_fp32_correctly_rounded_divide_sqrt:
     case OPT_COMPILE_cl_opt_disable:

tests/CMakeLists.txt

@@ -0,0 +1,29 @@
+set(LLVM_TOOLS_DIR ${LLVM_TOOLS_BINARY_DIR})
+
+set(OPENCL_CLANG_TEST_DEVICE "DEFAULT" CACHE STRING "Device name for opencl-clang lit tests (section key in ConfExt.ini)")
+
+add_subdirectory(occ-cli)
+
+# Set the depends list as a variable so that it can grow conditionally.
+# NOTE: Sync the substitutions in test/lit.cfg when adding to this list.
+set(OPENCL_CLANG_TEST_DEPENDS
+  FileCheck
+  llvm-dis
+  count
+  not
+  occ-cli
+  ${TARGET_NAME}
+)
+
+configure_lit_site_cfg(
+  ${CMAKE_CURRENT_SOURCE_DIR}/lit.site.cfg.py.in
+  ${CMAKE_CURRENT_BINARY_DIR}/lit.site.cfg.py
+  MAIN_CONFIG
+  ${CMAKE_CURRENT_SOURCE_DIR}/lit.cfg.py
+)
+
+add_lit_testsuite(check-opencl-clang "Running the OpenCL Clang regression tests"
+  ${CMAKE_CURRENT_BINARY_DIR}
+  DEPENDS ${OPENCL_CLANG_TEST_DEPENDS}
+  )
+set_target_properties(check-opencl-clang PROPERTIES FOLDER "OpenCL-Clang Tests")

tests/Conformance/api/kernel0.cl

@@ -0,0 +1,7 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global int *dst, int count) {
+  int tid = get_global_id(0);
+
+  for (int i = 0; i < count; i++)
+    dst[i] = tid + i;
+}

tests/Conformance/api/kernel1.cl

@@ -0,0 +1,6 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global float *src, __global int *dst) {
+  int tid = get_global_id(0);
+
+  dst[tid] = (int)src[tid] + 1;
+}

tests/Conformance/api/kernel10.cl

@@ -0,0 +1,11 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global float *src, __global int *dst) {
+  int tid = get_global_id(0);
+
+  dst[tid] = (int)src[tid];
+}
+__kernel void sample_test2(__global int *src, __global float *dst) {
+  int tid = get_global_id(0);
+
+  dst[tid] = (float)src[tid];
+}

tests/Conformance/api/kernel100.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global float4 *src1, __global float8 *src2,
+                          __global int8 *src3, __global float4 *dst1,
+                          __global float8 *dst2, __global int8 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel101.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global short4 *src1, __global short8 *src2,
+                          __global char8 *src3, __global short4 *dst1,
+                          __global short8 *dst2, __global char8 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel102.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global short4 *src1, __global short8 *src2,
+                          __global float8 *src3, __global short4 *dst1,
+                          __global short8 *dst2, __global float8 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel103.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global short4 *src1, __global short8 *src2,
+                          __global short8 *src3, __global short4 *dst1,
+                          __global short8 *dst2, __global short8 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel104.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global char8 *src1, __global short2 *src2,
+                          __global char2 *src3, __global char8 *dst1,
+                          __global short2 *dst2, __global char2 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel105.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global int8 *src1, __global short2 *src2,
+                          __global int2 *src3, __global int8 *dst1,
+                          __global short2 *dst2, __global int2 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel106.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global float8 *src1, __global float2 *src2,
+                          __global char2 *src3, __global float8 *dst1,
+                          __global float2 *dst2, __global char2 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel107.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global short8 *src1, __global float2 *src2,
+                          __global short2 *src3, __global short8 *dst1,
+                          __global float2 *dst2, __global short2 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel108.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global float8 *src1, __global int2 *src2,
+                          __global char4 *src3, __global float8 *dst1,
+                          __global int2 *dst2, __global char4 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel109.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global short8 *src1, __global float2 *src2,
+                          __global char4 *src3, __global short8 *dst1,
+                          __global float2 *dst2, __global char4 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel110.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global char8 *src1, __global float2 *src2,
+                          __global int4 *src3, __global char8 *dst1,
+                          __global float2 *dst2, __global int4 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel111.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global int8 *src1, __global int2 *src2,
+                          __global char4 *src3, __global int8 *dst1,
+                          __global int2 *dst2, __global char4 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel112.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global int8 *src1, __global float2 *src2,
+                          __global char8 *src3, __global int8 *dst1,
+                          __global float2 *dst2, __global char8 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel113.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global int8 *src1, __global char2 *src2,
+                          __global float8 *src3, __global int8 *dst1,
+                          __global char2 *dst2, __global float8 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel114.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global char8 *src1, __global int2 *src2,
+                          __global float8 *src3, __global char8 *dst1,
+                          __global int2 *dst2, __global float8 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel115.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global char8 *src1, __global float2 *src2,
+                          __global char8 *src3, __global char8 *dst1,
+                          __global float2 *dst2, __global char8 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel116.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global short8 *src1, __global float4 *src2,
+                          __global char2 *src3, __global short8 *dst1,
+                          __global float4 *dst2, __global char2 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel117.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global float8 *src1, __global char4 *src2,
+                          __global short2 *src3, __global float8 *dst1,
+                          __global char4 *dst2, __global short2 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel118.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global float8 *src1, __global char4 *src2,
+                          __global float2 *src3, __global float8 *dst1,
+                          __global char4 *dst2, __global float2 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}

tests/Conformance/api/kernel119.cl

@@ -0,0 +1,9 @@
+// RUN: %occ-cli %s --cl-options= --cl-device=%cl_device %cfg_path
+__kernel void sample_test(__global int8 *src1, __global int4 *src2,
+                          __global char2 *src3, __global int8 *dst1,
+                          __global int4 *dst2, __global char2 *dst3) {
+  int tid = get_global_id(0);
+  dst1[tid] = src1[tid];
+  dst2[tid] = src2[tid];
+  dst3[tid] = src3[tid];
+}