Fix/fnptr declarators

src/fe/parser.c

@@ -318,6 +318,8 @@ static uint32_t parse_expr(parser_t *P, int min_prec);
 static uint32_t parse_stmt(parser_t *P);
 static uint32_t parse_type_spec(parser_t *P, uint16_t *quals, uint16_t *cuda);
 static uint32_t parse_decl_or_stmt(parser_t *P);
+static uint32_t fnptr(parser_t *P, int *depth);
+static int is_fnptr(parser_t *P);
 
 static int64_t parse_int_text(const char *s, int len)
 {
@@ -894,7 +896,12 @@ static uint32_t parse_param_list(parser_t *P)
             P->nodes[param].d.oper.flags = ptr_depth;
         }
 
-        if (cur_type(P) == TOK_IDENT) {
+        if (is_fnptr(P)) {
+            int d = P->nodes[param].d.oper.flags;
+            uint32_t fname = fnptr(P, &d);
+            P->nodes[param].d.oper.flags = d;
+            if (fname) add_child(P, param, fname);
+        } else if (cur_type(P) == TOK_IDENT) {
             uint32_t name = alloc_node(P, AST_IDENT);
             P->nodes[name].d.text.offset = cur(P)->offset;
             P->nodes[name].d.text.len = cur(P)->len;
@@ -922,6 +929,57 @@ static uint32_t parse_param_list(parser_t *P)
     return first;
 }
 
+/* Steps over a balanced (...) without building nodes for it. */
+static void skippar(parser_t *P)
+{
+    int depth = 0;
+    while (cur_type(P) != TOK_EOF) {
+        int t = cur_type(P);
+        if (t == TOK_LPAREN) depth++;
+        else if (t == TOK_RPAREN && --depth == 0) { advance(P); return; }
+        advance(P);
+    }
+}
+
+/* T (*name)(sig) — inner stars add to *depth, and the signature is skipped
+   because Booth has no indirect calls to lower it to. 0 if abstract. */
+static uint32_t fnptr(parser_t *P, int *depth)
+{
+    uint32_t name = 0;
+    advance(P);                                     /* ( */
+    while (cur_type(P) == TOK_STAR) { advance(P); (*depth)++; }
+    while (cur_type(P) == TOK_CONST || cur_type(P) == TOK_CU_RESTRICT)
+        advance(P);
+    if (cur_type(P) == TOK_IDENT) {
+        name = alloc_node(P, AST_IDENT);
+        P->nodes[name].d.text.offset = cur(P)->offset;
+        P->nodes[name].d.text.len = cur(P)->len;
+        advance(P);
+    }
+    expect(P, TOK_RPAREN);
+    if (cur_type(P) == TOK_LPAREN) skippar(P);
+    return name;
+}
+
+/* Distinguishes T (*f)(...) from a parenthesised expression. */
+static int is_fnptr(parser_t *P)
+{
+    return cur_type(P) == TOK_LPAREN && peek_type(P, 1) == TOK_STAR;
+}
+
+/* S() or ~S() directly inside struct S. Both lack a return type, so the
+   type spec would otherwise eat the name and leave a stray (. */
+static int is_ctor(parser_t *P)
+{
+    if (!P->cs_len) return 0;
+    if (cur_type(P) == TOK_TILDE)
+        return peek_type(P, 1) == TOK_IDENT && peek_type(P, 2) == TOK_LPAREN;
+    if (cur_type(P) != TOK_IDENT || peek_type(P, 1) != TOK_LPAREN) return 0;
+    return cur(P)->len == P->cs_len
+        && memcmp(P->src + cur(P)->offset,
+                  P->src + P->cs_off, P->cs_len) == 0;
+}
+
 static int starts_declaration(parser_t *P)
 {
     int t = cur_type(P);
@@ -1033,7 +1091,13 @@ static uint32_t parse_declaration(parser_t *P)
         return u;
     }
 
-    uint32_t type_node = parse_type_spec(P, &quals, &cuda);
+    uint32_t type_node;
+    if (is_ctor(P)) {
+        type_node = alloc_node(P, AST_TYPE_SPEC);
+        P->nodes[type_node].d.btype.kind = TYPE_VOID;
+    } else {
+        type_node = parse_type_spec(P, &quals, &cuda);
+    }
     if (!type_node) {
         parse_error(P, BC_E024);
         advance(P);
@@ -1074,6 +1138,17 @@ static uint32_t parse_declaration(parser_t *P)
                 if (!match(P, TOK_COMMA)) break;
             }
         } else {
+            /* Saved and restored so nested structs each get their own.
+             * Synthetic names live outside src, so memcmp must skip those. */
+            uint32_t sv_off = P->cs_off;
+            uint16_t sv_len = P->cs_len;
+            uint32_t nm = P->nodes[type_node].first_child;
+            P->cs_len = 0;
+            if (nm && P->nodes[nm].type == AST_IDENT &&
+                P->nodes[nm].d.text.offset < BC_ANON_BASE) {
+                P->cs_off = P->nodes[nm].d.text.offset;
+                P->cs_len = (uint16_t)P->nodes[nm].d.text.len;
+            }
             while (cur_type(P) != TOK_RBRACE && cur_type(P) != TOK_EOF) {
                 if (cur_type(P) == TOK_PUBLIC || cur_type(P) == TOK_PRIVATE ||
                     cur_type(P) == TOK_PROTECTED) {
@@ -1083,6 +1158,8 @@ static uint32_t parse_declaration(parser_t *P)
                 uint32_t member = parse_declaration(P);
                 if (member) add_child(P, def, member);
             }
+            P->cs_off = sv_off;
+            P->cs_len = sv_len;
         }
         expect(P, TOK_RBRACE);
 
@@ -1163,6 +1240,25 @@ static uint32_t parse_declaration(parser_t *P)
         }
     }
 
