Merge branch 'main' into object_comparability_context_devices

.github/workflows/presubmit.yml

@@ -27,10 +27,29 @@ jobs:
             arch: aarch64
             debug: 1
             extra: " debug"
+          - os: ubuntu-22.04
+            mainmatrix: false
+            arch: android-arm
+            android_arch_abi: armeabi-v7a
+          - os: ubuntu-22.04
+            mainmatrix: false
+            arch: android-aarch64
+            android_arch_abi: arm64-v8a
     steps:
       - uses: actions/checkout@v4
       - name: Setup Ninja
         uses: seanmiddleditch/gha-setup-ninja@master
+      - name: Install Arm and AArch64 compilers
+        if: ${{ matrix.arch == 'arm' || matrix.arch == 'aarch64' }}
+        run: |
+          sudo apt-get update
+          sudo apt-get install -y \
+            gcc-12-arm-linux-gnueabihf g++-12-arm-linux-gnueabihf \
+            gcc-12-aarch64-linux-gnu g++-12-aarch64-linux-gnu
+          sudo update-alternatives --install /usr/bin/arm-linux-gnueabihf-gcc arm-linux-gnueabihf-gcc /usr/bin/arm-linux-gnueabihf-gcc-12 12
+          sudo update-alternatives --install /usr/bin/arm-linux-gnueabihf-g++ arm-linux-gnueabihf-g++ /usr/bin/arm-linux-gnueabihf-g++-12 12
+          sudo update-alternatives --install /usr/bin/aarch64-linux-gnu-gcc aarch64-linux-gnu-gcc /usr/bin/aarch64-linux-gnu-gcc-12 12
+          sudo update-alternatives --install /usr/bin/aarch64-linux-gnu-g++ aarch64-linux-gnu-g++ /usr/bin/aarch64-linux-gnu-g++-12 12
       - name: Setup OpenGL build dependencies
         if: ${{ matrix.gl }}
         run: |
@@ -55,6 +74,15 @@ jobs:
         with:
           version: 1.3.275.0
           cache: true
+      - name: Install Android NDK
+        if: ${{ matrix.arch == 'android-arm' || matrix.arch == 'android-aarch64' }}
+        run: |
+          wget https://dl.google.com/android/repository/android-ndk-r27c-linux.zip -O android-ndk.zip
+          unzip android-ndk.zip -d $HOME
+          export ANDROID_NDK=$HOME/android-ndk-r27c
+          echo "ANDROID_NDK=$ANDROID_NDK" >> $GITHUB_ENV
+          export ANDROID_ARCH_ABI=${{ matrix.android_arch_abi }}
+          echo "ANDROID_ARCH_ABI=$ANDROID_ARCH_ABI" >> $GITHUB_ENV
       - name: Build
         shell: bash
         run: ./presubmit.sh

.gitignore

@@ -0,0 +1,2 @@
+# Build dir
+[Bb]uild/

CMakeLists.txt

