From patchwork Fri Dec 24 08:51:04 2021 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 7bit X-Patchwork-Submitter: "Jiang, Haochen" X-Patchwork-Id: 1573000 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@bilbo.ozlabs.org Authentication-Results: bilbo.ozlabs.org; dkim=pass (1024-bit key; unprotected) header.d=gcc.gnu.org header.i=@gcc.gnu.org header.a=rsa-sha256 header.s=default header.b=BToJgNM2; dkim-atps=neutral Authentication-Results: ozlabs.org; spf=pass (sender SPF authorized) smtp.mailfrom=gcc.gnu.org (client-ip=8.43.85.97; helo=sourceware.org; envelope-from=gcc-patches-bounces+incoming=patchwork.ozlabs.org@gcc.gnu.org; receiver=) Received: from sourceware.org (server2.sourceware.org [8.43.85.97]) (using TLSv1.3 with cipher TLS_AES_256_GCM_SHA384 (256/256 bits) key-exchange X25519 server-signature RSA-PSS (4096 bits) server-digest SHA256) (No client certificate requested) by bilbo.ozlabs.org (Postfix) with ESMTPS id 4JL1500820z9sRK for ; Fri, 24 Dec 2021 19:51:30 +1100 (AEDT) Received: from server2.sourceware.org (localhost [IPv6:::1]) by sourceware.org (Postfix) with ESMTP id D30953858424 for ; Fri, 24 Dec 2021 08:51:27 +0000 (GMT) DKIM-Filter: OpenDKIM Filter v2.11.0 sourceware.org D30953858424 DKIM-Signature: v=1; a=rsa-sha256; c=relaxed/relaxed; d=gcc.gnu.org; s=default; t=1640335887; bh=D/ofpHcqnYA0YG4TijGkbY0IRVD44293fwAYBF33GHA=; h=To:Subject:Date:List-Id:List-Unsubscribe:List-Archive:List-Post: List-Help:List-Subscribe:From:Reply-To:Cc:From; b=BToJgNM2s2gE+rMCUJbPf2lbe9/8plHD92dlKXJT9z7+0Rh0ucH+Pk5f2OJ2adeAg O1t+U6oTYYLthjgdnXmC+R5n7ucDzjk6U72sf9roNdMwSvQ8WWAMpFrpZZtZlLyhrT aCkIko0SBrXzimoMUscaZfnj04Mh4RfUG0D7f1G8= X-Original-To: gcc-patches@gcc.gnu.org Delivered-To: gcc-patches@gcc.gnu.org Received: from mga14.intel.com (mga14.intel.com [192.55.52.115]) by sourceware.org (Postfix) with ESMTPS id 42B243858D28 for ; Fri, 24 Dec 2021 08:51:07 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.4.1 sourceware.org 42B243858D28 X-IronPort-AV: E=McAfee;i="6200,9189,10207"; a="241176902" X-IronPort-AV: E=Sophos;i="5.88,232,1635231600"; d="scan'208";a="241176902" Received: from fmsmga001.fm.intel.com ([10.253.24.23]) by fmsmga103.fm.intel.com with ESMTP/TLS/ECDHE-RSA-AES256-GCM-SHA384; 24 Dec 2021 00:51:06 -0800 X-ExtLoop1: 1 X-IronPort-AV: E=Sophos;i="5.88,232,1635231600"; d="scan'208";a="664815610" Received: from scymds01.sc.intel.com ([10.148.94.138]) by fmsmga001.fm.intel.com with ESMTP; 24 Dec 2021 00:51:06 -0800 Received: from shliclel320.sh.intel.com (shliclel320.sh.intel.com [10.239.236.50]) by scymds01.sc.intel.com with ESMTP id 1BO8p4SU009751; Fri, 24 Dec 2021 00:51:05 -0800 To: gcc-patches@gcc.gnu.org Subject: [PATCH] [i386]Fix tdpbf16ps testcase Date: Fri, 24 Dec 2021 16:51:04 +0800 Message-Id: <20211224085104.68236-1-haochen.jiang@intel.com> X-Mailer: git-send-email 2.18.1 X-Spam-Status: No, score=-12.4 required=5.0 tests=BAYES_00, DKIMWL_WL_HIGH, DKIM_SIGNED, DKIM_VALID, DKIM_VALID_AU, DKIM_VALID_EF, GIT_PATCH_0, KAM_SHORT, SPF_HELO_NONE, SPF_NONE, TXREP autolearn=ham autolearn_force=no version=3.4.4 X-Spam-Checker-Version: SpamAssassin 3.4.4 (2020-01-24) on server2.sourceware.org X-BeenThere: gcc-patches@gcc.gnu.org X-Mailman-Version: 2.1.29 Precedence: list List-Id: Gcc-patches mailing list List-Unsubscribe: , List-Archive: List-Post: List-Help: List-Subscribe: , X-Patchwork-Original-From: Haochen Jiang via Gcc-patches From: "Jiang, Haochen" Reply-To: Haochen Jiang Cc: hongtao.liu@intel.com Errors-To: gcc-patches-bounces+incoming=patchwork.ozlabs.org@gcc.gnu.org Sender: "Gcc-patches" Hi all, This patch fix the testcase of amxbf16-dpbf16ps-2.c. Previously the type convert has some issue. Ok for trunk? BRs, Haochen gcc/testsuite/ChangeLog: * gcc.target/i386/amx-check.h (check_float_tile_register): New check function for float to prevent precision loss. * gcc.target/i386/amxbf16-dpbf16ps-2.c: Correct the type convert and byte offset. Use the new check function. --- gcc/testsuite/gcc.target/i386/amx-check.h | 23 ++++++++++++-- .../gcc.target/i386/amxbf16-dpbf16ps-2.c | 30 ++++++++++++------- 2 files changed, 41 insertions(+), 12 deletions(-) diff --git a/gcc/testsuite/gcc.target/i386/amx-check.h b/gcc/testsuite/gcc.target/i386/amx-check.h index 03616ff0b8e..434b0e59703 100644 --- a/gcc/testsuite/gcc.target/i386/amx-check.h +++ b/gcc/testsuite/gcc.target/i386/amx-check.h @@ -139,8 +139,27 @@ int check_tile_register (__tile* ref, __tile* target) for (i = 0; i < rows; i++) for (j = 0; j < colsb; j++) - if (ref->buf[i * colsb + j] != target->buf[i * colsb + j]) - return 0; + if (ref->buf[i * colsb + j] != target->buf[i * colsb + j]) + return 0; + + return 1; +} + +/* Compare float tile register value with __tile variable */ +int check_float_tile_register (__tile* ref, __tile* target) +{ + /* Tile register should be stored from tmm to + memory and compare with emulation results. */ + int rows = target->rows; + int colsb = target->colsb / 4; + int i, j; + uint32_t *ref_buf = (uint32_t *) ref->buf; + uint32_t *target_buf = (uint32_t *) target->buf; + + for (i = 0; i < rows; i++) + for (j = 0; j < colsb; j++) + if (abs(ref_buf[i * colsb + j] - target_buf[i * colsb + j]) > 1) + return 0; return 1; } diff --git a/gcc/testsuite/gcc.target/i386/amxbf16-dpbf16ps-2.c b/gcc/testsuite/gcc.target/i386/amxbf16-dpbf16ps-2.c index f7002ca5ea5..b00bc13ec78 100644 --- a/gcc/testsuite/gcc.target/i386/amxbf16-dpbf16ps-2.c +++ b/gcc/testsuite/gcc.target/i386/amxbf16-dpbf16ps-2.c @@ -12,15 +12,25 @@ void test_amx_bf16_dpbf16ps (); /* Transformation functions between bf16/float */ static uint16_t make_bf16 (float f) { - uint32_t u = (uint32_t)f; - u = (u >> 16) & 0xffff; - return (uint16_t)u; + union + { + float f; + uint32_t u; + } fu; + fu.f = f; + fu.u = (fu.u >> 16) & 0xffff; + return (uint16_t) fu.u; } static float make_f32 (uint16_t bf) { - uint32_t u = (uint32_t)(bf << 16); - return (float)u; + union + { + float f; + uint32_t u; + } fu; + fu.u = (uint32_t) bf << 16; + return fu.f; } /* Init tile buffer with bf16 pairs */ @@ -54,10 +64,10 @@ void calc_matrix_dpbf16ps (__tile *dst, __tile *src1, __tile *src2) for (t = 0; t < 2; t+=2) { dst_buf[i * N + k] += - (make_f32(src1_buf[i * 4 * N + 4 * j + t]) * - make_f32(src2_buf[j * 4 * K + 4 * k + t])) + - (make_f32(src1_buf[i * 4 * N + 4 * j + t + 1]) * - make_f32(src2_buf[j * 4 * K + 4 * k + t + 1])); + (make_f32(src1_buf[i * 2 * N + 2 * j + t]) * + make_f32(src2_buf[j * 2 * K + 2 * k + t])) + + (make_f32(src1_buf[i * 2 * N + 2 * j + t + 1]) * + make_f32(src2_buf[j * 2 * K + 2 * k + t + 1])); } } @@ -80,6 +90,6 @@ void test_amx_bf16_dpbf16ps () _tile_dpbf16ps (1, 2, 3); _tile_stored (1, dst_ref.buf, _STRIDE); - if (!check_tile_register (&dst_ref, &dst)) + if (!check_float_tile_register (&dst_ref, &dst)) abort(); }