+    if (is_fnptr(P)) {
+        uint32_t fname = fnptr(P, &ptr_depth);
+        if (fname) {
+            decl_node = alloc_node(P, AST_VAR_DECL);
+            P->nodes[decl_node].qualifiers = quals;
+            P->nodes[decl_node].cuda_flags = cuda;
+            P->nodes[decl_node].d.oper.flags = ptr_depth;
+            add_child(P, decl_node, type_node);
+            add_child(P, decl_node, fname);
+            if (quals & QUAL_TYPEDEF)
+                reg_tname(P, P->nodes[fname].d.text.offset,
+                          (uint16_t)P->nodes[fname].d.text.len);
+            if (!expect(P, TOK_SEMI)) sync_past_semi(P);
+            return decl_node;
+        }
+        match(P, TOK_SEMI);
+        return type_node;
+    }
+
     if (cur_type(P) == TOK_IDENT || cur_type(P) == TOK_TILDE
         || cur_type(P) == TOK_OPERATOR) {
         int is_dtor = (cur_type(P) == TOK_TILDE);

src/fe/parser.h

@@ -32,6 +32,11 @@ typedef struct {
     uint32_t        anon_len;
     uint32_t        anon_cnt;
 
+    /* Enclosing struct name, so a constructor can be told apart from a
+     * declaration that happens to start with a type name. len 0 = not in one. */
+    uint32_t        cs_off;
+    uint16_t        cs_len;
+
     bc_error_t      errors[BC_MAX_ERRORS];
     int             num_errors;
 } parser_t;

tests/ctor.cu

@@ -0,0 +1,20 @@
+/* Constructors and destructors, which carry no return type. */
+
+struct vec3 {
+    float x, y, z;
+    vec3() { x = 0.0f; y = 0.0f; z = 0.0f; }
+    vec3(float v) { x = v; y = v; z = v; }
+    ~vec3() { }
+    float sum() { return x + y + z; }
+};
+
+struct outer {
+    struct inner { inner() { } };
+    int n;
+    outer() { n = 0; }
+};
+
+__global__ void k(float *out)
+{
+    out[0] = 1.0f;
+}

tests/fnptr.cu

@@ -0,0 +1,18 @@
+/* Function pointer declarators in each position they can turn up. */
+
+typedef void (*cb_t)(int);
+typedef int (*cmp_t)(int, int);
+typedef void (*void_t)();
+
+struct disp {
+    void (*hook)(float);
+    int n;
+};
+
+cb_t g_cb;
+cmp_t g_cmp;
+
+__global__ void k(int *x, void (*unused)(int))
+{
+    x[0] = 1;
+}

tests/tphase.c

@@ -45,6 +45,34 @@ static void pha_ast(void)
 }
 TH_REG("phase", pha_ast)
 
+/* A function pointer declarator used to parse as a call expression, which
+   cost nothing at parse time and everything later. */
+static void pha_fnptr(void)
+{
+    int rc = th_run(BC_BIN " --parse tests/fnptr.cu", obuf, TH_BUFSZ);
+    CHEQ(rc, 0);
+    CHECK(strstr(obuf, "0 parse error(s)") != NULL);
+    CHECK(strstr(obuf, "(ident cb_t)") != NULL);
+    CHECK(strstr(obuf, "(ident hook)") != NULL);
+    CHECK(strstr(obuf, "call") == NULL);
+    PASS();
+}
+TH_REG("phase", pha_fnptr)
+
+/* Nested structs matter here: the enclosing name is saved and restored,
+   so inner's constructor must not be mistaken for outer's. */
+static void pha_ctor(void)
+{
+    int rc = th_run(BC_BIN " --parse tests/ctor.cu", obuf, TH_BUFSZ);
+    CHEQ(rc, 0);
+    CHECK(strstr(obuf, "0 parse error(s)") != NULL);
+    CHECK(strstr(obuf, "(ident vec3)") != NULL);
+    CHECK(strstr(obuf, "(ident inner)") != NULL);
+    CHECK(strstr(obuf, "(ident sum)") != NULL);
+    PASS();
+}
+TH_REG("phase", pha_ctor)
+
 /* ---- phase: IR ---- */
 
 static void pha_ir(void)