@@ -104,7 +104,6 @@ if(CMAKE_COMPILER_IS_GNUCC OR "${CMAKE_CXX_COMPILER_ID}" MATCHES "(Apple)?Clang"
         # Enable more warnings if not doing a release build.
         add_cxx_flag_if_supported(-Wall)
     endif()
-    add_cxx_flag_if_supported(-Wno-format)
     add_cxx_flag_if_supported(-Wno-error=cpp) # Allow #warning directive
     add_cxx_flag_if_supported(-Wno-unknown-pragmas) # Issue #785
     add_cxx_flag_if_supported(-Wno-error=asm-operand-widths) # Issue #784

presubmit.sh

@@ -4,40 +4,37 @@ set -e
 
 export TOP=$(pwd)
 
-TOOLCHAIN_URL_arm="https://releases.linaro.org/components/toolchain/binaries/7.5-2019.12/arm-linux-gnueabihf/gcc-linaro-7.5.0-2019.12-x86_64_arm-linux-gnueabihf.tar.xz"
-TOOLCHAIN_URL_aarch64="https://releases.linaro.org/components/toolchain/binaries/7.5-2019.12/aarch64-linux-gnu/gcc-linaro-7.5.0-2019.12-x86_64_aarch64-linux-gnu.tar.xz"
-
 TOOLCHAIN_PREFIX_arm=arm-linux-gnueabihf
 TOOLCHAIN_PREFIX_aarch64=aarch64-linux-gnu
 
-TOOLCHAIN_FILE=${TOP}/toolchain.cmake
-touch ${TOOLCHAIN_FILE}
+if [[ ${JOB_ARCHITECTURE} == android-* ]]; then
+    TOOLCHAIN_FILE=${ANDROID_NDK}/build/cmake/android.toolchain.cmake
+    CMAKE_CONFIG_ARGS_ANDROID="-DCMAKE_ANDROID_ARCH_ABI=${ANDROID_ARCH_ABI}"
+else
+    TOOLCHAIN_FILE=${TOP}/toolchain.cmake
+    touch ${TOOLCHAIN_FILE}
+fi
+
 BUILD_OPENGL_TEST="OFF"
 BUILD_VULKAN_TEST="ON"
 
 cmake --version
 echo
 
 # Prepare toolchain if needed
-if [[ ${JOB_ARCHITECTURE} != "" && ${RUNNER_OS} != "Windows" ]]; then
-    TOOLCHAIN_URL_VAR=TOOLCHAIN_URL_${JOB_ARCHITECTURE}
-    TOOLCHAIN_URL=${!TOOLCHAIN_URL_VAR}
-    wget ${TOOLCHAIN_URL}
-    TOOLCHAIN_ARCHIVE=${TOOLCHAIN_URL##*/}
-    tar xf ${TOOLCHAIN_ARCHIVE}
-    TOOLCHAIN_DIR=${TOP}/${TOOLCHAIN_ARCHIVE%.tar.xz}
-    export PATH=${TOOLCHAIN_DIR}/bin:${PATH}
+if [[ ${JOB_ARCHITECTURE} != android-* ]]; then
+    if [[ ${JOB_ARCHITECTURE} != "" && ${RUNNER_OS} != "Windows" ]]; then
+        TOOLCHAIN_PREFIX_VAR=TOOLCHAIN_PREFIX_${JOB_ARCHITECTURE}
+        TOOLCHAIN_PREFIX=${!TOOLCHAIN_PREFIX_VAR}
 
-    TOOLCHAIN_PREFIX_VAR=TOOLCHAIN_PREFIX_${JOB_ARCHITECTURE}
-    TOOLCHAIN_PREFIX=${!TOOLCHAIN_PREFIX_VAR}
-
-    echo "SET(CMAKE_SYSTEM_NAME Linux)" >> ${TOOLCHAIN_FILE}
-    echo "SET(CMAKE_SYSTEM_PROCESSOR ${JOB_ARCHITECTURE})" >> ${TOOLCHAIN_FILE}
-    echo "SET(CMAKE_C_COMPILER   ${TOOLCHAIN_PREFIX}-gcc)" >> ${TOOLCHAIN_FILE}
-    echo "SET(CMAKE_CXX_COMPILER ${TOOLCHAIN_PREFIX}-g++)" >> ${TOOLCHAIN_FILE}
-    echo "SET(CMAKE_FIND_ROOT_PATH_MODE_PROGRAM NEVER)" >> ${TOOLCHAIN_FILE}
-    echo "SET(CMAKE_FIND_ROOT_PATH_MODE_LIBRARY ONLY)" >> ${TOOLCHAIN_FILE}
-    echo "SET(CMAKE_FIND_ROOT_PATH_MODE_INCLUDE ONLY)" >> ${TOOLCHAIN_FILE}
+        echo "SET(CMAKE_SYSTEM_NAME Linux)" >> ${TOOLCHAIN_FILE}
+        echo "SET(CMAKE_SYSTEM_PROCESSOR ${JOB_ARCHITECTURE})" >> ${TOOLCHAIN_FILE}
+        echo "SET(CMAKE_C_COMPILER   ${TOOLCHAIN_PREFIX}-gcc)" >> ${TOOLCHAIN_FILE}
+        echo "SET(CMAKE_CXX_COMPILER ${TOOLCHAIN_PREFIX}-g++)" >> ${TOOLCHAIN_FILE}
+        echo "SET(CMAKE_FIND_ROOT_PATH_MODE_PROGRAM NEVER)" >> ${TOOLCHAIN_FILE}
+        echo "SET(CMAKE_FIND_ROOT_PATH_MODE_LIBRARY ONLY)" >> ${TOOLCHAIN_FILE}
+        echo "SET(CMAKE_FIND_ROOT_PATH_MODE_INCLUDE ONLY)" >> ${TOOLCHAIN_FILE}
+    fi
 fi
 
 if [[ ( ${JOB_ARCHITECTURE} == "" && ${JOB_ENABLE_GL} == "1" ) ]]; then
@@ -61,34 +58,42 @@ cd build
 cmake .. -G Ninja \
       -DCMAKE_BUILD_TYPE=Release \
       -DCMAKE_TOOLCHAIN_FILE=${TOOLCHAIN_FILE} \
-      -DOPENCL_ICD_LOADER_HEADERS_DIR=${TOP}/OpenCL-Headers/
+      -DOPENCL_ICD_LOADER_HEADERS_DIR=${TOP}/OpenCL-Headers/ \
+      "${CMAKE_CONFIG_ARGS_ANDROID}"
 cmake --build . --parallel
 
 #Vulkan Loader
-cd ${TOP}
-git clone https://github.com/KhronosGroup/Vulkan-Loader.git
-cd Vulkan-Loader
-mkdir build
-cd build
-python3 ../scripts/update_deps.py
-cmake .. -G Ninja \
-      -DCMAKE_BUILD_TYPE=Release \
-      -DCMAKE_TOOLCHAIN_FILE=${TOOLCHAIN_FILE} \
-      -DBUILD_WSI_XLIB_SUPPORT=OFF \
-      -DBUILD_WSI_XCB_SUPPORT=OFF \
-      -DBUILD_WSI_WAYLAND_SUPPORT=OFF \
-      -C helper.cmake ..
-cmake --build . --parallel
+if [[ ${JOB_ARCHITECTURE} != android-* ]]; then
+    # Building the Vulkan loader is not supported on Android,
+    # instead, the loader is shipped as part of the operating system
+    cd ${TOP}
+    git clone https://github.com/KhronosGroup/Vulkan-Loader.git
+    cd Vulkan-Loader
+    mkdir build
+    cd build
+    python3 ../scripts/update_deps.py
+    cmake .. -G Ninja \
+          -DCMAKE_BUILD_TYPE=Release \
+          -DCMAKE_TOOLCHAIN_FILE=${TOOLCHAIN_FILE} \
+          -DBUILD_WSI_XLIB_SUPPORT=OFF \
+          -DBUILD_WSI_XCB_SUPPORT=OFF \
+          -DBUILD_WSI_WAYLAND_SUPPORT=OFF \
+          -C helper.cmake ..
+    cmake --build . --parallel
+fi
 
 # Build CTS
 cd ${TOP}
 ls -l
 mkdir build
 cd build
 if [[ ${RUNNER_OS} == "Windows" ]]; then
-  CMAKE_OPENCL_LIBRARIES_OPTION="OpenCL"
+    CMAKE_OPENCL_LIBRARIES_OPTION="OpenCL"
 else
-  CMAKE_OPENCL_LIBRARIES_OPTION="-lOpenCL -lpthread"
+    CMAKE_OPENCL_LIBRARIES_OPTION="-lOpenCL"
+    if [[ ${JOB_ARCHITECTURE} != android-* ]]; then
+        CMAKE_OPENCL_LIBRARIES_OPTION="${CMAKE_OPENCL_LIBRARIES_OPTION} -lpthread"
+    fi
 fi
 cmake .. -G Ninja \
       -DCMAKE_BUILD_TYPE="${BUILD_CONFIG}" \
@@ -102,5 +107,6 @@ cmake .. -G Ninja \
       -DGL_IS_SUPPORTED=${BUILD_OPENGL_TEST} \
       -DVULKAN_IS_SUPPORTED=${BUILD_VULKAN_TEST} \
       -DVULKAN_INCLUDE_DIR=${TOP}/Vulkan-Headers/include/ \
-      -DVULKAN_LIB_DIR=${TOP}/Vulkan-Loader/build/loader/
+      -DVULKAN_LIB_DIR=${TOP}/Vulkan-Loader/build/loader/ \
+      "${CMAKE_CONFIG_ARGS_ANDROID}"
 cmake --build . --parallel

test_common/harness/imageHelpers.cpp

@@ -490,6 +490,32 @@ size_t compare_scanlines(const image_descriptor *imageInfo, const char *aPtr,
             }
             break;
 
+            case CL_SNORM_INT8: {
+                cl_uchar aPixel = *(cl_uchar *)aPtr;
+                cl_uchar bPixel = *(cl_uchar *)bPtr;
+                // -1.0 is defined as 0x80 and 0x81
+                aPixel = (aPixel == 0x80) ? 0x81 : aPixel;
+                bPixel = (bPixel == 0x80) ? 0x81 : bPixel;
+                if (aPixel != bPixel)
+                {
+                    return column;
+                }
+            }
+            break;
+
+            case CL_SNORM_INT16: {
+                cl_ushort aPixel = *(cl_ushort *)aPtr;
+                cl_ushort bPixel = *(cl_ushort *)bPtr;
+                // -1.0 is defined as 0x8000 and 0x8001
+                aPixel = (aPixel == 0x8000) ? 0x8001 : aPixel;
+                bPixel = (bPixel == 0x8000) ? 0x8001 : bPixel;
+                if (aPixel != bPixel)
+                {
+                    return column;
+                }
+            }
+            break;
+
             default:
                 if (memcmp(aPtr, bPtr, pixel_size) != 0) return column;
                 break;

test_common/harness/kernelHelpers.h

@@ -39,21 +39,6 @@
 
 #include <functional>
 
-/*
- *  The below code is intended to be used at the top of kernels that appear
- * inline in files to set line and file info for the kernel:
- *
- *  const char *source = {
- *      INIT_OPENCL_DEBUG_INFO
- *      "__kernel void foo( int x )\n"
- *      "{\n"
- *      "   ...\n"
- *      "}\n"
- *  };
- */
-#define INIT_OPENCL_DEBUG_INFO SET_OPENCL_LINE_INFO(__LINE__, __FILE__)
-#define SET_OPENCL_LINE_INFO(_line, _file)                                     \
-    "#line " STRINGIFY(_line) " " STRINGIFY(_file) "\n"
 #ifndef STRINGIFY_VALUE
 #define STRINGIFY_VALUE(_x) STRINGIFY(_x)
 #endif
@@ -85,13 +70,6 @@ extern int create_single_kernel_helper_create_program_for_device(
     unsigned int numKernelLines, const char **kernelProgram,
     const char *buildOptions = NULL);
 
-/* Creates OpenCL C++ program. This one must be used for creating OpenCL C++
- * program. */
-extern int create_openclcpp_program(cl_context context, cl_program *outProgram,
-                                    unsigned int numKernelLines,
-                                    const char **kernelProgram,
-                                    const char *buildOptions = NULL);
-
 /* Builds program (outProgram) and creates one kernel */
 int build_program_create_kernel_helper(
     cl_context context, cl_program *outProgram, cl_kernel *outKernel,

test_common/harness/testHarness.cpp

@@ -62,6 +62,27 @@ bool gCoreILProgram = true;
 
 #define DEFAULT_NUM_ELEMENTS 0x4000
 
+test_definition *test_registry::definitions() { return &m_definitions[0]; }
+
+size_t test_registry::num_tests() { return m_definitions.size(); }
+
+void test_registry::add_test(test *t, const char *name, Version version)
+{
+
+    m_tests.push_back(t);
+    test_definition testDef;
+    testDef.func = t->getFunction();
+    testDef.name = name;
+    testDef.min_version = version;
+    m_definitions.push_back(testDef);
+}
+
+test_registry &test_registry::getInstance()
+{
+    static test_registry instance;
+    return instance;
+}
+
 static int saveResultsToJson(const char *suiteName, test_definition testList[],
                              unsigned char selectedTestList[],
                              test_status resultTestList[], int testNum)
@@ -689,6 +710,7 @@ static void print_results(int failed, int count, const char *name)
             log_error("FAILED %s.\n", name);
         }
     }
+    fflush(stdout);
 }
 
 int parseAndCallCommandLineTests(int argc, const char *argv[],

test_common/harness/testHarness.h

@@ -21,11 +21,15 @@
 #include <sstream>
 
 #include <string>
+#include <vector>
 
 class Version {
 public:
     Version(): m_major(0), m_minor(0) {}
+
     Version(cl_uint major, cl_uint minor): m_major(major), m_minor(minor) {}
+    int major() const { return m_major; }
+    int minor() const { return m_minor; }
     bool operator>(const Version &rhs) const
     {
         return to_uint() > rhs.to_uint();
@@ -99,6 +103,64 @@ struct test_harness_config
     unsigned numWorkerThreads;
 };
 
+
+struct test
+{
+    virtual test_function_pointer getFunction() = 0;
+};
+
+class test_registry {
+private:
+    std::vector<test *> m_tests;
+    std::vector<test_definition> m_definitions;
+
+public:
+    static test_registry &getInstance();
+
+    test_definition *definitions();
+
+    size_t num_tests();
+
+    void add_test(test *t, const char *name, Version version);
+    test_registry() {}
+};
+
+template <typename T> T *register_test(const char *name, Version version)
+{
+    T *t = new T();
+    test_registry::getInstance().add_test((test *)t, name, version);
+    return t;
+}
+
+#define REGISTER_TEST_VERSION(name, version)                                   \
+    extern int test_##name(cl_device_id device, cl_context context,            \
+                           cl_command_queue queue, int num_elements);          \
+    class test_##name##_class : public test {                                  \
+    private:                                                                   \
+        test_function_pointer fn;                                              \
+                                                                               \
+    public:                                                                    \
+        test_##name##_class(): fn(test_##name) {}                              \
+        test_function_pointer getFunction() { return fn; }                     \
+    };                                                                         \
+    test_##name##_class *var_##name =                                          \
+        register_test<test_##name##_class>(#name, version);                    \
+    int test_##name(cl_device_id device, cl_context context,                   \
+                    cl_command_queue queue, int num_elements)
+
+#define REGISTER_TEST(name) REGISTER_TEST_VERSION(name, Version(1, 2))
+
+#define REQUIRE_EXTENSION(name)                                                \
+    do                                                                         \
+    {                                                                          \
+        if (!is_extension_available(deviceID, name))                           \
+        {                                                                      \
+            log_info(name                                                      \
+                     " is not supported on this device. Skipping test.\n");    \
+            return TEST_SKIPPED_ITSELF;                                        \
+        }                                                                      \
+    } while (0)
+
 extern int gFailCount;
 extern int gTestCount;
 extern cl_uint gReSeed;

