From patchwork Fri Oct 18 06:22:25 2024 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 7bit X-Patchwork-Submitter: Tejas Belagod X-Patchwork-Id: 1998928 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@legolas.ozlabs.org Authentication-Results: legolas.ozlabs.org; dkim=pass (1024-bit key; unprotected) header.d=arm.com header.i=@arm.com header.a=rsa-sha256 header.s=selector1 header.b=nDl65WID; dkim=pass (1024-bit key) header.d=arm.com header.i=@arm.com header.a=rsa-sha256 header.s=selector1 header.b=nDl65WID; dkim-atps=neutral Authentication-Results: legolas.ozlabs.org; spf=pass (sender SPF authorized) smtp.mailfrom=gcc.gnu.org (client-ip=2620:52:3:1:0:246e:9693:128c; helo=server2.sourceware.org; envelope-from=gcc-patches-bounces~incoming=patchwork.ozlabs.org@gcc.gnu.org; receiver=patchwork.ozlabs.org) Received: from server2.sourceware.org (server2.sourceware.org [IPv6:2620:52:3:1:0:246e:9693:128c]) (using TLSv1.3 with cipher TLS_AES_256_GCM_SHA384 (256/256 bits) key-exchange X25519 server-signature ECDSA (secp384r1) server-digest SHA384) (No client certificate requested) by legolas.ozlabs.org (Postfix) with ESMTPS id 4XVF9m1rgRz1xw2 for ; Fri, 18 Oct 2024 17:27:24 +1100 (AEDT) Received: from server2.sourceware.org (localhost [IPv6:::1]) by sourceware.org (Postfix) with ESMTP id 6ADB13858423 for ; Fri, 18 Oct 2024 06:27:22 +0000 (GMT) X-Original-To: gcc-patches@gcc.gnu.org Delivered-To: gcc-patches@gcc.gnu.org Received: from EUR02-DB5-obe.outbound.protection.outlook.com (mail-db5eur02on20620.outbound.protection.outlook.com [IPv6:2a01:111:f403:2608::620]) by sourceware.org (Postfix) with ESMTPS id 543703858C2B for ; Fri, 18 Oct 2024 06:23:09 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.4.2 sourceware.org 543703858C2B Authentication-Results: sourceware.org; dmarc=pass (p=none dis=none) header.from=arm.com Authentication-Results: sourceware.org; spf=pass smtp.mailfrom=arm.com ARC-Filter: OpenARC Filter v1.0.0 sourceware.org 543703858C2B Authentication-Results: server2.sourceware.org; arc=pass smtp.remote-ip=2a01:111:f403:2608::620 ARC-Seal: i=3; a=rsa-sha256; d=sourceware.org; s=key; t=1729232600; cv=pass; b=mhDbiKRriWgMM3dUUOTpYmrTEEqNuVrbfmaOlfeWV2x+v8s9PMMcmuL6vbRb32n91BIh7ectUuLgmEcAwdBWW3udmzYIKz821IgAhgDy0YubS3vXylsANLbmkT5OzVFhwJcwQIOvkkiNYILlPj6hNJQf/QaTfpe1p3YZ0/JxJm4= ARC-Message-Signature: i=3; a=rsa-sha256; d=sourceware.org; s=key; t=1729232600; c=relaxed/simple; bh=yB04L0Tr9qZPI22nHsbZMQYTC7913TCwuoK47jNuzZ8=; h=DKIM-Signature:DKIM-Signature:From:To:Subject:Date:Message-ID: MIME-Version; b=w2h2wJUJ2rxDOHnL9DdaSgubT3PpiAH9JazWnZg4fHYOo7Q1e7Wenx6Ij6VpUzyc/2z3/Ld47Xky8pRg9nInG9f3cDyKYCf5vMlVSXLkTY9UyhYSzR4+QPifeYxKgk7guxVKAUdEG92mBDaeV5h7ANTt4fvrdOXEF5NPAYh32FQ= ARC-Authentication-Results: i=3; server2.sourceware.org ARC-Seal: i=2; a=rsa-sha256; s=arcselector10001; d=microsoft.com; cv=pass; b=PEtrkCk1S8fdDcT+6i4kxuKRPGeGfB5TptSb0sqVxhMZEynP0jyDJb0rieBvDG0QaqrLOz8ZcK0oZw11OD1DjagPKK+RXfGmqa3Huki1cAmVvNxW00D9JsoCDHI/OfyIW5nWnxjstj6cz8Bv5FK9J9CV9RtYdM9+MtcFga3HDG2GU9N/ZIxRj4o4Oe9Bxmimragb4yD97uN9RK4SS/FU0QT3fAvAGSkN2QhPdChK8q+SR2ZyLZX1s8GyqQYtzSfJb7qGYdTcTZJ/Z2S17otVzh81XyjPVaxfncqrknVZjaWwWoGqzboKMV1wKS2tOhADSIWZuUbvq0Dbw4nIX4zaZQ== ARC-Message-Signature: i=2; a=rsa-sha256; c=relaxed/relaxed; d=microsoft.com; s=arcselector10001; h=From:Date:Subject:Message-ID:Content-Type:MIME-Version:X-MS-Exchange-AntiSpam-MessageData-ChunkCount:X-MS-Exchange-AntiSpam-MessageData-0:X-MS-Exchange-AntiSpam-MessageData-1; bh=IdOQ2GnDp5k6jtANNBHdn0NqMquYisFOUfHbEX8S4z4=; b=St3YMIB/7Ti3Ox5PrrPNgu/EaQBRRii8Sv+7qt2f5o1RqLD1mq7kwvtgFadPA6c7e4B/BvTZtcv0st6rkbk0rk7YvU2tcm6Q2NFechWUMavVwFSKar1TB3fmC4V+ttHNAdv/BHcJq+NT3AyjMel6RVMAf1QZsdwtUM5AchCY8lvCxoIFXxlHa93W/e5Q7DGc1nVL21I6DSGHj8Erp+RE+1KaTpCmrqBfFYsNatJmHyHMNhARaP5pT70kgXL34mu8D2tEXB13EizTIKJn+cRpbEYk83ZSnX0Vs+Ne365sD7V+Tt6ZzOYWEYloinX9NN1Kk8um3hTH/dGPnZ236HLffA== ARC-Authentication-Results: i=2; mx.microsoft.com 1; spf=pass (sender ip is 63.35.35.123) smtp.rcpttodomain=gcc.gnu.org smtp.mailfrom=arm.com; dmarc=pass (p=none sp=none pct=100) action=none header.from=arm.com; dkim=pass (signature was verified) header.d=arm.com; arc=pass (0 oda=1 ltdi=1 spf=[1,1,smtp.mailfrom=arm.com] dmarc=[1,1,header.from=arm.com]) DKIM-Signature: v=1; a=rsa-sha256; c=relaxed/relaxed; d=arm.com; s=selector1; h=From:Date:Subject:Message-ID:Content-Type:MIME-Version:X-MS-Exchange-SenderADCheck; bh=IdOQ2GnDp5k6jtANNBHdn0NqMquYisFOUfHbEX8S4z4=; b=nDl65WIDgs26okfdVLhdusBNnYUGYcVpN166tJFjqnQs4zUPYocof1s/Z8Qj/H5Cn/h70G7KhaSb8ZWULVWcLJZLTUgYMqZa9n45wWEYCiBYVP38fwp1nTG/eyqU9n3rPpxybmjJ4xH1GNf7E73Z+RR5K+wTXy0Uc3/Cm2aCuWw= Received: from DBBPR09CA0017.eurprd09.prod.outlook.com (2603:10a6:10:c0::29) by PA4PR08MB6271.eurprd08.prod.outlook.com (2603:10a6:102:eb::19) with Microsoft SMTP Server (version=TLS1_2, cipher=TLS_ECDHE_RSA_WITH_AES_256_GCM_SHA384) id 15.20.8048.26; Fri, 18 Oct 2024 06:23:02 +0000 Received: from DB5PEPF00014B8C.eurprd02.prod.outlook.com (2603:10a6:10:c0:cafe::28) by DBBPR09CA0017.outlook.office365.com (2603:10a6:10:c0::29) with Microsoft SMTP Server (version=TLS1_2, cipher=TLS_ECDHE_RSA_WITH_AES_256_GCM_SHA384) id 15.20.8069.23 via Frontend Transport; Fri, 18 Oct 2024 06:23:02 +0000 X-MS-Exchange-Authentication-Results: spf=pass (sender IP is 63.35.35.123) smtp.mailfrom=arm.com; dkim=pass (signature was verified) header.d=arm.com;dmarc=pass action=none header.from=arm.com; Received-SPF: Pass (protection.outlook.com: domain of arm.com designates 63.35.35.123 as permitted sender) receiver=protection.outlook.com; client-ip=63.35.35.123; helo=64aa7808-outbound-1.mta.getcheckrecipient.com; pr=C Received: from 64aa7808-outbound-1.mta.getcheckrecipient.com (63.35.35.123) by DB5PEPF00014B8C.mail.protection.outlook.com (10.167.8.200) with Microsoft SMTP Server (version=TLS1_3, cipher=TLS_AES_256_GCM_SHA384) id 15.20.8069.17 via Frontend Transport; Fri, 18 Oct 2024 06:23:01 +0000 Received: ("Tessian outbound 60a4253641a2:v473"); Fri, 18 Oct 2024 06:23:01 +0000 X-CheckRecipientChecked: true X-CR-MTA-CID: 0cdbc22f0c44e9d9 X-TessianGatewayMetadata: Kde2r9BftBsTYbS5dxQ51mnRSGN0jum/2v3UNYrT3LerPH1CXg/dIFwtSXmg4gg3LGF2g7Sxzfz6JBvdpjKrxrxRcWxAfJMHwsRp+0RJ5/kydI0leXZkyxJRONNo1bWp7Vpny5AxFiEq6mG0M7Lwv8MQWlnbmjVWcDaIr/5M9sc= X-CR-MTA-TID: 64aa7808 Received: from Ldaadf3b0757a.1 by 64aa7808-outbound-1.mta.getcheckrecipient.com id 4B1FED84-1321-4C2A-8FDB-6D931637FFC4.1; Fri, 18 Oct 2024 06:22:55 +0000 Received: from EUR05-VI1-obe.outbound.protection.outlook.com by 64aa7808-outbound-1.mta.getcheckrecipient.com with ESMTPS id Ldaadf3b0757a.1 (version=TLSv1.2 cipher=ECDHE-RSA-AES256-GCM-SHA384); Fri, 18 Oct 2024 06:22:55 +0000 ARC-Seal: i=1; a=rsa-sha256; s=arcselector10001; d=microsoft.com; cv=none; b=hPhVbji7VztK4Yay5x2/aKkEmZ1GLYiZ+K6ANfsCJUM3hcBETwfdqSBKgbbdzG6vvnnWsQb/tnOQeOYJmku25QFHY++JyLUablj30etZlRjbmZq14rHaoAvV7fCt3O5WhDjjnztmMXvIN0a442TsKqllFvoo/WfsdjVfhnDp0y8I3l2pQ+MpF2+7FbqxaHBqDSbQnGKpPBg71Fk2t6A5IwfSQMN591MOciUtVsz/YPM0v81NgsbkNJSViSv9/grztC7tqGFud6T+7OOYnKzAW/BaDxVuwePOMoUTo1BTiQLIQFE9hUQG9LaGeRp16WtzRW6VBmWpOp1ln9392XVP+A== ARC-Message-Signature: i=1; a=rsa-sha256; c=relaxed/relaxed; d=microsoft.com; s=arcselector10001; h=From:Date:Subject:Message-ID:Content-Type:MIME-Version:X-MS-Exchange-AntiSpam-MessageData-ChunkCount:X-MS-Exchange-AntiSpam-MessageData-0:X-MS-Exchange-AntiSpam-MessageData-1; bh=IdOQ2GnDp5k6jtANNBHdn0NqMquYisFOUfHbEX8S4z4=; b=ey+jgrB0OVX1/L+uHnydMLtgIAiV73oJ5aeWBGy5pHrQCGFtQBFPhRZzE0NuQwNuSnodWwaCdeqSUSy3eAeuHweJMv0CpN2XhDNrPTHgv3ASqRr8167xYF4WBEl8qaw9g3e2NtSDmSbk6Vhq7SkUK7knu/imaO1S9SDbaZLj1gEfafzAz0OfPYBFL6VKgbYIs6Zddo4CDRPHvZxNay/jm9cuc/WWei7sjL5KHgTlNw8GEJNXZI/DT44fIj8kOWr/wKheD9Q3hG6A38hSukPNBCkBfeNIQdkmxsBm/0Xl+mbEhvpkcNBNw+77TvGAcF5HLqvkLoEhZf8fE9dZ898O+Q== ARC-Authentication-Results: i=1; mx.microsoft.com 1; spf=pass (sender ip is 40.67.248.234) smtp.rcpttodomain=gcc.gnu.org smtp.mailfrom=arm.com; dmarc=pass (p=none sp=none pct=100) action=none header.from=arm.com; dkim=none (message not signed); arc=none (0) DKIM-Signature: v=1; a=rsa-sha256; c=relaxed/relaxed; d=arm.com; s=selector1; h=From:Date:Subject:Message-ID:Content-Type:MIME-Version:X-MS-Exchange-SenderADCheck; bh=IdOQ2GnDp5k6jtANNBHdn0NqMquYisFOUfHbEX8S4z4=; b=nDl65WIDgs26okfdVLhdusBNnYUGYcVpN166tJFjqnQs4zUPYocof1s/Z8Qj/H5Cn/h70G7KhaSb8ZWULVWcLJZLTUgYMqZa9n45wWEYCiBYVP38fwp1nTG/eyqU9n3rPpxybmjJ4xH1GNf7E73Z+RR5K+wTXy0Uc3/Cm2aCuWw= Received: from DB8PR06CA0066.eurprd06.prod.outlook.com (2603:10a6:10:120::40) by AS8PR08MB6261.eurprd08.prod.outlook.com (2603:10a6:20b:295::18) with Microsoft SMTP Server (version=TLS1_2, cipher=TLS_ECDHE_RSA_WITH_AES_256_GCM_SHA384) id 15.20.8048.26; Fri, 18 Oct 2024 06:22:48 +0000 Received: from DU2PEPF00028D07.eurprd03.prod.outlook.com (2603:10a6:10:120:cafe::68) by DB8PR06CA0066.outlook.office365.com (2603:10a6:10:120::40) with Microsoft SMTP Server (version=TLS1_2, cipher=TLS_ECDHE_RSA_WITH_AES_256_GCM_SHA384) id 15.20.8048.27 via Frontend Transport; Fri, 18 Oct 2024 06:22:48 +0000 X-MS-Exchange-Authentication-Results: spf=pass (sender IP is 40.67.248.234) smtp.mailfrom=arm.com; dkim=none (message not signed) header.d=none;dmarc=pass action=none header.from=arm.com; Received-SPF: Pass (protection.outlook.com: domain of arm.com designates 40.67.248.234 as permitted sender) receiver=protection.outlook.com; client-ip=40.67.248.234; helo=nebula.arm.com; pr=C Received: from nebula.arm.com (40.67.248.234) by DU2PEPF00028D07.mail.protection.outlook.com (10.167.242.167) with Microsoft SMTP Server (version=TLS1_2, cipher=TLS_ECDHE_RSA_WITH_AES_128_GCM_SHA256) id 15.20.8069.17 via Frontend Transport; Fri, 18 Oct 2024 06:22:48 +0000 Received: from AZ-NEU-EX04.Arm.com (10.251.24.32) by AZ-NEU-EX03.Arm.com (10.251.24.31) with Microsoft SMTP Server (version=TLS1_2, cipher=TLS_ECDHE_RSA_WITH_AES_128_GCM_SHA256) id 15.1.2507.39; Fri, 18 Oct 2024 06:22:46 +0000 Received: from a078660.blr.arm.com (10.162.46.14) by mail.arm.com (10.251.24.32) with Microsoft SMTP Server id 15.1.2507.39 via Frontend Transport; Fri, 18 Oct 2024 06:22:45 +0000 From: Tejas Belagod To: CC: Tejas Belagod , , , Andrea Corallo Subject: [PATCH v2 04/12] AArch64: Diagnose OpenMP offloading when SVE types involved. Date: Fri, 18 Oct 2024 11:52:25 +0530 Message-ID: <20241018062233.243950-5-tejas.belagod@arm.com> X-Mailer: git-send-email 2.25.1 In-Reply-To: <20241018062233.243950-1-tejas.belagod@arm.com> References: <20241018062233.243950-1-tejas.belagod@arm.com> MIME-Version: 1.0 X-EOPAttributedMessage: 1 X-MS-TrafficTypeDiagnostic: DU2PEPF00028D07:EE_|AS8PR08MB6261:EE_|DB5PEPF00014B8C:EE_|PA4PR08MB6271:EE_ X-MS-Office365-Filtering-Correlation-Id: cb8c28c0-189e-4c5c-17d0-08dcef3d517c x-checkrecipientrouted: true NoDisclaimer: true X-MS-Exchange-SenderADCheck: 1 X-MS-Exchange-AntiSpam-Relay: 0 X-Microsoft-Antispam-Untrusted: BCL:0; ARA:13230040|376014|36860700013|82310400026|1800799024; X-Microsoft-Antispam-Message-Info-Original: asPoqYqlBR/CAKrqWvWGO+3tSJZHTl2epFVOLu8jDY+7m/6DvcywnatDAOiC/a2PiQ7qRuqDsrXs6PKZW7REHJrHqQHzC2Sq+SOwp2OXU/t1lhujmDzR0PZbf8dYjP6QejeYA33rAl4Zs6E5TCnOE81TsnSFk7RF0UEfA06/YreCvJLL7khIro34mgi3mBw0RevCqf9nlKiVryj/8WU61KLJdVpCnHCeo+I32jRutbsDUrSqYYwskkZDZqECXvm9ZDCx0cdq7vRAhQNB8XmbfiJPIxiyCayrb2QgZ6henyMuPvLqUeHKjtuBOHM93R52x4l66u7OrhmHf2xfbz2Egtk8EePPHwz6aTW9qnvJwaUFnyovUw0KNSbNXlGMhXcsnVV+7JqBcQTfMiIkVGzOyDzRZW0b3yAZjY5Bbqd8+iXYAGfvSy87scud+lfbwoVzcoxjHVuTwjU0hlZSHqtPywnqsCXkHYAaCLsalsP80maX3pP3szzZrW8Z0dJ3v4gHGoj3Wq/nRs9CJVJKPrrSb7cFP4jpJ/XU5V5BC68S+JRVOcrBlqetdbiET9fhjdFmSPgKmPemGOw3EoSrwvg+bdbUzzBoTeRkGm+iayroRm8Zo6PsB7Qs4CRkZTeJljetnglSa0Tq2rtLubkWz6wwXB7qsY2r56L8MhGJ7dGKRDUGj96F4KwQjei7o+TVFNijcme/JPHrsQMnFAEbFxqGIbCOWjAsr5UBrAdQ1wI58bLSswEtjeShEeQegSxiKzqS6iRTOgrt8iyYt/wIBv6aGnXjz/5NLG3bA6tAzcfZtG8ZZihqA9Sl3Aw6fduKddVZ+8X3DPuudS9w16Q2Q6qzcteRKc4s8E+kY5GCTWi48kiE35GohFCWx9Wj81CcxQIs2fgcGvvGMEQb1B+iiCWcdjcVmB4tjcW1dOPynh7pNAN59FFhWKJSQm7PJlr5IfWb4fT382ihBjptc1QiiooN/RcLsy5eLjE1fWz/ZJHm2qWoC1+EIFDM5v2+pVf+C7xiWhTglwWYR6e97II8kxttOLRzAA9zxFYLbBBLgDb+mWBuvBs5BhEQjnrEWyymUzn07J3uB+IsMD8j7mCi1JGI6VrTXcXaP2vLB+Pj+ISSpeKRmDER/MIIl8mUyikkvM7E56N2NJgvkvBzopG91sw6Ah5yMcFbivXhetsXmGPPTEQOC6idkKvxlhh2auePpHr1/5ctVAsDWasoOKrRS8IXOSnogAPN57W84ml26vLljhSu8A/SPLI3hu0KXHCmLjn7O9jNKxNUkoqC3YAmcdsLgq5LrDa9BuLzZMru41a9P/nOPcj5wUSNd7CJeinlN60jA+8bMjZggXlXk4itm0s5wVKX+RF7byoJdQ1zhopWUoY= X-Forefront-Antispam-Report-Untrusted: CIP:40.67.248.234; CTRY:IE; LANG:en; SCL:1; SRV:; IPV:NLI; SFV:NSPM; H:nebula.arm.com; PTR:InfoDomainNonexistent; CAT:NONE; SFS:(13230040)(376014)(36860700013)(82310400026)(1800799024); DIR:OUT; SFP:1101; X-MS-Exchange-Transport-CrossTenantHeadersStamped: AS8PR08MB6261 X-MS-Exchange-SkipListedInternetSender: ip=[2603:10a6:10:120::40]; domain=DB8PR06CA0066.eurprd06.prod.outlook.com X-MS-Exchange-Transport-CrossTenantHeadersStripped: DB5PEPF00014B8C.eurprd02.prod.outlook.com X-MS-PublicTrafficType: Email X-MS-Office365-Filtering-Correlation-Id-Prvs: 38a32cf8-1b21-438c-8d8a-08dcef3d495b X-Microsoft-Antispam: BCL:0; ARA:13230040|1800799024|36860700013|35042699022|376014|82310400026; X-Microsoft-Antispam-Message-Info: y5lagLXccPsbjrDtcHCp3rOuAkwhItnyOjpCSSDQx33C3bSSt/dqygYldKk28oAuQ/WqFX1BpWpDhM9sxq0ogjhoIimm+N1mUTTrzErfiCgE27S6tf5J1/Yfpii0OxhsCvx/3iJ/VLx8rnzefQHqHtFbLDtHxoiSAdhJW1Z6ViA44TuVn8wq6ugZaaoXJ9OCoUb9Aen9exrmmH8LPrLiiQAs5y65ohyGwAbm1yxqzAgw1PSwZB3lmKN7Jce3zhDh/8xuX1ymyraLNpt5r0MiosskHMbrsU4iTAFZkEgmX+8iKyxIE/uLG0Tjidsjb2+cP1bGXHNiAufJOV8814x57ueV7Ic/N5tQGsWda/VEqKGIv7UyFZk+ae/bGn+xSZzr9QPYgmsgTWx+pt/GcgRezJbRkRm5J8Ue9MQPjtqLtSXtNFmpV89UspAQsPW++r34eMb5N/sfUMyppO5lzCRmVyczyYXmbqz6f8/lOinR4J90b4RRcUPBGOMrv7xQYKb05/MjAasnvNlGqUR7fX/7fAJtu7kzS/yj8x2QVFwPecCprCnzgWahIRmYXVAOQACfBH8Mm3Dh9Xk2xykcKXaK7seOKaJuxs0mEwolI4aBfLJwHfQ/CCA1b9k2O90SYV1XFrrriIwMTYA+L4Lt15ZiHrnWzUXYThLtuYgGMwZ0CE0oXY9cUXACQAkFstdV4EShgUO0Lc/IorJLuJrzKlPuahbVoJEYi7pn0M/+Djpn+/PHI/2PcGltWaMCVJrMkqxpXeQi86E4h5bPwV9VhcIg8XrEVMfVunBA40gOg/XK9rMLsfa+EuUC2ToBXjSuuxXzyU0qyhTCU6UPwY5o8GVlj+IhEmEp7U6PG8NOLpRzTr9mHXBEqsWYtOEgUGCr72BFsVsErNOqE5FbPQrtvkyynnjdFDTLkEC91N/kt7g/qj/BImQegfdnil/C6ckidEY0oFwnUSZHXxTXAHLYEYdwUjDjkUwh+Udn7eJ0RFFT5xdAzxu6X/B8JUOuXn3RnDEv0yGp9wAqHrXelrByJOWsnatbV2/PMpgQhvHSP21kBhfkm/86K38vecIzIoNvzgqaA1IJYXZRturaDvMNh5/c52IwNuiSofCX65zMrazBatTOl0Bmsn3eBruj7Z6Zu5fNnkmNcUrBT2G0pr6eCsBf8nWAdDKxFGzaU+jHy6mTLklYxJZyMG8ayI7sd4rPjIMrgYNxApV2qB/yuCg69cTJZQKWtol9R/ll5qNURUDbNDfNK8pmTbUCgdab28vaNnJXjElb6d7K72YLekceRWUPf7V3BFIDZD3wcLg+aso02rgf+c+zkuTj6MB/aHr/y7UcUF2zfJAA1nQ+4Jt6DLCE5ZFrgOeUI4hAcFO+kaWxww0= X-Forefront-Antispam-Report: CIP:63.35.35.123; CTRY:IE; LANG:en; SCL:1; SRV:; IPV:CAL; SFV:NSPM; H:64aa7808-outbound-1.mta.getcheckrecipient.com; PTR:ec2-63-35-35-123.eu-west-1.compute.amazonaws.com; CAT:NONE; SFS:(13230040)(1800799024)(36860700013)(35042699022)(376014)(82310400026); DIR:OUT; SFP:1101; X-OriginatorOrg: arm.com X-MS-Exchange-CrossTenant-OriginalArrivalTime: 18 Oct 2024 06:23:01.8898 (UTC) X-MS-Exchange-CrossTenant-Network-Message-Id: cb8c28c0-189e-4c5c-17d0-08dcef3d517c X-MS-Exchange-CrossTenant-Id: f34e5979-57d9-4aaa-ad4d-b122a662184d X-MS-Exchange-CrossTenant-OriginalAttributedTenantConnectingIp: TenantId=f34e5979-57d9-4aaa-ad4d-b122a662184d; Ip=[63.35.35.123]; Helo=[64aa7808-outbound-1.mta.getcheckrecipient.com] X-MS-Exchange-CrossTenant-AuthSource: DB5PEPF00014B8C.eurprd02.prod.outlook.com X-MS-Exchange-CrossTenant-AuthAs: Anonymous X-MS-Exchange-CrossTenant-FromEntityHeader: HybridOnPrem X-MS-Exchange-Transport-CrossTenantHeadersStamped: PA4PR08MB6271 X-Spam-Status: No, score=-10.2 required=5.0 tests=BAYES_00, DKIM_SIGNED, DKIM_VALID, DKIM_VALID_AU, DKIM_VALID_EF, FORGED_SPF_HELO, GIT_PATCH_0, KAM_SHORT, RCVD_IN_DNSWL_NONE, SCC_10_SHORT_WORD_LINES, SCC_5_SHORT_WORD_LINES, SPF_HELO_PASS, SPF_NONE, TXREP, UNPARSEABLE_RELAY autolearn=ham autolearn_force=no version=3.4.6 X-Spam-Checker-Version: SpamAssassin 3.4.6 (2021-04-09) on server2.sourceware.org X-BeenThere: gcc-patches@gcc.gnu.org X-Mailman-Version: 2.1.30 Precedence: list List-Id: Gcc-patches mailing list List-Unsubscribe: , List-Archive: List-Post: List-Help: List-Subscribe: , Errors-To: gcc-patches-bounces~incoming=patchwork.ozlabs.org@gcc.gnu.org The target clause in OpenMP is used to offload loop kernels to accelarator peripeherals. target's 'map' clause is used to move data from and to the accelarator. When the data is SVE type, it may not be suitable because of various reasons i.e. the two SVE targets may not agree on vector size or some targets don't support variable vector size. This makes SVE unsuitable for use in OMP's 'map' clause. This patch diagnoses all such cases and issues an error where SVE types are not suitable. Co-authored-by: Andrea Corallo gcc/ChangeLog: * target.h (type_context_kind): Add new context kinds for target clauses. * config/aarch64/aarch64-sve-builtins.cc (verify_type_context): Diagnose SVE types for a given OpenMP context. (omp_type_context): New. * gimplify.cc (omp_notice_variable): Diagnose implicitly-mapped SVE objects in OpenMP regions. (gimplify_scan_omp_clauses): Diagnose SVE types for various target clauses. gcc/testsuite/ChangeLog: * gcc.target/aarch64/sve/omp/offload.c: New test. * gcc.target/aarch64/sve/omp/offload-parallel-loop.c: Likewise. * gcc.target/aarch64/sve/omp/offload-parallel.c: Likewise. * gcc.target/aarch64/sve/omp/offload-simd.c: Likewise. * gcc.target/aarch64/sve/omp/offload-teams-distribute-simd.c: Likewise. * gcc.target/aarch64/sve/omp/offload-teams-distribute.c: Likewise. * gcc.target/aarch64/sve/omp/offload-teams-loop.c: Likewise. * gcc.target/aarch64/sve/omp/offload-teams.c: Likewise. * gcc.target/aarch64/sve/omp/target-device.c: Likewise. * gcc.target/aarch64/sve/omp/target-link.c: Likewise. --- gcc/config/aarch64/aarch64-sve-builtins.cc | 52 +- gcc/gimplify.cc | 34 +- gcc/target.h | 19 +- .../aarch64/sve/omp/offload-parallel-loop.c | 442 +++++++++++++++++ .../aarch64/sve/omp/offload-parallel.c | 376 +++++++++++++++ .../gcc.target/aarch64/sve/omp/offload-simd.c | 442 +++++++++++++++++ .../sve/omp/offload-teams-distribute-simd.c | 442 +++++++++++++++++ .../sve/omp/offload-teams-distribute.c | 442 +++++++++++++++++ .../aarch64/sve/omp/offload-teams-loop.c | 442 +++++++++++++++++ .../aarch64/sve/omp/offload-teams.c | 365 ++++++++++++++ .../gcc.target/aarch64/sve/omp/offload.c | 452 ++++++++++++++++++ .../aarch64/sve/omp/target-device.c | 186 +++++++ .../gcc.target/aarch64/sve/omp/target-link.c | 54 +++ 13 files changed, 3745 insertions(+), 3 deletions(-) create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel-loop.c create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel.c create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/offload-simd.c create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute-simd.c create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute.c create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-loop.c create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams.c create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/offload.c create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/target-device.c create mode 100644 gcc/testsuite/gcc.target/aarch64/sve/omp/target-link.c diff --git a/gcc/config/aarch64/aarch64-sve-builtins.cc b/gcc/config/aarch64/aarch64-sve-builtins.cc index e7c703c987e..2c169ea3806 100644 --- a/gcc/config/aarch64/aarch64-sve-builtins.cc +++ b/gcc/config/aarch64/aarch64-sve-builtins.cc @@ -4956,12 +4956,35 @@ handle_arm_sve_vector_bits_attribute (tree *node, tree, tree args, int, return NULL_TREE; } + +/* Return true if OpenMP context types. */ + +static bool +omp_type_context (type_context_kind context) +{ + switch (context) + { + case TCTX_OMP_MAP: + case TCTX_OMP_MAP_IMP_REF: + case TCTX_OMP_PRIVATE: + case TCTX_OMP_FIRSTPRIVATE: + case TCTX_OMP_DEVICE_ADDR: + return true; + default: + return false;; + } +} + /* Implement TARGET_VERIFY_TYPE_CONTEXT for SVE types. */ bool verify_type_context (location_t loc, type_context_kind context, const_tree type, bool silent_p) { - if (!sizeless_type_p (type)) + const_tree tmp = type; + if (omp_type_context (context) && POINTER_TYPE_P (type)) + tmp = strip_pointer_types (tmp); + + if (!sizeless_type_p (tmp)) return true; switch (context) @@ -5021,6 +5044,33 @@ verify_type_context (location_t loc, type_context_kind context, if (!silent_p) error_at (loc, "capture by copy of SVE type %qT", type); return false; + + case TCTX_OMP_MAP: + if (!silent_p) + error_at (loc, "SVE type %qT not allowed in map clause", type); + return false; + + case TCTX_OMP_MAP_IMP_REF: + /* The diagnosis is done in the caller. */ + return false; + + case TCTX_OMP_PRIVATE: + if (!silent_p) + error_at (loc, "SVE type %qT not allowed in target private clause", type); + return false; + + case TCTX_OMP_FIRSTPRIVATE: + if (!silent_p) + error_at (loc, "SVE type %qT not allowed in target firstprivate clause", type); + return false; + + case TCTX_OMP_DEVICE_ADDR: + if (!silent_p) + error_at (loc, "SVE type %qT not allowed in target device clauses", type); + return false; + + default: + break; } gcc_unreachable (); } diff --git a/gcc/gimplify.cc b/gcc/gimplify.cc index 3f602469d57..ace43cf78a0 100644 --- a/gcc/gimplify.cc +++ b/gcc/gimplify.cc @@ -8430,11 +8430,13 @@ omp_notice_variable (struct gimplify_omp_ctx *ctx, tree decl, bool in_code) | GOVD_MAP_ALLOC_ONLY)) == flags) { tree type = TREE_TYPE (decl); + location_t dummy = UNKNOWN_LOCATION; if (gimplify_omp_ctxp->target_firstprivatize_array_bases && omp_privatize_by_reference (decl)) type = TREE_TYPE (type); - if (!omp_mappable_type (type)) + if (!omp_mappable_type (type) + || !verify_type_context (dummy, TCTX_OMP_MAP_IMP_REF, type)) { error ("%qD referenced in target region does not have " "a mappable type", decl); @@ -12165,6 +12167,8 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p, unsigned int flags; tree decl; auto_vec addr_tokens; + tree op = NULL_TREE; + location_t loc = OMP_CLAUSE_LOCATION (c); if (grp_end && c == OMP_CLAUSE_CHAIN (grp_end)) { @@ -12172,6 +12176,34 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p, grp_end = NULL_TREE; } + if (code == OMP_TARGET || code == OMP_TARGET_DATA + || code == OMP_TARGET_ENTER_DATA || code == OMP_TARGET_EXIT_DATA) + /* Do some target-specific type checks for map operands. */ + switch (OMP_CLAUSE_CODE (c)) + { + case OMP_CLAUSE_MAP: + op = OMP_CLAUSE_OPERAND (c, 0); + verify_type_context (loc, TCTX_OMP_MAP, TREE_TYPE (op)); + break; + case OMP_CLAUSE_PRIVATE: + op = OMP_CLAUSE_OPERAND (c, 0); + verify_type_context (loc, TCTX_OMP_PRIVATE, TREE_TYPE (op)); + break; + case OMP_CLAUSE_FIRSTPRIVATE: + op = OMP_CLAUSE_OPERAND (c, 0); + verify_type_context (loc, TCTX_OMP_FIRSTPRIVATE, TREE_TYPE (op)); + break; + case OMP_CLAUSE_IS_DEVICE_PTR: + case OMP_CLAUSE_USE_DEVICE_ADDR: + case OMP_CLAUSE_USE_DEVICE_PTR: + case OMP_CLAUSE_HAS_DEVICE_ADDR: + op = OMP_CLAUSE_OPERAND (c, 0); + verify_type_context (loc, TCTX_OMP_DEVICE_ADDR, TREE_TYPE (op)); + break; + default: + break; + } + switch (OMP_CLAUSE_CODE (c)) { case OMP_CLAUSE_PRIVATE: diff --git a/gcc/target.h b/gcc/target.h index 837651d273a..7791daf6315 100644 --- a/gcc/target.h +++ b/gcc/target.h @@ -271,7 +271,24 @@ enum type_context_kind { TCTX_EXCEPTIONS, /* Capturing objects of type T by value in a closure. */ - TCTX_CAPTURE_BY_COPY + TCTX_CAPTURE_BY_COPY, + + /* Objects of type T appearing in OpenMP map clause. */ + TCTX_OMP_MAP, + + /* Objects of type T appearing in OpenMP target region + without explicit map. */ + TCTX_OMP_MAP_IMP_REF, + + /* Objects of type T appearing in OpenMP private clause. */ + TCTX_OMP_PRIVATE, + + /* Objects of type T appearing in OpenMP firstprivate clause. */ + TCTX_OMP_FIRSTPRIVATE, + + /* Objects of type T appearing in OpenMP device clauses. */ + TCTX_OMP_DEVICE_ADDR + }; enum poly_value_estimate_kind diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel-loop.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel-loop.c new file mode 100644 index 00000000000..b8e078fc816 --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel-loop.c @@ -0,0 +1,442 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define N __ARM_FEATURE_SVE_BITS + +svint32_t +omp_target_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel loop + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + + return va; +} + +svint32_t +omp_target_data_map_1_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel loop map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +svint32_t +omp_target_data_map_2_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel loop map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target parallel loop map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +svint32_t +omp_target_map_data_enter_exit_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target parallel loop map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target parallel loop map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + + return va; +} + +svint32_t +omp_target_map_data_alloc_update_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ +{ +#pragma omp target parallel loop + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target parallel loop map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } +} + return va; +} + +int64_t +omp_target_private_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +/* Combined construct scenario: here private applies to the parallel loop + construct, so no error. */ +#pragma omp target parallel loop private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vla (svbool_t vp) +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel loop firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N))) +typedef svint32_t v8si FIXED_ATTR; +typedef svbool_t v8bi FIXED_ATTR; + +v8si +omp_target_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel loop + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_1_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel loop map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_2_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel loop map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target parallel loop map(to: b, c) map(tofrom: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +v8si +omp_target_map_data_enter_exit_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target parallel loop map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target parallel loop map(to: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) + + return va; +} + +v8si +omp_target_map_data_alloc_update_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) +{ +#pragma omp target parallel loop + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target parallel loop map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } +} + return va; +} + +int64_t +omp_target_private_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel loop private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vls (v8bi vp) +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel loop firstprivate (vp) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); + vc = svld1_s32 (vp, c); + va = svadd_s32_z (vp, vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel.c new file mode 100644 index 00000000000..b8edaff6755 --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel.c @@ -0,0 +1,376 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define parallel parallel +#define N __ARM_FEATURE_SVE_BITS + +svint32_t +omp_target_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + } + + return va; +} + +svint32_t +omp_target_data_map_1_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + + return va; +} + +svint32_t +omp_target_data_map_2_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + + +#pragma omp target parallel map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +#pragma omp target parallel map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + + return va; +} + +svint32_t +omp_target_map_data_enter_exit_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target parallel map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +#pragma omp target parallel map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + +#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + + return va; +} + +svint32_t +omp_target_map_data_alloc_update_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ +{ +#pragma omp target parallel + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target parallel map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } +} + return va; +} + +#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N))) +typedef svint32_t v8si FIXED_ATTR; + +v8si +omp_target_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + + return va; +} + +v8si +omp_target_data_map_1_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target parallel map(to: b, c) map(from: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + + return va; +} + +v8si +omp_target_data_map_2_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + + +#pragma omp target parallel map(to: b, c) map(from: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +#pragma omp target parallel map(to: b, c) map(tofrom: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + + return va; +} + +v8si +omp_target_map_data_enter_exit_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target parallel map(from: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +#pragma omp target parallel map(to: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + +#pragma omp target exit data map(from: va) + + return va; +} + +v8si +omp_target_map_data_alloc_update_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) +{ +#pragma omp target parallel + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target parallel map(from: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } +} + return va; +} diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-simd.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-simd.c new file mode 100644 index 00000000000..a09aa5399f4 --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-simd.c @@ -0,0 +1,442 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define N __ARM_FEATURE_SVE_BITS + +svint32_t +omp_target_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target simd + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + + return va; +} + +svint32_t +omp_target_data_map_1_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target simd map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +svint32_t +omp_target_data_map_2_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target simd map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target simd map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +svint32_t +omp_target_map_data_enter_exit_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target simd map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target simd map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + + return va; +} + +svint32_t +omp_target_map_data_alloc_update_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ +{ +#pragma omp target simd + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target simd map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } +} + return va; +} + +int64_t +omp_target_private_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +/* Combined construct scenario: here private applies to the simd construct so + no error. */ +#pragma omp target simd private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vla (svbool_t vp) +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target simd firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N))) +typedef svint32_t v8si FIXED_ATTR; +typedef svbool_t v8bi FIXED_ATTR; + +v8si +omp_target_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target simd + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_1_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target simd map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_2_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target simd map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target simd map(to: b, c) map(tofrom: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +v8si +omp_target_map_data_enter_exit_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target simd map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target simd map(to: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) + + return va; +} + +v8si +omp_target_map_data_alloc_update_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) +{ +#pragma omp target simd + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target simd map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } +} + return va; +} + +int64_t +omp_target_private_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target simd private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vls (v8bi vp) +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target simd firstprivate (vp) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); + vc = svld1_s32 (vp, c); + va = svadd_s32_z (vp, vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute-simd.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute-simd.c new file mode 100644 index 00000000000..3a998caeefd --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute-simd.c @@ -0,0 +1,442 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define N __ARM_FEATURE_SVE_BITS + +svint32_t +omp_target_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute simd + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + + return va; +} + +svint32_t +omp_target_data_map_1_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute simd map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +svint32_t +omp_target_data_map_2_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute simd map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams distribute simd map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +svint32_t +omp_target_map_data_enter_exit_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target teams distribute simd map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams distribute simd map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + + return va; +} + +svint32_t +omp_target_map_data_alloc_update_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { +#pragma omp target teams distribute simd + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target teams distribute simd map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + return va; +} + +int64_t +omp_target_private_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +/* Combined construct scenario: here private applies to the distribute simd + construct, so no error. */ +#pragma omp target teams distribute simd private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vla (svbool_t vp) +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute simd firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N))) +typedef svint32_t v8si FIXED_ATTR; +typedef svbool_t v8bi FIXED_ATTR; + +v8si +omp_target_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute simd + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_1_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute simd map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_2_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute simd map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams distribute simd map(to: b, c) map(tofrom: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +v8si +omp_target_map_data_enter_exit_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target teams distribute simd map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams distribute simd map(to: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) + + return va; +} + +v8si +omp_target_map_data_alloc_update_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) + { +#pragma omp target teams distribute simd + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target teams distribute simd map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + return va; +} + +int64_t +omp_target_private_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute simd private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vls (v8bi vp) +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute simd firstprivate (vp) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); + vc = svld1_s32 (vp, c); + va = svadd_s32_z (vp, vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute.c new file mode 100644 index 00000000000..dfb78ef69ee --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute.c @@ -0,0 +1,442 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define N __ARM_FEATURE_SVE_BITS + +svint32_t +omp_target_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + return va; +} + +svint32_t +omp_target_data_map_1_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +svint32_t +omp_target_data_map_2_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams distribute map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +svint32_t +omp_target_map_data_enter_exit_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target teams distribute map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams distribute map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + + return va; +} + +svint32_t +omp_target_map_data_alloc_update_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { +#pragma omp target teams distribute + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target teams distribute map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + + return va; +} + +int64_t +omp_target_private_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +/* Combined construct scenario: here private applies to the teams distribute + construct, so no error. */ +#pragma omp target teams distribute private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate (svbool_t vp) +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N))) +typedef svint32_t v8si FIXED_ATTR; +typedef svbool_t v8bi FIXED_ATTR; + +v8si +omp_target_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + return va; +} + +v8si +omp_target_data_map_1_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_2_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams distribute map(to: b, c) map(tofrom: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +v8si +omp_target_map_data_enter_exit_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target teams distribute map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams distribute map(to: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) + + return va; +} + +v8si +omp_target_map_data_alloc_update_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) + { +#pragma omp target teams distribute + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target teams distribute map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + + return va; +} + +int64_t +omp_target_private_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vls (v8bi vp) +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams distribute firstprivate (vp) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); + vc = svld1_s32 (vp, c); + va = svadd_s32_z (vp, vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-loop.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-loop.c new file mode 100644 index 00000000000..4c96f5a0fc8 --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-loop.c @@ -0,0 +1,442 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define N __ARM_FEATURE_SVE_BITS + +svint32_t +omp_target_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams loop + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + + return va; +} + +svint32_t +omp_target_data_map_1_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams loop map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +svint32_t +omp_target_data_map_2_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams loop map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams loop map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +svint32_t +omp_target_map_data_enter_exit_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target teams loop map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams loop map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + + return va; +} + +svint32_t +omp_target_map_data_alloc_update_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ +{ +#pragma omp target teams loop + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target teams loop map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } +} + return va; +} + +int64_t +omp_target_private_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +/* Combined construct scenario: here private applies to the teams loop + construct, so no error. */ +#pragma omp target teams loop private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vla (svbool_t vp) +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams loop firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N))) +typedef svint32_t v8si FIXED_ATTR; +typedef svbool_t v8bi FIXED_ATTR; + +v8si +omp_target_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams loop + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_1_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams loop map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_2_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams loop map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams loop map(to: b, c) map(tofrom: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +v8si +omp_target_map_data_enter_exit_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target teams loop map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target teams loop map(to: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) + + return va; +} + +v8si +omp_target_map_data_alloc_update_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) +{ +#pragma omp target teams loop + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target teams loop map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } +} + return va; +} + +int64_t +omp_target_private_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams loop private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vls (v8bi vp) +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams loop firstprivate (vp) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); + vc = svld1_s32 (vp, c); + va = svadd_s32_z (vp, vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams.c new file mode 100644 index 00000000000..2c5bf7e8926 --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams.c @@ -0,0 +1,365 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define N __ARM_FEATURE_SVE_BITS + +svint32_t +omp_target_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + } + + return va; +} + +svint32_t +omp_target_data_map_1_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + + return va; +} + +svint32_t +omp_target_data_map_2_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + + +#pragma omp target teams map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +#pragma omp target teams map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + + return va; +} + +svint32_t +omp_target_map_data_enter_exit_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target teams map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +#pragma omp target teams map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + +#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + + return va; +} + +svint32_t +omp_target_map_data_alloc_update_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ +{ +#pragma omp target teams + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target teams map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } +} + return va; +} + +#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N))) +typedef svint32_t v8si FIXED_ATTR; + +v8si +omp_target_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + + return va; +} + +v8si +omp_target_data_map_1_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target teams map(to: b, c) map(from: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + + return va; +} + +v8si +omp_target_data_map_2_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + + +#pragma omp target teams map(to: b, c) map(from: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +#pragma omp target teams map(to: b, c) map(tofrom: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + + return va; +} + +v8si +omp_target_map_data_enter_exit_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target teams map(from: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +#pragma omp target teams map(to: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } + +#pragma omp target exit data map(from: va) + + return va; +} + +v8si +omp_target_map_data_alloc_update_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) +{ +#pragma omp target teams + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target teams map(from: va) + { + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + } +} + return va; +} diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload.c new file mode 100644 index 00000000000..b2f6e543531 --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload.c @@ -0,0 +1,452 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define N __ARM_FEATURE_SVE_BITS +#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N))) + +svint32_t +omp_target_vla () +{ + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + + return va; +} + +svint32_t +omp_target_data_map_1_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +svint32_t +omp_target_data_map_2_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +svint32_t +omp_target_map_data_enter_exit_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + + return va; +} + +svint32_t +omp_target_map_data_alloc_update_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ +{ +#pragma omp target + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } +} + return va; +} + +int64_t +omp_target_private_vla () +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target private (va, vb, vc) map (to: b, c) map (from: res) /* { dg-error {SVE type 'svint32_t' not allowed in target private clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vla (svbool_t vp) +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +typedef svint32_t v8si FIXED_ATTR; +typedef svbool_t v8bi FIXED_ATTR; + +v8si +omp_target_vls () +{ + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_1_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + + return va; +} + +v8si +omp_target_data_map_2_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target map(to: b, c) map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target map(to: b, c) map(tofrom: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + + return va; +} + +v8si +omp_target_map_data_enter_exit_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target enter data map(to: b, c) + +#pragma omp target map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +#pragma omp target map(to: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } + +#pragma omp target exit data map(from: va) + + return va; +} + +v8si +omp_target_map_data_alloc_update_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data map(to: b, c) map(alloc: va) +{ +#pragma omp target + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + } + +/* Update va on the host from target. */ +#pragma omp target update from(va) + +#pragma omp target map(from: va) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, va); + va = svadd_s32_z (svptrue_b32 (), vc, va); + } +} + return va; +} + +int64_t +omp_target_private_vls () +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target private (va, vb, vc) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (svptrue_b32 (), b); + vc = svld1_s32 (svptrue_b32 (), c); + va = svadd_s32_z (svptrue_b32 (), vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t +omp_target_firstprivate_vls (v8bi vp) +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target firstprivate (vp) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); + vc = svld1_s32 (vp, c); + va = svadd_s32_z (vp, vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/target-device.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/target-device.c new file mode 100644 index 00000000000..a20129cb42b --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/target-device.c @@ -0,0 +1,186 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define N __ARM_FEATURE_SVE_BITS +#define FIXED_ATTR __attribute__ ((arm_sve_vector_bits (N))) + +typedef __SVInt32_t v8si FIXED_ATTR; +typedef svbool_t v8bi FIXED_ATTR; + +int64_t __attribute__ ((noinline)) +omp_target_device_ptr_vls (v8bi vp, v8si *vptr) +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data use_device_ptr (vptr) map (to: b, c) +#pragma omp target is_device_ptr (vptr) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = *vptr; + vc = svld1_s32 (vp, c); + va = svadd_s32_z (vp, vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t __attribute__ ((noinline)) +omp_target_device_addr_vls (v8bi vp, v8si *vptr) +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data use_device_addr (vb) map (to: b, c) +#pragma omp target is_device_ptr (vptr) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = *vptr; + vc = svld1_s32 (vp, c); + va = svadd_s32_z (vp, vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t __attribute__ ((noinline)) +omp_target_has_device_addr_vls (v8bi vp, v8si *vptr) +{ + + int a[N], b[N], c[N]; + v8si va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data use_device_addr (vb) map (to: b, c) +#pragma omp target has_device_addr (vb) map (to: b, c) map (from: res) + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); + vc = svld1_s32 (vp, c); + va = svadd_s32_z (vp, vb, vc); + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t __attribute__ ((noinline)) +omp_target_device_ptr_vla (svbool_t vp, svint32_t *vptr) +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data use_device_ptr (vptr) map (to: b, c) /* { dg-error {SVE type 'svint32_t \*' not allowed in target device clauses} } */ +#pragma omp target is_device_ptr (vptr) map (to: b, c) map (from: res) /* { dg-error {SVE type 'svint32_t \*' not allowed in target device clauses} } */ + for (i = 0; i < 8; i++) + { + vb = *vptr; /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + /* { dg-error {'vp' referenced in target region does not have a mappable type} "" { target *-*-* } .-1 } */ + va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t __attribute__ ((noinline)) +omp_target_device_addr_vla (svbool_t vp, svint32_t *vptr) +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data use_device_addr (vb) map (to: b, c) /* { dg-error {SVE type 'svint32_t' not allowed in target device clauses} } */ +#pragma omp target is_device_ptr (vptr) map (to: b, c) map (from: res) /* { dg-error {SVE type 'svint32_t \*' not allowed in target device clauses} } */ + for (i = 0; i < 8; i++) + { + vb = *vptr; /* { dg-error {'vb' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + /* { dg-error {'vp' referenced in target region does not have a mappable type} "" { target *-*-* } .-1 } */ + va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} + +int64_t __attribute__ ((noinline)) +omp_target_has_device_addr_vla (svbool_t vp, svint32_t *vptr) +{ + + int a[N], b[N], c[N]; + svint32_t va, vb, vc; + int64_t res; + int i; + +#pragma omp parallel for + for (i = 0; i < N; i++) + { + b[i] = i; + c[i] = i + 1; + } + +#pragma omp target data use_device_addr (vb) map (to: b, c) /* { dg-error {SVE type 'svint32_t' not allowed in target device clauses} } */ +#pragma omp target has_device_addr (vb) map (to: b, c) map (from: res) /* { dg-error {SVE type 'svint32_t' not allowed in target device clauses} } */ + for (i = 0; i < 8; i++) + { + vb = svld1_s32 (vp, b); /* { dg-error {'vp' referenced in target region does not have a mappable type} } */ + vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */ + va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */ + res = svaddv_s32 (svptrue_b32 (), va); + } + + return res; +} diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/target-link.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/target-link.c new file mode 100644 index 00000000000..afd9cf4fb05 --- /dev/null +++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/target-link.c @@ -0,0 +1,54 @@ +/* { dg-do compile } */ +/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */ + +#include + +#define N __ARM_FEATURE_SVE_BITS +#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N))) + +typedef __SVInt32_t v8si FIXED_ATTR; + +static v8si local_vec; +#pragma omp declare target link(local_vec) + +v8si global_vec; +#pragma omp declare target link(global_vec) + +static svint32_t slocal_vec; /* { dg-error {SVE type 'svint32_t' does not have a fixed size} } */ +#pragma omp declare target link(slocal_vec) /* { dg-error {'slocal_vec' does not have a mappable type in 'link' clause} } */ + +void +one_get_inc2_local_vec_vls () +{ + v8si res, res2, tmp; + +#pragma omp target map(from: res, res2) + { + res = local_vec; + local_vec = svadd_s32_z (svptrue_b32 (), local_vec, local_vec); + res2 = local_vec; + } + + tmp = svadd_s32_z (svptrue_b32 (), res, res); + svbool_t p = svcmpne_s32 (svptrue_b32 (), tmp, res2); + if (svptest_any (svptrue_b32 (), p)) + __builtin_abort (); +} + +void +one_get_inc3_global_vec_vls () +{ + v8si res, res2, tmp; + +#pragma omp target map(from: res, res2) + { + res = global_vec; + global_vec = svadd_s32_z (svptrue_b32 (), global_vec, global_vec); + res2 = global_vec; + } + + tmp = svadd_s32_z (svptrue_b32 (), res, res); + svbool_t p = svcmpne_s32 (svptrue_b32 (), tmp, res2); + if (svptest_any (svptrue_b32 (), p)) + __builtin_abort (); +}