Add VLM/VSM tests
This set of tests cover the allowed usage reachable from intrinsics.
These only pass after change 74304.
Change-Id: I6fbd9ad14e86aeb4827dd4dc57323602b22a815a
diff --git a/tests/cocotb/BUILD b/tests/cocotb/BUILD
index 369baeb..0b030d8 100644
--- a/tests/cocotb/BUILD
+++ b/tests/cocotb/BUILD
@@ -172,6 +172,7 @@
# BEGIN_TESTCASES_FOR_rvv_load_store_test
RVV_LOAD_STORE_TESTCASES = [
+ "load_store_bits",
"load8_stride2_m1",
"load8_stride2_m1_partial",
"load8_stride2_mf4",
diff --git a/tests/cocotb/rvv/load_store/BUILD b/tests/cocotb/rvv/load_store/BUILD
index 167cd30..31e659f 100644
--- a/tests/cocotb/rvv/load_store/BUILD
+++ b/tests/cocotb/rvv/load_store/BUILD
@@ -20,6 +20,9 @@
template_rule(
kelvin_v2_binary,
{
+ "load_store_bits": {
+ "srcs": ["load_store_bits.cc"],
+ },
"load8_indexed_m1": {
"srcs": ["load8_indexed_m1.cc"],
},
@@ -92,6 +95,7 @@
filegroup(
name = "rvv_load_store_tests",
srcs = [
+ ":load_store_bits.elf",
":load8_indexed_m1.elf",
":load8_segment2_unit_m1.elf",
":load16_segment2_unit_m1.elf",
diff --git a/tests/cocotb/rvv/load_store/load_store_bits.cc b/tests/cocotb/rvv/load_store/load_store_bits.cc
new file mode 100644
index 0000000..f4a73a7
--- /dev/null
+++ b/tests/cocotb/rvv/load_store/load_store_bits.cc
@@ -0,0 +1,63 @@
+// Copyright 2025 Google LLC
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+// http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+#include <riscv_vector.h>
+#include <stdint.h>
+
+namespace {
+constexpr size_t buf_size = 16;
+}
+
+size_t vl __attribute__((section(".data"))) = buf_size;
+uint8_t in_buf[buf_size] __attribute__((section(".data")));
+uint8_t out_buf[buf_size] __attribute__((section(".data")));
+
+extern "C" {
+__attribute__((used, retain)) void vlm_vsm_v_b1() {
+ auto data = __riscv_vlm_v_b1(in_buf, vl);
+ __riscv_vsm_v_b1(out_buf, data, vl);
+}
+
+__attribute__((used, retain)) void vlm_vsm_v_b2() {
+ auto data = __riscv_vlm_v_b2(in_buf, vl);
+ __riscv_vsm_v_b2(out_buf, data, vl);
+}
+
+__attribute__((used, retain)) void vlm_vsm_v_b4() {
+ auto data = __riscv_vlm_v_b4(in_buf, vl);
+ __riscv_vsm_v_b4(out_buf, data, vl);
+}
+
+__attribute__((used, retain)) void vlm_vsm_v_b8() {
+ auto data = __riscv_vlm_v_b8(in_buf, vl);
+ __riscv_vsm_v_b8(out_buf, data, vl);
+}
+
+__attribute__((used, retain)) void vlm_vsm_v_b16() {
+ auto data = __riscv_vlm_v_b16(in_buf, vl);
+ __riscv_vsm_v_b16(out_buf, data, vl);
+}
+
+__attribute__((used, retain)) void vlm_vsm_v_b32() {
+ auto data = __riscv_vlm_v_b32(in_buf, vl);
+ __riscv_vsm_v_b32(out_buf, data, vl);
+}
+}
+
+void (*impl)() __attribute__((section(".data"))) = &vlm_vsm_v_b1;
+
+int main(int argc, char** argv) {
+ impl();
+ return 0;
+}
diff --git a/tests/cocotb/rvv_load_store_test.py b/tests/cocotb/rvv_load_store_test.py
index 78da0b4..2f3f11a 100644
--- a/tests/cocotb/rvv_load_store_test.py
+++ b/tests/cocotb/rvv_load_store_test.py
@@ -145,6 +145,66 @@
assert (actual_outputs == expected_outputs).all()
+
+@cocotb.test()
+async def load_store_bits(dut):
+ """Test vlm/vsm usage accessible from intrinsics."""
+ # mask is not accessible from here.
+ fixture = await Fixture.Create(dut)
+ r = runfiles.Create()
+ cases = [
+ {'impl': 'vlm_vsm_v_b1', 'vl': 128},
+ {'impl': 'vlm_vsm_v_b1', 'vl': 121},
+ {'impl': 'vlm_vsm_v_b1', 'vl': 120},
+ {'impl': 'vlm_vsm_v_b2', 'vl': 64},
+ {'impl': 'vlm_vsm_v_b2', 'vl': 57},
+ {'impl': 'vlm_vsm_v_b2', 'vl': 56},
+ {'impl': 'vlm_vsm_v_b4', 'vl': 32},
+ {'impl': 'vlm_vsm_v_b4', 'vl': 25},
+ {'impl': 'vlm_vsm_v_b4', 'vl': 24},
+ {'impl': 'vlm_vsm_v_b8', 'vl': 16},
+ {'impl': 'vlm_vsm_v_b8', 'vl': 9},
+ {'impl': 'vlm_vsm_v_b8', 'vl': 8},
+ {'impl': 'vlm_vsm_v_b16', 'vl': 8},
+ {'impl': 'vlm_vsm_v_b16', 'vl': 1},
+ {'impl': 'vlm_vsm_v_b32', 'vl': 4},
+ {'impl': 'vlm_vsm_v_b32', 'vl': 1},
+ ]
+ await fixture.load_elf_and_lookup_symbols(
+ r.Rlocation(
+ 'kelvin_hw/tests/cocotb/rvv/load_store/load_store_bits.elf'),
+ ['vl', 'in_buf', 'out_buf', 'impl'] + [c['impl'] for c in cases],
+ )
+ rng = np.random.default_rng()
+ for c in cases:
+ impl = c['impl']
+ vl = c['vl']
+ in_bytes = (vl + 7) // 8
+ last_byte_mask = (1 << (vl % 8) - 1) if vl % 8 else 0xFF
+
+ input_data = rng.integers(
+ low=0, high=256, size=in_bytes, dtype=np.uint8)
+ expected_output = input_data
+ expected_output[-1] = expected_output[-1] & last_byte_mask
+
+ await fixture.write_ptr('impl', impl)
+ await fixture.write_word('vl', vl)
+ await fixture.write('in_buf', input_data)
+ await fixture.write('out_buf', np.zeros([in_bytes], dtype=np.uint8))
+
+ await fixture.run_to_halt()
+
+ actual_output = (await fixture.read('out_buf', in_bytes)).view(np.uint8)
+
+ debug_msg = str({
+ 'impl': impl,
+ 'input': input_data,
+ 'expected': expected_output,
+ 'actual': actual_output,
+ })
+ assert (actual_output == expected_output).all(), debug_msg
+
+
@cocotb.test()
async def load8_stride2_m1(dut):
await vector_load_store(