test_conformance/SVM/common.h

@@ -81,25 +81,6 @@ extern cl_int        verify_linked_lists_on_device(int qi, cl_command_queue q, c
 extern cl_int create_linked_lists_on_device_no_map(int qi, cl_command_queue q, size_t *pAllocator,   cl_kernel k, size_t numLists  );
 extern cl_int verify_linked_lists_on_device_no_map(int qi, cl_command_queue q, cl_int *pNum_correct, cl_kernel k, cl_int ListLength, size_t numLists  );
 
-extern int    test_svm_byte_granularity(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_set_kernel_exec_info_svm_ptrs(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_fine_grain_memory_consistency(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_fine_grain_sync_buffers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_shared_address_space_coarse_grain_old_api(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_shared_address_space_coarse_grain_new_api(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_shared_address_space_fine_grain_buffers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_shared_address_space_fine_grain(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_cross_buffer_pointers_coarse_grain(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_pointer_passing(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_allocate_shared_buffer(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int test_svm_allocate_shared_buffer_negative(cl_device_id deviceID,
-                                                    cl_context context,
-                                                    cl_command_queue queue,
-                                                    int num_elements);
-extern int    test_svm_shared_sub_buffers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_enqueue_api(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-extern int    test_svm_migrate(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
-
 extern cl_int create_cl_objects(cl_device_id device_from_harness, const char** ppCodeString, cl_context* context, cl_program *program, cl_command_queue *queues, cl_uint *num_devices, cl_device_svm_capabilities required_svm_caps, std::vector<std::string> extensions_list = std::vector<std::string>());
 
 extern const char *linked_list_create_and_verify_kernels[];